diff --git a/sycl/include/sycl/kernel_bundle.hpp b/sycl/include/sycl/kernel_bundle.hpp index 6eed7b270edca..12651407a537f 100644 --- a/sycl/include/sycl/kernel_bundle.hpp +++ b/sycl/include/sycl/kernel_bundle.hpp @@ -308,17 +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; + 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 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.