From 886a68a193a1a177b55ed8c78d70e3cbcfb6ddb7 Mon Sep 17 00:00:00 2001 From: Andrei Elovikov Date: Mon, 21 Jul 2025 07:54:03 -0700 Subject: [PATCH 1/2] [NFC][SYCL] Better "managed" `ur_program_handle_t` There was `ProgramManager::ProgramPtr` alias over `std::unique_ptr` with a custom deleter to RAII-manage `ur_program_handle_t` lifetime but it was applied in just a few places with the rest of the usage left with C-style explicit management. This PR introduce a dedicated helper class to manage all UR handle types that I think is more convenient than `ProgramManager::ProgramPtr`. I'm also switching all the objects that stored `ur_program_handle_t` and then `urProgramRelease`d them to use that new helper, while leaving the full refactoring (i.e., create those `Managed` objects at `urProgramCreate*`/`urProgramRetain` point) to a subsequent PRs to ease review process. Other `ur*_handle_t`s are left to subsequent changes as well. --- sycl/source/backend.cpp | 2 +- sycl/source/detail/adapter_impl.hpp | 50 ++++++++++ sycl/source/detail/context_impl.cpp | 5 +- sycl/source/detail/context_impl.hpp | 2 +- sycl/source/detail/device_image_impl.cpp | 4 +- sycl/source/detail/device_image_impl.hpp | 46 ++++----- sycl/source/detail/kernel_bundle_impl.hpp | 4 +- .../detail/kernel_name_based_cache_t.hpp | 25 ++--- sycl/source/detail/kernel_program_cache.cpp | 2 +- sycl/source/detail/kernel_program_cache.hpp | 34 +++---- sycl/source/detail/memory_manager.cpp | 2 +- .../program_manager/program_manager.cpp | 95 +++++++++---------- .../program_manager/program_manager.hpp | 15 ++- sycl/source/detail/scheduler/commands.cpp | 4 +- .../arg_mask/EliminatedArgMask.cpp | 2 +- 15 files changed, 157 insertions(+), 135 deletions(-) diff --git a/sycl/source/backend.cpp b/sycl/source/backend.cpp index ff1258b7c2dbf..eaa558ef8bc35 100644 --- a/sycl/source/backend.cpp +++ b/sycl/source/backend.cpp @@ -343,7 +343,7 @@ kernel make_kernel(const context &TargetContext, const device_image &DeviceImage = *KernelBundle.begin(); device_image_impl &DeviceImageImpl = *getSyclObjImpl(DeviceImage); - UrProgram = DeviceImageImpl.get_ur_program_ref(); + UrProgram = DeviceImageImpl.get_ur_program(); } // Create UR kernel first. diff --git a/sycl/source/detail/adapter_impl.hpp b/sycl/source/detail/adapter_impl.hpp index ce4fd8fcf8017..639eff7ab702d 100644 --- a/sycl/source/detail/adapter_impl.hpp +++ b/sycl/source/detail/adapter_impl.hpp @@ -239,6 +239,56 @@ class adapter_impl { UrFuncPtrMapT UrFuncPtrs; }; // class adapter_impl +template class Managed { + static constexpr auto Release = []() constexpr { + if constexpr (std::is_same_v) + return UrApiKind::urProgramRelease; + }(); + +public: + Managed() = default; + Managed(URResource R, adapter_impl &Adapter) : R(R), Adapter(&Adapter) {} + Managed(adapter_impl &Adapter) : Adapter(&Adapter) {} + Managed(const Managed &) = delete; + Managed(Managed &&Other) : Adapter(Other.Adapter) { + R = Other.R; + Other.R = nullptr; + } + Managed &operator=(const Managed &) = delete; + Managed &operator=(Managed &&Other) { + if (R) + Adapter->call(R); + R = Other.R; + Other.R = nullptr; + Adapter = Other.Adapter; + return *this; + } + + operator URResource() const { return R; } + + URResource release() { + URResource Res = R; + R = nullptr; + return Res; + } + + URResource *operator&() { + assert(!R && "Already initialized!"); + assert(Adapter && "Adapter must be set for this API!"); + return &R; + } + + ~Managed() { + if (!R) + return; + + Adapter->call(R); + } + +private: + URResource R = nullptr; + adapter_impl *Adapter = nullptr; +}; } // namespace detail } // namespace _V1 } // namespace sycl diff --git a/sycl/source/detail/context_impl.cpp b/sycl/source/detail/context_impl.cpp index 855c1f8c27295..8800f854f384e 100644 --- a/sycl/source/detail/context_impl.cpp +++ b/sycl/source/detail/context_impl.cpp @@ -128,10 +128,7 @@ context_impl::~context_impl() { DeviceGlobal); DGEntry->removeAssociatedResources(this); } - for (auto LibProg : MCachedLibPrograms) { - assert(LibProg.second && "Null program must not be kept in the cache"); - getAdapter().call(LibProg.second); - } + MCachedLibPrograms.clear(); // TODO catch an exception and put it to list of asynchronous exceptions getAdapter().call_nocheck(MContext); } catch (std::exception &e) { diff --git a/sycl/source/detail/context_impl.hpp b/sycl/source/detail/context_impl.hpp index 2441da11bd210..6d97f1c9ca47e 100644 --- a/sycl/source/detail/context_impl.hpp +++ b/sycl/source/detail/context_impl.hpp @@ -134,7 +134,7 @@ class context_impl : public std::enable_shared_from_this { using CachedLibProgramsT = std::map, - ur_program_handle_t>; + Managed>; /// In contrast to user programs, which are compiled from user code, library /// programs come from the SYCL runtime. They are identified by the diff --git a/sycl/source/detail/device_image_impl.cpp b/sycl/source/detail/device_image_impl.cpp index be62bbf3bbf9b..20b7465f30851 100644 --- a/sycl/source/detail/device_image_impl.cpp +++ b/sycl/source/detail/device_image_impl.cpp @@ -30,7 +30,7 @@ std::shared_ptr device_image_impl::tryGetExtensionKernel( if (!KID || !has_kernel(*KID)) continue; - auto UrProgram = get_ur_program_ref(); + auto UrProgram = get_ur_program(); auto [UrKernel, CacheMutex, ArgMask] = PM.getOrCreateKernel(Context, AdjustedName, /*PropList=*/{}, UrProgram); @@ -41,7 +41,7 @@ std::shared_ptr device_image_impl::tryGetExtensionKernel( return nullptr; } - ur_program_handle_t UrProgram = get_ur_program_ref(); + ur_program_handle_t UrProgram = get_ur_program(); detail::adapter_impl &Adapter = getSyclObjImpl(Context)->getAdapter(); ur_kernel_handle_t UrKernel = nullptr; Adapter.call(UrProgram, AdjustedName.c_str(), diff --git a/sycl/source/detail/device_image_impl.hpp b/sycl/source/detail/device_image_impl.hpp index 0da6ea9bf9216..9955af9e71aee 100644 --- a/sycl/source/detail/device_image_impl.hpp +++ b/sycl/source/detail/device_image_impl.hpp @@ -260,7 +260,8 @@ class device_image_impl ur_program_handle_t Program, uint8_t Origins, private_tag) : MBinImage(BinImage), MContext(std::move(Context)), MDevices(Devices.to>()), MState(State), - MProgram(Program), MKernelIDs(std::move(KernelIDs)), + MProgram(Program, getSyclObjImpl(MContext)->getAdapter()), + MKernelIDs(std::move(KernelIDs)), MSpecConstsDefValBlob(getSpecConstsDefValBlob()), MOrigins(Origins) { updateSpecConstSymMap(); if (BinImage && (MOrigins & ImageOriginSYCLBIN)) { @@ -294,8 +295,8 @@ class device_image_impl std::unique_ptr &&MergedImageStorage, private_tag) : MBinImage(BinImage), MContext(std::move(Context)), MDevices(Devices.to>()), MState(State), - MProgram(Program), MKernelIDs(std::move(KernelIDs)), - MKernelNames{std::move(KernelNames)}, + MProgram(Program, getSyclObjImpl(MContext)->getAdapter()), + MKernelIDs(std::move(KernelIDs)), MKernelNames{std::move(KernelNames)}, MEliminatedKernelArgMasks{std::move(EliminatedKernelArgMasks)}, MSpecConstsBlob(SpecConstsBlob), MSpecConstsDefValBlob(getSpecConstsDefValBlob()), @@ -311,7 +312,8 @@ class device_image_impl private_tag) : MBinImage(BinImage), MContext(std::move(Context)), MDevices(Devices.to>()), MState(State), - MProgram(Program), MKernelNames{std::move(KernelNames)}, + MProgram(Program, getSyclObjImpl(MContext)->getAdapter()), + MKernelNames{std::move(KernelNames)}, MEliminatedKernelArgMasks{std::move(EliminatedKernelArgMasks)}, MSpecConstsDefValBlob(getSpecConstsDefValBlob()), MOrigins(ImageOriginKernelCompiler), @@ -329,8 +331,7 @@ class device_image_impl private_tag) : MBinImage(BinImage), MContext(std::move(Context)), MDevices(Devices.to>()), MState(State), - MProgram(nullptr), MKernelIDs(std::move(KernelIDs)), - MKernelNames{std::move(KernelNames)}, + MKernelIDs(std::move(KernelIDs)), MKernelNames{std::move(KernelNames)}, MSpecConstsDefValBlob(getSpecConstsDefValBlob()), MOrigins(ImageOriginKernelCompiler), MRTCBinInfo(KernelCompilerBinaryInfo{ @@ -344,7 +345,7 @@ class device_image_impl include_pairs_t &&IncludePairsVec, private_tag) : MBinImage(Src), MContext(std::move(Context)), MDevices(Devices.to>()), - MState(bundle_state::ext_oneapi_source), MProgram(nullptr), + MState(bundle_state::ext_oneapi_source), MSpecConstsDefValBlob(getSpecConstsDefValBlob()), MOrigins(ImageOriginKernelCompiler), MRTCBinInfo( @@ -357,7 +358,7 @@ class device_image_impl private_tag) : MBinImage(Bytes), MContext(std::move(Context)), MDevices(Devices.to>()), - MState(bundle_state::ext_oneapi_source), MProgram(nullptr), + MState(bundle_state::ext_oneapi_source), MSpecConstsDefValBlob(getSpecConstsDefValBlob()), MOrigins(ImageOriginKernelCompiler), MRTCBinInfo(KernelCompilerBinaryInfo{Lang}) { @@ -371,7 +372,8 @@ class device_image_impl : MBinImage(static_cast(nullptr)), MContext(std::move(Context)), MDevices(Devices.to>()), MState(State), - MProgram(Program), MKernelNames{std::move(KernelNames)}, + MProgram(Program, getSyclObjImpl(MContext)->getAdapter()), + MKernelNames{std::move(KernelNames)}, MSpecConstsDefValBlob(getSpecConstsDefValBlob()), MOrigins(ImageOriginKernelCompiler), MRTCBinInfo(KernelCompilerBinaryInfo{Lang}) {} @@ -558,9 +560,7 @@ class device_image_impl return get_devices().contains(Dev); } - const ur_program_handle_t &get_ur_program_ref() const noexcept { - return MProgram; - } + ur_program_handle_t get_ur_program() const noexcept { return MProgram; } const RTDeviceBinaryImage *const &get_bin_image_ref() const { return std::get(MBinImage); @@ -617,20 +617,14 @@ class device_image_impl return NativeProgram; } - ~device_image_impl() { - try { - if (MProgram) { - adapter_impl &Adapter = getSyclObjImpl(MContext)->getAdapter(); - Adapter.call(MProgram); - } - if (MSpecConstsBuffer) { - std::lock_guard Lock{MSpecConstAccessMtx}; - adapter_impl &Adapter = getSyclObjImpl(MContext)->getAdapter(); - memReleaseHelper(Adapter, MSpecConstsBuffer); - } - } catch (std::exception &e) { - __SYCL_REPORT_EXCEPTION_TO_STREAM("exception in ~device_image_impl", e); + ~device_image_impl() try { + if (MSpecConstsBuffer) { + std::lock_guard Lock{MSpecConstAccessMtx}; + adapter_impl &Adapter = getSyclObjImpl(MContext)->getAdapter(); + memReleaseHelper(Adapter, MSpecConstsBuffer); } + } catch (std::exception &e) { + __SYCL_REPORT_EXCEPTION_TO_STREAM("exception in ~device_image_impl", e); } std::string adjustKernelName(std::string_view Name) const { @@ -1298,7 +1292,7 @@ class device_image_impl std::vector MDevices; bundle_state MState; // Native program handler which this device image represents - ur_program_handle_t MProgram = nullptr; + Managed MProgram; // List of kernel ids available in this image, elements should be sorted // according to LessByNameComp. Shared between images for performance reasons diff --git a/sycl/source/detail/kernel_bundle_impl.hpp b/sycl/source/detail/kernel_bundle_impl.hpp index 4000bafaf96b2..0363fec749fce 100644 --- a/sycl/source/detail/kernel_bundle_impl.hpp +++ b/sycl/source/detail/kernel_bundle_impl.hpp @@ -1014,11 +1014,11 @@ class kernel_bundle_impl auto [Kernel, CacheMutex, ArgMask] = detail::ProgramManager::getInstance().getOrCreateKernel( MContext, KernelID.get_name(), /*PropList=*/{}, - SelectedImage->get_ur_program_ref()); + SelectedImage->get_ur_program()); return std::make_shared( Kernel, *detail::getSyclObjImpl(MContext), SelectedImage, *this, - ArgMask, SelectedImage->get_ur_program_ref(), CacheMutex); + ArgMask, SelectedImage->get_ur_program(), CacheMutex); } std::shared_ptr diff --git a/sycl/source/detail/kernel_name_based_cache_t.hpp b/sycl/source/detail/kernel_name_based_cache_t.hpp index 3c2b3fb03cf6f..ef92112dcf98e 100644 --- a/sycl/source/detail/kernel_name_based_cache_t.hpp +++ b/sycl/source/detail/kernel_name_based_cache_t.hpp @@ -27,31 +27,26 @@ struct FastKernelCacheVal { caching is disabled, the pointer is nullptr. */ const KernelArgMask *MKernelArgMask; /* Eliminated kernel argument mask. */ - ur_program_handle_t MProgramHandle; /* UR program handle corresponding to - this kernel. */ - const adapter_impl &MAdapterPtr; /* We can keep reference to the adapter - because during 2-stage shutdown the kernel - cache is destroyed deliberately before the - adapter. */ + Managed MProgramHandle; /* UR program handle + corresponding to this kernel. */ + adapter_impl &MAdapter; /* We can keep reference to the adapter + because during 2-stage shutdown the kernel + cache is destroyed deliberately before the + adapter. */ FastKernelCacheVal(ur_kernel_handle_t KernelHandle, std::mutex *Mutex, const KernelArgMask *KernelArgMask, - ur_program_handle_t ProgramHandle, - const adapter_impl &AdapterPtr) + ur_program_handle_t ProgramHandle, adapter_impl &Adapter) : MKernelHandle(KernelHandle), MMutex(Mutex), - MKernelArgMask(KernelArgMask), MProgramHandle(ProgramHandle), - MAdapterPtr(AdapterPtr) {} + MKernelArgMask(KernelArgMask), MProgramHandle(ProgramHandle, Adapter), + MAdapter(Adapter) {} ~FastKernelCacheVal() { if (MKernelHandle) - MAdapterPtr.call(MKernelHandle); - if (MProgramHandle) - MAdapterPtr.call( - MProgramHandle); + MAdapter.call(MKernelHandle); MKernelHandle = nullptr; MMutex = nullptr; MKernelArgMask = nullptr; - MProgramHandle = nullptr; } FastKernelCacheVal(const FastKernelCacheVal &) = delete; diff --git a/sycl/source/detail/kernel_program_cache.cpp b/sycl/source/detail/kernel_program_cache.cpp index 33946dcfa66ab..15bcef7e6a78f 100644 --- a/sycl/source/detail/kernel_program_cache.cpp +++ b/sycl/source/detail/kernel_program_cache.cpp @@ -12,7 +12,7 @@ namespace sycl { inline namespace _V1 { namespace detail { -const adapter_impl &KernelProgramCache::getAdapter() { +adapter_impl &KernelProgramCache::getAdapter() { return MParentContext->getAdapter(); } diff --git a/sycl/source/detail/kernel_program_cache.hpp b/sycl/source/detail/kernel_program_cache.hpp index 4c139f99c8211..41e91ff9f26ac 100644 --- a/sycl/source/detail/kernel_program_cache.hpp +++ b/sycl/source/detail/kernel_program_cache.hpp @@ -111,27 +111,17 @@ class KernelProgramCache { } }; - struct ProgramBuildResult : public BuildResult { - const adapter_impl &MAdapter; - ProgramBuildResult(const adapter_impl &Adapter) : MAdapter(Adapter) { - Val = nullptr; + struct ProgramBuildResult : public BuildResult> { + ProgramBuildResult(adapter_impl &Adapter) { + Val = Managed{Adapter}; } - ProgramBuildResult(const adapter_impl &Adapter, BuildState InitialState) - : MAdapter(Adapter) { - Val = nullptr; + ProgramBuildResult(adapter_impl &Adapter, BuildState InitialState) { + Val = Managed{Adapter}; this->State.store(InitialState); } - ~ProgramBuildResult() { - try { - if (Val) { - ur_result_t Err = - MAdapter.call_nocheck(Val); - __SYCL_CHECK_UR_CODE_NO_EXC(Err, MAdapter.getBackend()); - } - } catch (std::exception &e) { - __SYCL_REPORT_EXCEPTION_TO_STREAM("exception in ~ProgramBuildResult", - e); - } + ~ProgramBuildResult() try { + } catch (std::exception &e) { + __SYCL_REPORT_EXCEPTION_TO_STREAM("exception in ~ProgramBuildResult", e); } }; using ProgramBuildResultPtr = std::shared_ptr; @@ -434,7 +424,7 @@ class KernelProgramCache { if (DidInsert) { It->second = std::make_shared(getAdapter(), BuildState::BS_Done); - It->second->Val = Program; + It->second->Val = Managed{Program, getAdapter()}; // Save reference between the common key and the full key. CommonProgramKeyT CommonKey = std::make_pair(CacheKey.first.second, CacheKey.second); @@ -794,7 +784,9 @@ class KernelProgramCache { // only the building thread will run this try { - BuildResult->Val = Build(); + // Remove `adapter_impl` from `ProgramBuildResult`'s ctors once `Build` + // returns `ManagedVal) = Build(); if constexpr (!std::is_same_v) EvictFunc(BuildResult->Val, /*IsBuilt=*/true); @@ -868,7 +860,7 @@ class KernelProgramCache { friend class ::MockKernelProgramCache; - const adapter_impl &getAdapter(); + adapter_impl &getAdapter(); ur_context_handle_t getURContext() const; }; } // namespace detail diff --git a/sycl/source/detail/memory_manager.cpp b/sycl/source/detail/memory_manager.cpp index 3a5c615364f47..44e553c086175 100644 --- a/sycl/source/detail/memory_manager.cpp +++ b/sycl/source/detail/memory_manager.cpp @@ -1152,7 +1152,7 @@ getOrBuildProgramForDeviceGlobal(queue_impl &Queue, PM.getDeviceImageFromBinaryImage(&Img, Context, Device); device_image_plain BuiltImage = PM.build(std::move(DeviceImage), {std::move(Device)}, {}); - return getSyclObjImpl(BuiltImage)->get_ur_program_ref(); + return getSyclObjImpl(BuiltImage)->get_ur_program(); } static void diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 7f70fa05c2c49..6cde984c034d5 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -920,10 +920,7 @@ ProgramManager::getBuiltURProgram(const BinImgWithDeps &ImgWithDeps, NativePrg, Adapter); } - UrFuncInfo programReleaseInfo; - auto programRelease = - programReleaseInfo.getFuncPtrFromModule(ur::getURLoaderLibrary()); - ProgramPtr ProgramManaged(NativePrg, programRelease); + Managed ProgramManaged(NativePrg, Adapter); // Link a fallback implementation of device libraries if they are not // supported by a device compiler. @@ -938,7 +935,7 @@ ProgramManager::getBuiltURProgram(const BinImgWithDeps &ImgWithDeps, if (UseDeviceLibs) DeviceLibReqMask = getDeviceLibReqMask(MainImg); - std::vector ProgramsToLink; + std::vector> ProgramsToLink; // If we had a program in cache, then it should have been the fully linked // program already. if (!DeviceCodeWasInCache) { @@ -950,8 +947,8 @@ ProgramManager::getBuiltURProgram(const BinImgWithDeps &ImgWithDeps, if (UseDeviceLibs) DeviceLibReqMask |= getDeviceLibReqMask(*BinImg); - ur_program_handle_t NativePrg = - createURProgram(*BinImg, ContextImpl, Devs); + Managed NativePrg{ + createURProgram(*BinImg, ContextImpl, Devs), Adapter}; if (BinImg->supportsSpecConstants()) { enableITTAnnotationsIfNeeded(NativePrg, Adapter); @@ -960,22 +957,22 @@ ProgramManager::getBuiltURProgram(const BinImgWithDeps &ImgWithDeps, *getSyclObjImpl(DevImgWithDeps->getAll()[I]), NativePrg, Adapter); } - ProgramsToLink.push_back(NativePrg); + ProgramsToLink.push_back(std::move(NativePrg)); } } auto URDevices = Devs.to>(); - ProgramPtr BuiltProgram = + Managed BuiltProgram = build(std::move(ProgramManaged), ContextImpl, CompileOpts, LinkOpts, URDevices, DeviceLibReqMask, ProgramsToLink, /*CreatedFromBinary*/ MainImg.getFormat() != SYCL_DEVICE_BINARY_TYPE_SPIRV); - // Those extra programs won't be used anymore, just the final linked result - for (ur_program_handle_t Prg : ProgramsToLink) - Adapter.call(Prg); - emitBuiltProgramInfo(BuiltProgram.get(), ContextImpl); + // Those extra programs won't be used anymore, just the final + // linked result: + ProgramsToLink.clear(); + emitBuiltProgramInfo(BuiltProgram, ContextImpl); { std::lock_guard Lock(MNativeProgramsMutex); @@ -983,20 +980,20 @@ ProgramManager::getBuiltURProgram(const BinImgWithDeps &ImgWithDeps, // so keys in the map can be invalid (reference count went to zero and the // underlying program disposed of). Protecting from incorrect values by // removal of map entries with same handle (obviously invalid entries). - std::ignore = NativePrograms.erase(BuiltProgram.get()); + std::ignore = NativePrograms.erase(BuiltProgram); for (const RTDeviceBinaryImage *Img : ImgWithDeps) { NativePrograms.insert( - {BuiltProgram.get(), {ContextImpl.shared_from_this(), Img}}); + {BuiltProgram, {ContextImpl.shared_from_this(), Img}}); } } - ContextImpl.addDeviceGlobalInitializer(BuiltProgram.get(), Devs, &MainImg); + ContextImpl.addDeviceGlobalInitializer(BuiltProgram, Devs, &MainImg); // Save program to persistent cache if it is not there if (!DeviceCodeWasInCache) { PersistentDeviceCodeCache::putItemToDisc( Devs, ImgWithDeps.getAll(), SpecConsts, CompileOpts + LinkOpts, - BuiltProgram.get()); + BuiltProgram); } return BuiltProgram.release(); @@ -1350,13 +1347,13 @@ loadDeviceLibFallback(context_impl &Context, DeviceLibExt Extension, "At least one device is expected in the input vector"); // Vector of devices that don't have the library cached. for (auto Dev : Devices) { - auto CacheResult = CachedLibPrograms.emplace( - std::make_pair(std::make_pair(Extension, Dev), nullptr)); - auto Cached = !CacheResult.second; + auto CacheResult = CachedLibPrograms.emplace(std::make_pair( + std::make_pair(Extension, Dev), Managed{})); + bool Cached = !CacheResult.second; if (!Cached) { DevicesToCompile.push_back(Dev); } else { - auto CachedURProgram = CacheResult.first->second; + ur_program_handle_t CachedURProgram = CacheResult.first->second; assert(CachedURProgram && "If device lib UR program was cached then is " "expected to be not a nullptr"); assert(((URProgram && URProgram == CachedURProgram) || (!URProgram)) && @@ -1395,7 +1392,8 @@ loadDeviceLibFallback(context_impl &Context, DeviceLibExt Extension, Adapter.call(URProgram); for (auto Dev : DevicesToCompile) - CachedLibPrograms[std::make_pair(Extension, Dev)] = URProgram; + CachedLibPrograms[std::make_pair(Extension, Dev)] = + Managed{URProgram, Adapter}; // TODO no spec constants are used in the std libraries, support in the future // Do not use compile options for library programs: it is not clear if user @@ -1710,15 +1708,16 @@ static inline bool isDeviceImageCompressed(sycl_device_binary Bin) { return currFormat == SYCL_DEVICE_BINARY_TYPE_COMPRESSED_NONE; } -ProgramManager::ProgramPtr ProgramManager::build( - ProgramPtr Program, context_impl &Context, +Managed ProgramManager::build( + Managed Program, context_impl &Context, const std::string &CompileOptions, const std::string &LinkOptions, std::vector &Devices, uint32_t DeviceLibReqMask, - const std::vector &ExtraProgramsToLink, + const std::vector> &ExtraProgramsToLink, bool CreatedFromBinary) { if constexpr (DbgProgMgr > 0) { - std::cerr << ">>> ProgramManager::build(" << Program.get() << ", " + std::cerr << ">>> ProgramManager::build(" + << static_cast(Program) << ", " << CompileOptions << ", " << LinkOptions << ", " << VecToString(Devices) << ", " << std::hex << DeviceLibReqMask << std::dec << ", " << VecToString(ExtraProgramsToLink) << ", " @@ -1748,16 +1747,16 @@ ProgramManager::ProgramPtr ProgramManager::build( ? CompileOptions : (CompileOptions + " " + LinkOptions); ur_result_t Error = Adapter.call_nocheck( - Program.get(), Devices.size(), Devices.data(), Options.c_str()); + Program, Devices.size(), Devices.data(), Options.c_str()); if (Error == UR_RESULT_ERROR_UNSUPPORTED_FEATURE) { Error = Adapter.call_nocheck( - Context.getHandleRef(), Program.get(), Options.c_str()); + Context.getHandleRef(), Program, Options.c_str()); } if (Error != UR_RESULT_SUCCESS) throw detail::set_ur_error( exception(make_error_code(errc::build), - getProgramBuildLog(Program.get(), Context)), + getProgramBuildLog(Program, Context)), Error); return Program; @@ -1765,11 +1764,13 @@ ProgramManager::ProgramPtr ProgramManager::build( // Include the main program and compile/link everything together if (!CreatedFromBinary) { - auto Res = doCompile(Adapter, Program.get(), Devices.size(), Devices.data(), + auto Res = doCompile(Adapter, Program, Devices.size(), Devices.data(), Context.getHandleRef(), CompileOptions.c_str()); Adapter.checkUrResult(Res); } - LinkPrograms.push_back(Program.get()); + // Should be `std::move(Program)` once `LinkPrograms` is switched to + // `Managed LinkedProg{Adapter}; auto doLink = [&] { auto Res = Adapter.call_nocheck( Context.getHandleRef(), Devices.size(), Devices.data(), @@ -1802,12 +1803,12 @@ ProgramManager::ProgramPtr ProgramManager::build( } // Link program call returns a new program object if all parameters are valid, - // or NULL otherwise. Release the original (user) program. - Program.reset(LinkedProg); + // or NULL otherwise. if (Error != UR_RESULT_SUCCESS) { if (LinkedProg) { // A non-trivial error occurred during linkage: get a build log, release - // an incomplete (but valid) LinkedProg, and throw. + // an incomplete (but valid) LinkedProg (via implicit dtor call), and + // throw. throw detail::set_ur_error( exception(make_error_code(errc::build), getProgramBuildLog(LinkedProg, Context)), @@ -1815,7 +1816,7 @@ ProgramManager::ProgramPtr ProgramManager::build( } Adapter.checkUrResult(Error); } - return Program; + return LinkedProg; } void ProgramManager::cacheKernelUsesAssertInfo(const RTDeviceBinaryImage &Img) { @@ -2864,15 +2865,14 @@ ProgramManager::compile(const DevImgPlainWithDeps &ImgWithDeps, CompileOptions, *(InputImpl.get_bin_image_ref()), Devs, Adapter); // Should always come last! appendCompileEnvironmentVariablesThatAppend(CompileOptions); - ur_result_t Error = - doCompile(Adapter, ObjectImpl->get_ur_program_ref(), Devs.size(), - URDevices.data(), - getSyclObjImpl(InputImpl.get_context()).get()->getHandleRef(), - CompileOptions.c_str()); + ur_result_t Error = doCompile( + Adapter, ObjectImpl->get_ur_program(), Devs.size(), URDevices.data(), + getSyclObjImpl(InputImpl.get_context()).get()->getHandleRef(), + CompileOptions.c_str()); if (Error != UR_RESULT_SUCCESS) throw sycl::exception( make_error_code(errc::build), - getProgramBuildLog(ObjectImpl->get_ur_program_ref(), + getProgramBuildLog(ObjectImpl->get_ur_program(), *getSyclObjImpl(ObjectImpl->get_context()))); CompiledImages.push_back( @@ -2956,7 +2956,7 @@ ProgramManager::link(const std::vector &Imgs, std::vector URPrograms; URPrograms.reserve(Imgs.size()); for (const device_image_plain &Img : Imgs) - URPrograms.push_back(getSyclObjImpl(Img)->get_ur_program_ref()); + URPrograms.push_back(getSyclObjImpl(Img)->get_ur_program()); auto URDevices = Devs.to>(); @@ -3255,16 +3255,13 @@ ur_kernel_handle_t ProgramManager::getOrCreateMaterializedKernel( auto Program = createURProgram(Img, ContextImpl, {Device}); detail::device_impl &DeviceImpl = *detail::getSyclObjImpl(Device); adapter_impl &Adapter = DeviceImpl.getAdapter(); - UrFuncInfo programReleaseInfo; - auto programRelease = - programReleaseInfo.getFuncPtrFromModule(ur::getURLoaderLibrary()); - ProgramPtr ProgramManaged(Program, programRelease); + Managed ProgramManaged(Program, Adapter); std::string CompileOpts; std::string LinkOpts; applyOptionsFromEnvironment(CompileOpts, LinkOpts); // No linking of extra programs reqruired. - std::vector ExtraProgramsToLink; + std::vector> ExtraProgramsToLink; std::vector Devs = {DeviceImpl.getHandleRef()}; auto BuildProgram = build(std::move(ProgramManaged), ContextImpl, CompileOpts, LinkOpts, Devs, @@ -3272,7 +3269,7 @@ ur_kernel_handle_t ProgramManager::getOrCreateMaterializedKernel( ExtraProgramsToLink); ur_kernel_handle_t UrKernel{nullptr}; Adapter.call( - BuildProgram.get(), KernelName.data(), &UrKernel); + BuildProgram, KernelName.data(), &UrKernel); { std::lock_guard KernelIDsGuard(m_KernelIDsMutex); m_MaterializedKernels[KernelName][SpecializationConsts] = UrKernel; diff --git a/sycl/source/detail/program_manager/program_manager.hpp b/sycl/source/detail/program_manager/program_manager.hpp index b7b0475457cfa..221abed5af865 100644 --- a/sycl/source/detail/program_manager/program_manager.hpp +++ b/sycl/source/detail/program_manager/program_manager.hpp @@ -393,15 +393,12 @@ class ProgramManager { ProgramManager(ProgramManager const &) = delete; ProgramManager &operator=(ProgramManager const &) = delete; - using ProgramPtr = std::unique_ptr, - decltype(&::urProgramRelease)>; - ProgramPtr build(ProgramPtr Program, context_impl &Context, - const std::string &CompileOptions, - const std::string &LinkOptions, - std::vector &Devices, - uint32_t DeviceLibReqMask, - const std::vector &ProgramsToLink, - bool CreatedFromBinary = false); + Managed + build(Managed Program, context_impl &Context, + const std::string &CompileOptions, const std::string &LinkOptions, + std::vector &Devices, uint32_t DeviceLibReqMask, + const std::vector> &ProgramsToLink, + bool CreatedFromBinary = false); /// Dumps image to current directory void dumpImage(const RTDeviceBinaryImage &Img, uint32_t SequenceID = 0) const; diff --git a/sycl/source/detail/scheduler/commands.cpp b/sycl/source/detail/scheduler/commands.cpp index 4ac11b25c9ff2..e427d851148c1 100644 --- a/sycl/source/detail/scheduler/commands.cpp +++ b/sycl/source/detail/scheduler/commands.cpp @@ -2719,7 +2719,7 @@ void enqueueImpKernel( Kernel = SyclKernelImpl->getHandleRef(); DeviceImageImpl = SyclKernelImpl->getDeviceImage(); - Program = DeviceImageImpl->get_ur_program_ref(); + Program = DeviceImageImpl->get_ur_program(); EliminatedArgMask = SyclKernelImpl->getKernelArgMask(); KernelMutex = SyclKernelImpl->getCacheMutex(); @@ -2806,7 +2806,7 @@ ur_result_t enqueueReadWriteHostPipe(queue_impl &Queue, hostPipeEntry->getDevBinImage(), Queue.get_context(), Device); device_image_plain BuiltImage = ProgramManager::getInstance().build( std::move(devImgPlain), {std::move(Device)}, {}); - Program = getSyclObjImpl(BuiltImage)->get_ur_program_ref(); + Program = getSyclObjImpl(BuiltImage)->get_ur_program(); } assert(Program && "Program for this hostpipe is not compiled."); diff --git a/sycl/unittests/program_manager/arg_mask/EliminatedArgMask.cpp b/sycl/unittests/program_manager/arg_mask/EliminatedArgMask.cpp index 3438f05206db5..f856703f30185 100644 --- a/sycl/unittests/program_manager/arg_mask/EliminatedArgMask.cpp +++ b/sycl/unittests/program_manager/arg_mask/EliminatedArgMask.cpp @@ -187,7 +187,7 @@ const sycl::detail::KernelArgMask *getKernelArgMaskFromBundle( EXPECT_TRUE(SyclKernelImpl != nullptr); std::shared_ptr DeviceImageImpl = SyclKernelImpl->getDeviceImage(); - ur_program_handle_t Program = DeviceImageImpl->get_ur_program_ref(); + ur_program_handle_t Program = DeviceImageImpl->get_ur_program(); EXPECT_TRUE(nullptr == ExecKernel->MSyclKernel || !ExecKernel->MSyclKernel->isCreatedFromSource()); From 7bd44b74b6ea42b5a9ea2566c1a5d20931f37fab Mon Sep 17 00:00:00 2001 From: Andrei Elovikov Date: Mon, 21 Jul 2025 13:56:02 -0700 Subject: [PATCH 2/2] Fix dtor try/catch --- sycl/source/detail/device_image_impl.hpp | 10 ++++++++++ sycl/source/detail/kernel_program_cache.hpp | 10 ++++++++++ 2 files changed, 20 insertions(+) diff --git a/sycl/source/detail/device_image_impl.hpp b/sycl/source/detail/device_image_impl.hpp index 9955af9e71aee..8a6875b51d873 100644 --- a/sycl/source/detail/device_image_impl.hpp +++ b/sycl/source/detail/device_image_impl.hpp @@ -617,6 +617,12 @@ class device_image_impl return NativeProgram; } +#ifdef _MSC_VER +#pragma warning(push) +// https://developercommunity.visualstudio.com/t/False-C4297-warning-while-using-function/1130300 +// https://godbolt.org/z/xsMvKf84f +#pragma warning(disable : 4297) +#endif ~device_image_impl() try { if (MSpecConstsBuffer) { std::lock_guard Lock{MSpecConstAccessMtx}; @@ -625,7 +631,11 @@ class device_image_impl } } catch (std::exception &e) { __SYCL_REPORT_EXCEPTION_TO_STREAM("exception in ~device_image_impl", e); + return; // Don't re-throw. } +#ifdef _MSC_VER +#pragma warning(pop) +#endif std::string adjustKernelName(std::string_view Name) const { if (MOrigins & ImageOriginSYCLBIN) { diff --git a/sycl/source/detail/kernel_program_cache.hpp b/sycl/source/detail/kernel_program_cache.hpp index 41e91ff9f26ac..73884d53385f0 100644 --- a/sycl/source/detail/kernel_program_cache.hpp +++ b/sycl/source/detail/kernel_program_cache.hpp @@ -119,10 +119,20 @@ class KernelProgramCache { Val = Managed{Adapter}; this->State.store(InitialState); } +#ifdef _MSC_VER +#pragma warning(push) +// https://developercommunity.visualstudio.com/t/False-C4297-warning-while-using-function/1130300 +// https://godbolt.org/z/xsMvKf84f +#pragma warning(disable : 4297) +#endif ~ProgramBuildResult() try { } catch (std::exception &e) { __SYCL_REPORT_EXCEPTION_TO_STREAM("exception in ~ProgramBuildResult", e); + return; // Don't re-throw. } +#ifdef _MSC_VER +#pragma warning(pop) +#endif }; using ProgramBuildResultPtr = std::shared_ptr;