diff --git a/sycl/source/backend.cpp b/sycl/source/backend.cpp index 091d36344654f..51e3b7ddc520b 100644 --- a/sycl/source/backend.cpp +++ b/sycl/source/backend.cpp @@ -323,7 +323,7 @@ kernel make_kernel(const context &TargetContext, backend Backend) { const auto &Adapter = getAdapter(Backend); const auto &ContextImpl = getSyclObjImpl(TargetContext); - const auto &KernelBundleImpl = getSyclObjImpl(KernelBundle); + kernel_bundle_impl &KernelBundleImpl = *getSyclObjImpl(KernelBundle); // For Level-Zero expect exactly one device image in the bundle. This is // natural for interop kernel to get created out of a single native @@ -334,7 +334,7 @@ kernel make_kernel(const context &TargetContext, // ur_program_handle_t UrProgram = nullptr; if (Backend == backend::ext_oneapi_level_zero) { - if (KernelBundleImpl->size() != 1) + if (KernelBundleImpl.size() != 1) throw sycl::exception( sycl::make_error_code(sycl::errc::runtime), "make_kernel: kernel_bundle must have single program image " + @@ -360,7 +360,7 @@ kernel make_kernel(const context &TargetContext, // Construct the SYCL queue from UR queue. return detail::createSyclObjFromImpl( - std::make_shared(UrKernel, *ContextImpl, KernelBundleImpl)); + std::make_shared(UrKernel, *ContextImpl, &KernelBundleImpl)); } kernel make_kernel(ur_native_handle_t NativeHandle, diff --git a/sycl/source/detail/device_image_impl.cpp b/sycl/source/detail/device_image_impl.cpp index 655a8dcc72c24..423df27bd489d 100644 --- a/sycl/source/detail/device_image_impl.cpp +++ b/sycl/source/detail/device_image_impl.cpp @@ -33,9 +33,9 @@ std::shared_ptr device_image_impl::tryGetExtensionKernel( auto [UrKernel, CacheMutex, ArgMask] = PM.getOrCreateKernel(Context, AdjustedName, /*PropList=*/{}, UrProgram); - return std::make_shared( - UrKernel, *getSyclObjImpl(Context), shared_from_this(), - OwnerBundle.shared_from_this(), ArgMask, UrProgram, CacheMutex); + return std::make_shared(UrKernel, *getSyclObjImpl(Context), + shared_from_this(), OwnerBundle, + ArgMask, UrProgram, CacheMutex); } return nullptr; } @@ -49,7 +49,7 @@ std::shared_ptr device_image_impl::tryGetExtensionKernel( return std::make_shared( UrKernel, *detail::getSyclObjImpl(Context), shared_from_this(), - OwnerBundle.shared_from_this(), + OwnerBundle, /*ArgMask=*/nullptr, UrProgram, /*CacheMutex=*/nullptr); } diff --git a/sycl/source/detail/handler_impl.hpp b/sycl/source/detail/handler_impl.hpp index 4712a14b49c97..0fda3dd4f2769 100644 --- a/sycl/source/detail/handler_impl.hpp +++ b/sycl/source/detail/handler_impl.hpp @@ -22,8 +22,6 @@ class dynamic_parameter_impl; } // namespace ext::oneapi::experimental::detail namespace detail { -using KernelBundleImplPtr = std::shared_ptr; - enum class HandlerSubmissionState : std::uint8_t { NO_STATE = 0, EXPLICIT_KERNEL_BUNDLE_STATE, diff --git a/sycl/source/detail/kernel_bundle_impl.hpp b/sycl/source/detail/kernel_bundle_impl.hpp index 59b7132b8db23..b873c77ae6df0 100644 --- a/sycl/source/detail/kernel_bundle_impl.hpp +++ b/sycl/source/detail/kernel_bundle_impl.hpp @@ -412,10 +412,9 @@ class kernel_bundle_impl removeDuplicateImages(); for (const kernel_bundle &Bundle : ObjectBundles) { - const KernelBundleImplPtr &BundlePtr = getSyclObjImpl(Bundle); - for (const std::pair> - &SpecConst : BundlePtr->MSpecConstValues) { - MSpecConstValues[SpecConst.first] = SpecConst.second; + kernel_bundle_impl &BundleImpl = *getSyclObjImpl(Bundle); + for (const auto &[Name, Values] : BundleImpl.MSpecConstValues) { + MSpecConstValues[Name] = Values; } } } @@ -567,7 +566,8 @@ class kernel_bundle_impl // SYCLBIN constructor kernel_bundle_impl(const context &Context, const std::vector &Devs, - const sycl::span &Bytes, bundle_state State) + const sycl::span Bytes, bundle_state State, + private_tag) : MContext(Context), MDevices(Devs), MState(State) { common_ctor_checks(); @@ -993,9 +993,8 @@ class kernel_bundle_impl SelectedImage->get_ur_program_ref()); return std::make_shared( - Kernel, *detail::getSyclObjImpl(MContext), SelectedImage, - shared_from_this(), ArgMask, SelectedImage->get_ur_program_ref(), - CacheMutex); + Kernel, *detail::getSyclObjImpl(MContext), SelectedImage, *this, + ArgMask, SelectedImage->get_ur_program_ref(), CacheMutex); } std::shared_ptr diff --git a/sycl/source/detail/kernel_impl.cpp b/sycl/source/detail/kernel_impl.cpp index 2977cffd28fc2..8ef45146fecd8 100644 --- a/sycl/source/detail/kernel_impl.cpp +++ b/sycl/source/detail/kernel_impl.cpp @@ -17,12 +17,14 @@ inline namespace _V1 { namespace detail { kernel_impl::kernel_impl(ur_kernel_handle_t Kernel, context_impl &Context, - KernelBundleImplPtr KernelBundleImpl, + kernel_bundle_impl *KernelBundleImpl, const KernelArgMask *ArgMask) : MKernel(Kernel), MContext(Context.shared_from_this()), MProgram(ProgramManager::getInstance().getUrProgramFromUrKernel(Kernel, Context)), - MCreatedFromSource(true), MKernelBundleImpl(std::move(KernelBundleImpl)), + MCreatedFromSource(true), + MKernelBundleImpl(KernelBundleImpl ? KernelBundleImpl->shared_from_this() + : nullptr), MIsInterop(true), MKernelArgMaskPtr{ArgMask} { ur_context_handle_t UrContext = nullptr; // Using the adapter from the passed ContextImpl @@ -39,14 +41,14 @@ kernel_impl::kernel_impl(ur_kernel_handle_t Kernel, context_impl &Context, kernel_impl::kernel_impl(ur_kernel_handle_t Kernel, context_impl &ContextImpl, DeviceImageImplPtr DeviceImageImpl, - KernelBundleImplPtr &&KernelBundleImpl, + const kernel_bundle_impl &KernelBundleImpl, const KernelArgMask *ArgMask, ur_program_handle_t Program, std::mutex *CacheMutex) : MKernel(Kernel), MContext(ContextImpl.shared_from_this()), MProgram(Program), MCreatedFromSource(DeviceImageImpl->isNonSYCLSourceBased()), MDeviceImageImpl(std::move(DeviceImageImpl)), - MKernelBundleImpl(std::move(KernelBundleImpl)), + MKernelBundleImpl(KernelBundleImpl.shared_from_this()), MIsInterop(MDeviceImageImpl->getOriginMask() & ImageOriginInterop), MKernelArgMaskPtr{ArgMask}, MCacheMutex{CacheMutex} { // Enable USM indirect access for interop and non-sycl-jit source kernels. diff --git a/sycl/source/detail/kernel_impl.hpp b/sycl/source/detail/kernel_impl.hpp index 3ac0da5977941..5a57f1b14fde4 100644 --- a/sycl/source/detail/kernel_impl.hpp +++ b/sycl/source/detail/kernel_impl.hpp @@ -40,7 +40,7 @@ class kernel_impl { /// \param Context is a valid SYCL context /// \param KernelBundleImpl is a valid instance of kernel_bundle_impl kernel_impl(ur_kernel_handle_t Kernel, context_impl &Context, - KernelBundleImplPtr KernelBundleImpl, + kernel_bundle_impl *KernelBundleImpl, const KernelArgMask *ArgMask = nullptr); /// Constructs a SYCL kernel_impl instance from a SYCL device_image, @@ -51,7 +51,7 @@ class kernel_impl { /// \param KernelBundleImpl is a valid instance of kernel_bundle_impl kernel_impl(ur_kernel_handle_t Kernel, context_impl &ContextImpl, DeviceImageImplPtr DeviceImageImpl, - KernelBundleImplPtr &&KernelBundleImpl, + const kernel_bundle_impl &KernelBundleImpl, const KernelArgMask *ArgMask, ur_program_handle_t Program, std::mutex *CacheMutex); diff --git a/sycl/source/detail/scheduler/commands.cpp b/sycl/source/detail/scheduler/commands.cpp index 2d9b651077ab8..12de67aac105f 100644 --- a/sycl/source/detail/scheduler/commands.cpp +++ b/sycl/source/detail/scheduler/commands.cpp @@ -2540,7 +2540,7 @@ getCGKernelInfo(const CGExecKernel &CommandGroup, context_impl &ContextImpl, ur_kernel_handle_t UrKernel = nullptr; std::shared_ptr DeviceImageImpl = nullptr; const KernelArgMask *EliminatedArgMask = nullptr; - auto &KernelBundleImplPtr = CommandGroup.MKernelBundle; + kernel_bundle_impl *KernelBundleImplPtr = CommandGroup.MKernelBundle.get(); if (auto Kernel = CommandGroup.MSyclKernel; Kernel != nullptr) { UrKernel = Kernel->getHandleRef(); diff --git a/sycl/source/kernel_bundle.cpp b/sycl/source/kernel_bundle.cpp index d10078622c941..a7e0f9289aac0 100644 --- a/sycl/source/kernel_bundle.cpp +++ b/sycl/source/kernel_bundle.cpp @@ -213,7 +213,7 @@ get_kernel_bundle_impl(const context &Ctx, const std::vector &Devs, detail::KernelBundleImplPtr get_kernel_bundle_impl(const context &Ctx, const std::vector &Devs, const sycl::span &Bytes, bundle_state State) { - return std::make_shared(Ctx, Devs, Bytes, State); + return detail::kernel_bundle_impl::create(Ctx, Devs, Bytes, State); } detail::KernelBundleImplPtr @@ -524,8 +524,8 @@ obj_kb compile_from_source( LogPtr = &Log; std::vector UniqueDevices = sycl::detail::removeDuplicateDevices(Devices); - std::shared_ptr sourceImpl = getSyclObjImpl(SourceKB); - std::shared_ptr KBImpl = sourceImpl->compile_from_source( + kernel_bundle_impl &sourceImpl = *getSyclObjImpl(SourceKB); + std::shared_ptr KBImpl = sourceImpl.compile_from_source( UniqueDevices, BuildOptions, LogPtr, RegisteredKernelNames); auto result = sycl::detail::createSyclObjFromImpl(KBImpl); if (LogView) @@ -548,9 +548,8 @@ exe_kb build_from_source( LogPtr = &Log; std::vector UniqueDevices = sycl::detail::removeDuplicateDevices(Devices); - const std::shared_ptr &sourceImpl = - getSyclObjImpl(SourceKB); - std::shared_ptr KBImpl = sourceImpl->build_from_source( + kernel_bundle_impl &sourceImpl = *getSyclObjImpl(SourceKB); + std::shared_ptr KBImpl = sourceImpl.build_from_source( UniqueDevices, BuildOptions, LogPtr, RegisteredKernelNames); auto result = sycl::detail::createSyclObjFromImpl(std::move(KBImpl)); if (LogView)