From 3cdd0ca33f2774e8fa29557042b1a7fc0a17e057 Mon Sep 17 00:00:00 2001 From: "Larsen, Steffen" Date: Mon, 6 Mar 2023 04:16:16 -0800 Subject: [PATCH 1/3] [SYCL] Fix get_specialization_constant segmentation fault In certain cases get_specialization_constant would cause a segmentation fault, likely due to strict aliasing violations. This commit changes the implementation to use memcpy of the data into the resulting object, as this can be assumed to be valid due to specialization constants being device-copyable. Signed-off-by: Larsen, Steffen --- sycl/include/sycl/kernel_bundle.hpp | 13 +++++++------ 1 file changed, 7 insertions(+), 6 deletions(-) diff --git a/sycl/include/sycl/kernel_bundle.hpp b/sycl/include/sycl/kernel_bundle.hpp index 8f71c118027a7..12651407a537f 100644 --- a/sycl/include/sycl/kernel_bundle.hpp +++ b/sycl/include/sycl/kernel_bundle.hpp @@ -308,18 +308,19 @@ class kernel_bundle : public detail::kernel_bundle_plain, template typename std::remove_reference_t::value_type get_specialization_constant() const { - const char *SpecSymName = detail::get_spec_constant_symbolic_ID(); - if (!is_specialization_constant_set(SpecSymName)) - return SpecName.getDefaultValue(); - using SCType = typename std::remove_reference_t::value_type; - std::array RetValue; + const char *SpecSymName = detail::get_spec_constant_symbolic_ID(); + SCType Res{SpecName.getDefaultValue()}; + if (!is_specialization_constant_set(SpecSymName)) + return Res; + std::array RetValue; get_specialization_constant_impl(SpecSymName, RetValue.data()); + std::memcpy(&Res, RetValue.data(), sizeof(SCType)); - return *reinterpret_cast(RetValue.data()); + return Res; } /// \returns an iterator to the first device image kernel_bundle contains From e622cc64dd688dc24a7082e388add22f29dc822b Mon Sep 17 00:00:00 2001 From: "Larsen, Steffen" Date: Mon, 6 Mar 2023 06:15:55 -0800 Subject: [PATCH 2/3] Fix cache test Signed-off-by: Larsen, Steffen --- sycl/unittests/kernel-and-program/Cache.cpp | 6 +++++- 1 file changed, 5 insertions(+), 1 deletion(-) diff --git a/sycl/unittests/kernel-and-program/Cache.cpp b/sycl/unittests/kernel-and-program/Cache.cpp index a4cc58ea360ae..0002aa75e332f 100644 --- a/sycl/unittests/kernel-and-program/Cache.cpp +++ b/sycl/unittests/kernel-and-program/Cache.cpp @@ -68,7 +68,11 @@ template <> const char *get_spec_constant_symbolic_ID() { static sycl::unittest::PiImage generateDefaultImage() { using namespace sycl::unittest; + std::vector SpecConstData; + PiProperty SC1 = makeSpecConstant(SpecConstData, "SC1", {0}, {0}, {42}); + PiPropertySet PropSet; + addSpecConstants({SC1}, std::move(SpecConstData), PropSet); std::vector Bin{0, 1, 2, 3, 4, 5}; // Random data @@ -256,7 +260,7 @@ TEST_F(KernelAndProgramCacheTest, SpecConstantCacheNegative) { detail::KernelProgramCache::ProgramCache &Cache = CtxImpl->getKernelProgramCache().acquireCachedPrograms().get(); - EXPECT_EQ(Cache.size(), 1U) << "Expect non-empty cache"; + EXPECT_EQ(Cache.size(), 2U) << "Expect an entry for each build in the cache."; } // Check that kernel_bundle created through join() is not cached. From 1f385fa8c17d383f2b845a032ee0be61668b29e8 Mon Sep 17 00:00:00 2001 From: "Larsen, Steffen" Date: Mon, 6 Mar 2023 06:22:58 -0800 Subject: [PATCH 3/3] Revert buffer size change Signed-off-by: Larsen, Steffen --- sycl/include/sycl/kernel_bundle.hpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/include/sycl/kernel_bundle.hpp b/sycl/include/sycl/kernel_bundle.hpp index 12651407a537f..d644affc6c435 100644 --- a/sycl/include/sycl/kernel_bundle.hpp +++ b/sycl/include/sycl/kernel_bundle.hpp @@ -316,7 +316,7 @@ class kernel_bundle : public detail::kernel_bundle_plain, if (!is_specialization_constant_set(SpecSymName)) return Res; - std::array RetValue; + std::array RetValue; get_specialization_constant_impl(SpecSymName, RetValue.data()); std::memcpy(&Res, RetValue.data(), sizeof(SCType));