[llvm] [libsycl] add queue ctors with context (PR #221763)
Kseniya Tikhomirova via llvm-commits
llvm-commits at lists.llvm.org
Tue Sep 8 06:12:14 PDT 2026
https://github.com/KseniyaTikhomirova updated https://github.com/llvm/llvm-project/pull/221763
>From e05f0feb1b76fdda862c96f0819a101f0a096f3c Mon Sep 17 00:00:00 2001
From: "Tikhomirova, Kseniya" <kseniya.tikhomirova at intel.com>
Date: Fri, 4 Sep 2026 09:26:48 -0700
Subject: [PATCH] [libsycl] add queue ctors with queue
Signed-off-by: Tikhomirova, Kseniya <kseniya.tikhomirova at intel.com>
draft 2
Signed-off-by: Tikhomirova, Kseniya <kseniya.tikhomirova at intel.com>
---
libsycl/include/sycl/__impl/queue.hpp | 67 ++++++
libsycl/src/detail/context_impl.cpp | 46 ++++
libsycl/src/detail/context_impl.hpp | 38 ++++
libsycl/src/detail/device_image_wrapper.cpp | 20 +-
libsycl/src/detail/device_image_wrapper.hpp | 40 ++--
libsycl/src/detail/device_kernel_info.hpp | 32 +--
libsycl/src/detail/program_manager.cpp | 65 ++++--
libsycl/src/detail/program_manager.hpp | 24 +-
libsycl/src/detail/queue_impl.cpp | 20 +-
libsycl/src/detail/queue_impl.hpp | 7 +-
libsycl/src/queue.cpp | 7 +-
libsycl/unittests/mock/helpers.cpp | 2 +-
.../unittests/program_manager/CMakeLists.txt | 3 +-
.../program_manager/program_cache.cpp | 208 ++++++++++++++++++
libsycl/unittests/queue/queue.cpp | 24 ++
15 files changed, 515 insertions(+), 88 deletions(-)
create mode 100644 libsycl/unittests/program_manager/program_cache.cpp
diff --git a/libsycl/include/sycl/__impl/queue.hpp b/libsycl/include/sycl/__impl/queue.hpp
index bd6613e95324b..3b81477047863 100644
--- a/libsycl/include/sycl/__impl/queue.hpp
+++ b/libsycl/include/sycl/__impl/queue.hpp
@@ -18,6 +18,7 @@
#include <sycl/__impl/async_handler.hpp>
#include <sycl/__impl/device.hpp>
#include <sycl/__impl/event.hpp>
+#include <sycl/__impl/platform.hpp>
#include <sycl/__impl/property_list.hpp>
#include <sycl/__impl/detail/config.hpp>
@@ -134,6 +135,72 @@ class _LIBSYCL_EXPORT queue {
/// \param asyncHandler is a SYCL asynchronous exception handler.
/// \param propList is a list of properties for queue construction.
explicit queue(const device &syclDevice, const async_handler &asyncHandler,
+ const property_list &propList = {})
+ : queue(syclDevice.get_platform().khr_get_default_context(), syclDevice,
+ asyncHandler, propList) {}
+
+ /// Constructs a SYCL queue instance that is associated with syclContext,
+ /// using the device identified by the device selector provided.
+ ///
+ /// \param syclContext is the context to associate the queue with.
+ /// \param deviceSelector is a SYCL 2020 Device Selector, a simple callable
+ /// that takes a device and returns an int
+ /// \param propList is a list of properties for queue construction.
+ /// \throw sycl::exception with sycl::errc::invalid if syclContext does not
+ /// contain the selected device.
+ template <
+ typename DeviceSelector,
+ typename = detail::EnableIfDeviceSelectorIsInvocable<DeviceSelector>>
+ explicit queue(const context &syclContext,
+ const DeviceSelector &deviceSelector,
+ const property_list &propList = {})
+ : queue(syclContext, detail::SelectDevice(deviceSelector),
+ detail::defaultAsyncHandler, propList) {}
+
+ /// Constructs a SYCL queue instance with an async_handler that is associated
+ /// with syclContext, using the device identified by the device selector
+ /// provided.
+ ///
+ /// \param syclContext is the context to associate the queue with.
+ /// \param deviceSelector is a SYCL 2020 Device Selector, a simple callable
+ /// that takes a device and returns an int
+ /// \param asyncHandler is a SYCL asynchronous exception handler.
+ /// \param propList is a list of properties for queue construction.
+ /// \throw sycl::exception with sycl::errc::invalid if syclContext does not
+ /// contain the selected device.
+ template <
+ typename DeviceSelector,
+ typename = detail::EnableIfDeviceSelectorIsInvocable<DeviceSelector>>
+ explicit queue(const context &syclContext,
+ const DeviceSelector &deviceSelector,
+ const async_handler &asyncHandler,
+ const property_list &propList = {})
+ : queue(syclContext, detail::SelectDevice(deviceSelector), asyncHandler,
+ propList) {}
+
+ /// Constructs a SYCL queue instance that is associated with syclContext,
+ /// using the device provided.
+ ///
+ /// \param syclContext is the context to associate the queue with.
+ /// \param syclDevice is an instance of SYCL device.
+ /// \param propList is a list of properties for queue construction.
+ /// \throw sycl::exception with sycl::errc::invalid if syclContext does not
+ /// contain syclDevice.
+ explicit queue(const context &syclContext, const device &syclDevice,
+ const property_list &propList = {})
+ : queue(syclContext, syclDevice, detail::defaultAsyncHandler, propList) {}
+
+ /// Constructs a SYCL queue instance with an async_handler that is associated
+ /// with syclContext, using the device provided.
+ ///
+ /// \param syclContext is the context to associate the queue with.
+ /// \param syclDevice is an instance of SYCL device.
+ /// \param asyncHandler is a SYCL asynchronous exception handler.
+ /// \param propList is a list of properties for queue construction.
+ /// \throw sycl::exception with sycl::errc::invalid if syclContext does not
+ /// contain syclDevice.
+ explicit queue(const context &syclContext, const device &syclDevice,
+ const async_handler &asyncHandler,
const property_list &propList = {});
/// \return the SYCL backend associated with this queue.
diff --git a/libsycl/src/detail/context_impl.cpp b/libsycl/src/detail/context_impl.cpp
index 66a63131b49bb..d577936660a0a 100644
--- a/libsycl/src/detail/context_impl.cpp
+++ b/libsycl/src/detail/context_impl.cpp
@@ -34,6 +34,11 @@ ContextImpl::ContextImpl(std::vector<DeviceImpl *> &&DeviceList,
ContextImpl::~ContextImpl() {
assert(MOffloadContext && "Context must be created in ctor");
+ // liboffload does not reference-count contexts: every resource tied to a
+ // context must be released before olDestroyContext, otherwise it is left in
+ // an undefined state. MPrograms is a member, so it would be destroyed only
+ // after this destructor body has run.
+ releaseAllPrograms();
std::ignore = olDestroyContext(MOffloadContext);
}
@@ -49,5 +54,46 @@ void ContextImpl::iterateDevices(
backend ContextImpl::getBackend() const { return MDevices[0]->getBackend(); }
+ol_symbol_handle_t
+ContextImpl::getOrCreateKernel(const DeviceImageManager &DeviceImage,
+ ol_device_handle_t DeviceHandle,
+ std::string_view KernelName) {
+ std::lock_guard<std::mutex> Guard(MProgramCacheMutex);
+
+ auto ImageIt = MPrograms.try_emplace(&DeviceImage).first;
+ ProgramsByDeviceT &ProgramsForImage = ImageIt->second;
+
+ auto ProgramIt = ProgramsForImage.find(DeviceHandle);
+ if (ProgramIt == ProgramsForImage.end()) {
+ // Constructing a ProgramWrapper calls olCreateProgram, so try_emplace is
+ // used rather than emplace: the latter would build a program even when one
+ // is already cached, only to destroy it again.
+ try {
+ ProgramIt = ProgramsForImage
+ .try_emplace(DeviceHandle, MOffloadContext, DeviceHandle,
+ DeviceImage)
+ .first;
+ } catch (...) {
+ // Do not leave an empty entry behind if program creation failed.
+ if (ProgramsForImage.empty())
+ MPrograms.erase(ImageIt);
+ throw;
+ }
+ }
+
+ return ProgramIt->second.getOrCreateKernel(KernelName);
+}
+
+void ContextImpl::releaseProgramsForImage(
+ const DeviceImageManager &DeviceImage) {
+ std::lock_guard<std::mutex> Guard(MProgramCacheMutex);
+ MPrograms.erase(&DeviceImage);
+}
+
+void ContextImpl::releaseAllPrograms() {
+ std::lock_guard<std::mutex> Guard(MProgramCacheMutex);
+ MPrograms.clear();
+}
+
} // namespace detail
_LIBSYCL_END_NAMESPACE_SYCL
diff --git a/libsycl/src/detail/context_impl.hpp b/libsycl/src/detail/context_impl.hpp
index ec9c8d10aca7e..2b01e911aa603 100644
--- a/libsycl/src/detail/context_impl.hpp
+++ b/libsycl/src/detail/context_impl.hpp
@@ -19,9 +19,14 @@
#include <sycl/__impl/context.hpp>
#include <sycl/__impl/detail/config.hpp>
+#include <detail/device_image_wrapper.hpp>
+
#include <OffloadAPI.h>
#include <functional>
+#include <mutex>
+#include <string_view>
+#include <unordered_map>
_LIBSYCL_BEGIN_NAMESPACE_SYCL
@@ -82,10 +87,43 @@ class ContextImpl : public std::enable_shared_from_this<ContextImpl> {
/// \return backend of the platform this context is associated with.
backend getBackend() const;
+ /// Returns the liboffload kernel symbol for the specified kernel, taken from
+ /// the program built in this context from the specified device image for the
+ /// specified device. Creates the program on first use.
+ /// This method is thread-safe.
+ /// \param DeviceImage the device image containing the kernel's device code.
+ /// \param DeviceHandle the liboffload handle of the device the program must
+ /// be compatible with.
+ /// \param KernelName the name of the kernel to look up.
+ /// \throw sycl::exception with sycl::errc::runtime when program creation or
+ /// symbol lookup fails.
+ /// \return the liboffload symbol handle of the kernel.
+ ol_symbol_handle_t getOrCreateKernel(const DeviceImageManager &DeviceImage,
+ ol_device_handle_t DeviceHandle,
+ std::string_view KernelName);
+
+ /// Destroys every program in this context that was created from the specified
+ /// device image, together with the kernel symbols taken from them. Called
+ /// while the image is being unregistered, before it is destroyed.
+ /// This method is thread-safe.
+ /// \param DeviceImage the device image whose programs must be released.
+ void releaseProgramsForImage(const DeviceImageManager &DeviceImage);
+
+ /// Destroys every program in this context, together with the kernel symbols
+ /// taken from them.
+ /// This method is thread-safe.
+ void releaseAllPrograms();
+
private:
const async_handler MAsyncHandler;
const std::vector<DeviceImpl *> MDevices;
ol_context_handle_t MOffloadContext{};
+
+ // TODO: later to replace with efficient kernel & program cache impl.
+ std::mutex MProgramCacheMutex;
+ using ProgramsByDeviceT =
+ std::unordered_map<ol_device_handle_t, ProgramWrapper>;
+ std::unordered_map<const DeviceImageManager *, ProgramsByDeviceT> MPrograms;
};
} // namespace detail
diff --git a/libsycl/src/detail/device_image_wrapper.cpp b/libsycl/src/detail/device_image_wrapper.cpp
index 17a830614bd0a..7f5582f3b680b 100644
--- a/libsycl/src/detail/device_image_wrapper.cpp
+++ b/libsycl/src/detail/device_image_wrapper.cpp
@@ -15,7 +15,7 @@ namespace detail {
ProgramWrapper::ProgramWrapper(ol_context_handle_t Context,
ol_device_handle_t Device,
- DeviceImageManager &DevImage) {
+ const DeviceImageManager &DevImage) {
assert(Context);
assert(Device);
@@ -30,13 +30,17 @@ ProgramWrapper::~ProgramWrapper() {
// TODO: define a way to report errors from dtors.
}
-ol_program_handle_t
-DeviceImageManager::getOrCreateProgram(ol_context_handle_t ContextHandle,
- ol_device_handle_t DeviceHandle) {
- const auto &[Iterator, Flag] = MPrograms.emplace(
- std::piecewise_construct, std::forward_as_tuple(DeviceHandle),
- std::forward_as_tuple(ContextHandle, DeviceHandle, *this));
- return Iterator->second.getOLHandle();
+ol_symbol_handle_t
+ProgramWrapper::getOrCreateKernel(std::string_view KernelName) {
+ auto It = MKernels.find(KernelName);
+ if (It != MKernels.end())
+ return It->second;
+
+ ol_symbol_handle_t Kernel{};
+ callAndThrow(olGetSymbol, MProgram, KernelName.data(), OL_SYMBOL_KIND_KERNEL,
+ &Kernel);
+ MKernels.emplace(KernelName, Kernel);
+ return Kernel;
}
} // namespace detail
diff --git a/libsycl/src/detail/device_image_wrapper.hpp b/libsycl/src/detail/device_image_wrapper.hpp
index 5dfd7d05ee6b2..5d639a2fe2970 100644
--- a/libsycl/src/detail/device_image_wrapper.hpp
+++ b/libsycl/src/detail/device_image_wrapper.hpp
@@ -22,6 +22,7 @@
#include <OffloadAPI.h>
#include <memory>
+#include <string_view>
#include <unordered_map>
_LIBSYCL_BEGIN_NAMESPACE_SYCL
@@ -41,7 +42,7 @@ class ProgramWrapper {
/// \throw sycl::exception with sycl::errc::runtime when failed to create the
/// program.
ProgramWrapper(ol_context_handle_t Context, ol_device_handle_t Device,
- DeviceImageManager &DevImage);
+ const DeviceImageManager &DevImage);
/// Releases the corresponding liboffload program handle by calling
/// olDestroyProgram.
@@ -55,12 +56,31 @@ class ProgramWrapper {
/// \return the corresponding liboffload program handle.
ol_program_handle_t getOLHandle() { return MProgram; }
+ /// Returns the liboffload kernel symbol for the specified kernel, looking it
+ /// up in this program on first use.
+ ///
+ /// Symbols belong to the program they were retrieved from: liboffload has no
+ /// olDestroySymbol, so they are released together with this program. Caching
+ /// them here rather than per device keeps a symbol from ever being handed out
+ /// for a program it does not belong to.
+ ///
+ /// \param KernelName the name of the kernel to look up.
+ /// \throw sycl::exception with sycl::errc::runtime when the symbol lookup
+ /// fails.
+ /// \return the liboffload symbol handle of the kernel.
+ ol_symbol_handle_t getOrCreateKernel(std::string_view KernelName);
+
private:
ol_program_handle_t MProgram{};
+
+ // Kernel names are backed by the "symbols" string of the device image this
+ // program was created from, so entries stay valid only while that image is
+ // registered. ContextImpl::releaseProgramsForImage() destroys this program
+ // before the image goes away.
+ std::unordered_map<std::string_view, ol_symbol_handle_t> MKernels;
};
-/// This class manages all work with device images: from data parsing to program
-/// creation.
+/// This class manages data parsing of device images.
class DeviceImageManager {
public:
DeviceImageManager(std::unique_ptr<llvm::object::OffloadBinary> Bin)
@@ -77,21 +97,7 @@ class DeviceImageManager {
/// \return a reference to the corresponding parsed OffloadBinary object.
const llvm::object::OffloadBinary &getOffloadBinary() const { return *MBin; }
- /// Returns a liboffload program which is compatible with the specified
- /// device. Searches among existing programs and creates a new one if no
- /// compatible image is found.
- /// \param ContextHandle the liboffload handle of the context to create the
- /// program in.
- /// \param DeviceHandle the liboffload handle of the device the program must
- /// be compatible with.
- /// \return the liboffload handle of the program compatible with the specified
- /// device.
- ol_program_handle_t getOrCreateProgram(ol_context_handle_t ContextHandle,
- ol_device_handle_t DeviceHandle);
-
protected:
- std::unordered_map<ol_device_handle_t, ProgramWrapper> MPrograms;
-
std::unique_ptr<llvm::object::OffloadBinary> MBin;
};
diff --git a/libsycl/src/detail/device_kernel_info.hpp b/libsycl/src/detail/device_kernel_info.hpp
index d5194f7619cd5..7899487b05369 100644
--- a/libsycl/src/detail/device_kernel_info.hpp
+++ b/libsycl/src/detail/device_kernel_info.hpp
@@ -20,10 +20,12 @@
#include <OffloadAPI.h>
+#include <string_view>
+
_LIBSYCL_BEGIN_NAMESPACE_SYCL
namespace detail {
-class ProgramAndKernelManager;
+class DeviceImageManager;
// TODO: Pointers to instances of this class are supported to be stored in
// header function templates as a static variable to avoid repeated runtime
@@ -44,36 +46,8 @@ class DeviceKernelInfo {
DeviceImageManager &getDeviceImage() const { return MDeviceImage; }
private:
- std::unordered_map<ol_device_handle_t, ol_symbol_handle_t> MBuiltKernels;
-
std::string_view MName;
DeviceImageManager &MDeviceImage;
-
- /// Searches for the existing kernel handle compatible with the specified
- /// device.
- /// \param Device the device the kernel must be compatible with.
- /// \return a liboffload kernel handle if a built kernel was found; otherwise
- /// returns nullptr.
- ol_symbol_handle_t getKernel(ol_device_handle_t Device) const {
- auto KernelIt = MBuiltKernels.find(Device);
- if (KernelIt == MBuiltKernels.end())
- return nullptr;
- return KernelIt->second;
- }
-
- /// Attaches a liboffload kernel handle to this device kernel info object.
- /// \param Device the device the kernel symbol was created for.
- /// \param Kernel the liboffload kernel symbol to attach.
- void addKernel(ol_device_handle_t Device, ol_symbol_handle_t Kernel) {
- assert(Kernel && "Invalid liboffload kernel handle");
- assert(Device && "Invalid liboffload device handle");
- assert((MBuiltKernels.find(Device) == MBuiltKernels.end()) &&
- "Kernel is being managed already");
- MBuiltKernels.insert({Device, Kernel});
- }
-
- /// Kernel info update is intended to be done only by ProgramAndKernelManager.
- friend class ProgramAndKernelManager;
};
} // namespace detail
diff --git a/libsycl/src/detail/program_manager.cpp b/libsycl/src/detail/program_manager.cpp
index 34e80435290e7..c594f8fe6542e 100644
--- a/libsycl/src/detail/program_manager.cpp
+++ b/libsycl/src/detail/program_manager.cpp
@@ -33,10 +33,38 @@ ProgramAndKernelManager::getDeviceKernelInfo(std::string_view KernelName) {
}
void ProgramAndKernelManager::releaseResources() {
+ std::lock_guard<std::mutex> Guard(MDataCollectionMutex);
+ // Contexts can outlive this call: platform default contexts are kept in the
+ // platform cache, which is static. Programs must not be left for
+ // their destructors to release, because olShutDown() follows this call.
+ for (const std::weak_ptr<ContextImpl> &WeakContext : MContextsWithPrograms) {
+ if (std::shared_ptr<ContextImpl> Context = WeakContext.lock())
+ Context->releaseAllPrograms();
+ }
+ MContextsWithPrograms.clear();
MDeviceKernelInfoMap.clear();
MDeviceImageManagers.clear();
}
+void ProgramAndKernelManager::trackContext(
+ const std::shared_ptr<ContextImpl> &Context) {
+ bool AlreadyTracked = false;
+ for (auto It = MContextsWithPrograms.begin();
+ It != MContextsWithPrograms.end();) {
+ std::shared_ptr<ContextImpl> TrackedContext = It->lock();
+ if (!TrackedContext) {
+ // Remove expired context from the tracking list.
+ It = MContextsWithPrograms.erase(It);
+ continue;
+ }
+ AlreadyTracked |= (TrackedContext == Context);
+ ++It;
+ }
+
+ if (!AlreadyTracked)
+ MContextsWithPrograms.push_back(Context);
+}
+
static inline bool
checkDeviceImageValidity(const llvm::object::OffloadBinary &OB) {
return (OB.getOffloadKind() == llvm::object::OFK_SYCL) &&
@@ -100,13 +128,21 @@ void ProgramAndKernelManager::unregisterFatBin(const void *BinaryStart,
return;
for (auto &Image : It->second) {
+ // Programs created from this image are owned by the contexts they were
+ // created in, so they have to be destroyed here: the image is about to go
+ // away, and the kernel names cached alongside those programs point into its
+ // memory, which may be unmapped right after this call.
+ for (const std::weak_ptr<ContextImpl> &WeakContext :
+ MContextsWithPrograms) {
+ if (std::shared_ptr<ContextImpl> Context = WeakContext.lock())
+ Context->releaseProgramsForImage(*Image);
+ }
+
llvm::StringRef Symbols = Image->getOffloadBinary().getString("symbols");
llvm::offloading::sycl::forEachSymbol(Symbols, [&](llvm::StringRef Name) {
if (auto KernelIt = MDeviceKernelInfoMap.find(std::string_view(Name));
KernelIt != MDeviceKernelInfoMap.end()) {
- // Programs are attached to the image and will be released with image
- // destruction. Clear only kernel specific data by destroying its kernel
- // info object.
+ // Clear kernel specific data by destroying its kernel info object.
MDeviceKernelInfoMap.erase(KernelIt);
}
});
@@ -129,29 +165,26 @@ static bool isImageCompatible(const DeviceImageManager &Image,
}
ol_symbol_handle_t ProgramAndKernelManager::getOrCreateKernel(
- DeviceKernelInfo &KernelInfo, ContextImpl &Context, DeviceImpl &Device) {
+ DeviceKernelInfo &KernelInfo, const std::shared_ptr<ContextImpl> &Context,
+ DeviceImpl &Device) {
+ assert(Context && "Context can't be nullptr");
std::lock_guard<std::mutex> KernelGuard(MDataCollectionMutex);
- if (auto Kernel = KernelInfo.getKernel(Device.getOLHandle()))
- return Kernel;
-
- auto &DeviceImage = KernelInfo.getDeviceImage();
+ DeviceImageManager &DeviceImage = KernelInfo.getDeviceImage();
if (!isImageCompatible(DeviceImage, Device))
throw exception(make_error_code(errc::runtime),
std::string("No compatible image for ") +
KernelInfo.getName().data() + " was found");
- auto DeviceHandle = Device.getOLHandle();
- auto Program =
- DeviceImage.getOrCreateProgram(Context.getOLHandleRef(), DeviceHandle);
+ // Track the context before it caches anything, so that unregisterFatBin() can
+ // reach the programs it is about to create.
+ trackContext(Context);
- ol_symbol_handle_t Kernel{};
- callAndThrow(olGetSymbol, Program, KernelInfo.getName().data(),
- OL_SYMBOL_KIND_KERNEL, &Kernel);
- KernelInfo.addKernel(DeviceHandle, Kernel);
- return Kernel;
+ // Lock order is MDataCollectionMutex -> ContextImpl::MProgramCacheMutex.
+ return Context->getOrCreateKernel(DeviceImage, Device.getOLHandle(),
+ KernelInfo.getName());
}
bool ProgramAndKernelManager::hasCompatibleImage(const DeviceImpl &Device) {
diff --git a/libsycl/src/detail/program_manager.hpp b/libsycl/src/detail/program_manager.hpp
index 59f77633debb1..1fe5df79a79d5 100644
--- a/libsycl/src/detail/program_manager.hpp
+++ b/libsycl/src/detail/program_manager.hpp
@@ -25,6 +25,7 @@
#include <OffloadAPI.h>
+#include <memory>
#include <mutex>
#include <unordered_map>
#include <vector>
@@ -86,9 +87,10 @@ class ProgramAndKernelManager {
/// \param Device the device for which this kernel must be compiled.
/// \return a liboffload kernel handle that is ready to be passed to kernel
/// execution methods.
- ol_symbol_handle_t getOrCreateKernel(DeviceKernelInfo &KernelInfo,
- ContextImpl &Context,
- DeviceImpl &Device);
+ ol_symbol_handle_t
+ getOrCreateKernel(DeviceKernelInfo &KernelInfo,
+ const std::shared_ptr<ContextImpl> &Context,
+ DeviceImpl &Device);
/// \return kernel info for the kernel with the specified name.
DeviceKernelInfo &getDeviceKernelInfo(std::string_view KernelName);
@@ -106,6 +108,12 @@ class ProgramAndKernelManager {
ProgramAndKernelManager(ProgramAndKernelManager const &) = delete;
ProgramAndKernelManager &operator=(ProgramAndKernelManager const &) = delete;
+ /// Adds the specified context to MContextsWithPrograms unless it is already
+ /// tracked, and drops the entries of contexts that have been destroyed.
+ /// MDataCollectionMutex must be held by the caller.
+ /// \param Context the context that is about to cache a program.
+ void trackContext(const std::shared_ptr<ContextImpl> &Context);
+
// Filled by registerFatBin(...).
// Map for storing device kernel information. Runtime lookup should be avoided
// by caching the pointers when possible.
@@ -121,6 +129,16 @@ class ProgramAndKernelManager {
std::unordered_map<BinaryStartKey, DeviceImageManagerVec>
MDeviceImageManagers;
+ // Contexts that may hold programs created from the device images above. A
+ // context owns its programs, so it is the only one that can destroy them, and
+ // it must do so before an image they were created from is destroyed.
+ //
+ // Entries are weak and pruned lazily: a context can be destroyed at any point
+ // and ~ContextImpl must not call back into this class, because that would
+ // take MDataCollectionMutex while holding ContextImpl::MProgramCacheMutex and
+ // invert the lock order used everywhere else.
+ std::vector<std::weak_ptr<ContextImpl>> MContextsWithPrograms;
+
// All work with device images and data related to it must be wrapped with a
// lock of this mutex.
std::mutex MDataCollectionMutex;
diff --git a/libsycl/src/detail/queue_impl.cpp b/libsycl/src/detail/queue_impl.cpp
index 881e4617b3a85..4e534e6480517 100644
--- a/libsycl/src/detail/queue_impl.cpp
+++ b/libsycl/src/detail/queue_impl.cpp
@@ -62,15 +62,21 @@ static void setKernelLaunchArgs(const detail::UnifiedRangeView &Range,
ArgsToSet.DynSharedMemory = 0;
}
-QueueImpl::QueueImpl(DeviceImpl &deviceImpl, const async_handler &asyncHandler,
+QueueImpl::QueueImpl(const std::shared_ptr<ContextImpl> &contextImpl,
+ DeviceImpl &deviceImpl, const async_handler &asyncHandler,
const property_list &propList, PrivateTag)
: MIsInorder(false), MAsyncHandler(asyncHandler), MPropList(propList),
- MDevice(deviceImpl),
- MContext(MDevice.getPlatformImpl().getDefaultContext()) {
- assert(MContext.getOLHandleRef() &&
- "Queue must be associated with a valid offload context");
- callAndThrow(olCreateQueue, MContext.getOLHandleRef(), MDevice.getOLHandle(),
- &MOffloadQueue);
+ MDevice(deviceImpl), MContext(contextImpl) {
+ assert(MContext && "Context impl ptr can't be nullptr");
+
+ ol_result_t Err = callNoCheck(olCreateQueue, MContext->getOLHandleRef(),
+ MDevice.getOLHandle(), &MOffloadQueue);
+ // liboffload guarantees OL_ERRC_INVALID_DEVICE when the device does not
+ // belong to the context.
+ if (isFailed(Err) && Err->Code == OL_ERRC_INVALID_DEVICE)
+ throw sycl::exception(sycl::make_error_code(sycl::errc::invalid),
+ "The device is not associated with the context.");
+ checkAndThrow(Err);
}
QueueImpl::~QueueImpl() {
diff --git a/libsycl/src/detail/queue_impl.hpp b/libsycl/src/detail/queue_impl.hpp
index c889a189a7ff4..66b2c00cf6466 100644
--- a/libsycl/src/detail/queue_impl.hpp
+++ b/libsycl/src/detail/queue_impl.hpp
@@ -47,7 +47,8 @@ class QueueImpl : public std::enable_shared_from_this<QueueImpl> {
/// submitted to the queue.
/// \param asyncHandler is a SYCL asynchronous exception handler.
/// \param propList is a list of properties to use for queue construction.
- explicit QueueImpl(DeviceImpl &deviceImpl, const async_handler &asyncHandler,
+ explicit QueueImpl(const std::shared_ptr<ContextImpl> &contextImpl,
+ DeviceImpl &deviceImpl, const async_handler &asyncHandler,
const property_list &propList, PrivateTag);
/// Constructs a QueueImpl with the provided arguments. Variadic helper.
@@ -61,7 +62,7 @@ class QueueImpl : public std::enable_shared_from_this<QueueImpl> {
backend getBackend() const noexcept;
/// \return the context implementation object this queue is associated with.
- ContextImpl &getContext() { return MContext; }
+ ContextImpl &getContext() { return *MContext; }
/// \return the device implementation object this queue is associated with.
DeviceImpl &getDevice() { return MDevice; }
@@ -136,7 +137,7 @@ class QueueImpl : public std::enable_shared_from_this<QueueImpl> {
const async_handler MAsyncHandler;
const property_list MPropList;
DeviceImpl &MDevice;
- ContextImpl &MContext;
+ const std::shared_ptr<ContextImpl> MContext;
// Submit data.
struct KernelSubmitInfo {
diff --git a/libsycl/src/queue.cpp b/libsycl/src/queue.cpp
index c257a09367f3f..f98858490f0c2 100644
--- a/libsycl/src/queue.cpp
+++ b/libsycl/src/queue.cpp
@@ -15,9 +15,10 @@
_LIBSYCL_BEGIN_NAMESPACE_SYCL
-queue::queue(const device &syclDevice, const async_handler &asyncHandler,
- const property_list &propList) {
- impl = detail::QueueImpl::create(*detail::getSyclObjImpl(syclDevice),
+queue::queue(const context &syclContext, const device &syclDevice,
+ const async_handler &asyncHandler, const property_list &propList) {
+ impl = detail::QueueImpl::create(detail::getSyclObjImpl(syclContext),
+ *detail::getSyclObjImpl(syclDevice),
asyncHandler, propList);
}
diff --git a/libsycl/unittests/mock/helpers.cpp b/libsycl/unittests/mock/helpers.cpp
index b3838daa00742..67f701790d6f2 100644
--- a/libsycl/unittests/mock/helpers.cpp
+++ b/libsycl/unittests/mock/helpers.cpp
@@ -222,7 +222,7 @@ void mock::MockLiboffload::initDefault() {
ON_CALL(*this, olCreateQueue)
.WillByDefault([](ol_context_handle_t Context, ol_device_handle_t Device,
ol_queue_handle_t *Queue) -> ol_result_t {
- std::ignore = Context;
+ EXPECT_NE(Context, nullptr);
EXPECT_NE(Device, nullptr);
EXPECT_NE(Queue, nullptr);
// Attach device as data to check what device queue belongs to if needed
diff --git a/libsycl/unittests/program_manager/CMakeLists.txt b/libsycl/unittests/program_manager/CMakeLists.txt
index e8169beb8a808..3604b86bf6ed2 100644
--- a/libsycl/unittests/program_manager/CMakeLists.txt
+++ b/libsycl/unittests/program_manager/CMakeLists.txt
@@ -1,3 +1,4 @@
-add_sycl_unittest(ProgramManagerTests
+add_sycl_unittest(ProgramManagerTests
register_and_unregister.cpp
+ program_cache.cpp
)
diff --git a/libsycl/unittests/program_manager/program_cache.cpp b/libsycl/unittests/program_manager/program_cache.cpp
new file mode 100644
index 0000000000000..76f05413e1a11
--- /dev/null
+++ b/libsycl/unittests/program_manager/program_cache.cpp
@@ -0,0 +1,208 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+///
+/// \file
+/// Tests for the per-context program cache: a program is identified by the
+/// (context, device, image) triple, and it must be destroyed before both the
+/// context it belongs to and the image it was created from.
+///
+//===----------------------------------------------------------------------===//
+
+#include <common/device_images.hpp>
+#include <common/scoped_binary_registration.hpp>
+#include <mock/helpers.hpp>
+
+#include <detail/context_impl.hpp>
+#include <detail/device_impl.hpp>
+#include <detail/program_manager.hpp>
+
+#include <sycl/__impl/device.hpp>
+#include <sycl/__impl/exception.hpp>
+#include <sycl/__impl/property_list.hpp>
+
+#include <gmock/gmock.h>
+#include <gtest/gtest.h>
+
+#include <array>
+#include <memory>
+#include <string>
+#include <string_view>
+#include <vector>
+
+#include <llvm/ADT/SmallVector.h>
+
+using namespace sycl;
+using namespace ::testing;
+
+namespace {
+
+/// Creates a context over a single device, bypassing sycl::context: SYCL 2020
+/// only exposes a platform's default context here, and these tests need several
+/// distinct contexts over the same device.
+std::shared_ptr<detail::ContextImpl> createContext(const device &Device) {
+ std::vector<detail::DeviceImpl *> Devices = {detail::getSyclObjImpl(Device)};
+ return detail::ContextImpl::create(
+ std::move(Devices), detail::defaultAsyncHandler, property_list{});
+}
+
+/// Allows the liboffload calls that these tests do not assert on, so that
+/// context and program teardown does not produce uninteresting call warnings.
+void allowContextAndProgramLifetimeCalls(mock::MockLiboffload &Mock) {
+ EXPECT_CALL(Mock, olCreateContext(_, _, _)).Times(AnyNumber());
+ EXPECT_CALL(Mock, olDestroyContext(_)).Times(AnyNumber());
+ EXPECT_CALL(Mock, olCreateProgram(_, _, _, _, _)).Times(AnyNumber());
+ EXPECT_CALL(Mock, olDestroyProgram(_)).Times(AnyNumber());
+ EXPECT_CALL(Mock, olGetSymbol(_, _, _, _)).Times(AnyNumber());
+}
+
+detail::DeviceKernelInfo &getKernelInfo(std::string_view KernelName) {
+ return detail::ProgramAndKernelManager::getInstance().getDeviceKernelInfo(
+ KernelName);
+}
+
+ol_symbol_handle_t
+getKernel(const std::shared_ptr<detail::ContextImpl> &Context,
+ const device &Device, std::string_view KernelName) {
+ return detail::ProgramAndKernelManager::getInstance().getOrCreateKernel(
+ getKernelInfo(KernelName), Context, *detail::getSyclObjImpl(Device));
+}
+
+} // namespace
+
+// A program belongs to the context it was created in, so two contexts over the
+// same device must not share one. Before the cache was keyed by context, the
+// second context silently reused the first context's program.
+TEST(ProgramCache, ProgramIsCreatedPerContext) {
+ mock::MockWrapper Mock;
+ allowContextAndProgramLifetimeCalls(Mock.get());
+
+ const std::string KernelName = "kernel";
+ sycl::unittests::ScopedKernelRegistration Registration(KernelName);
+
+ const device Device;
+ std::shared_ptr<detail::ContextImpl> FirstContext = createContext(Device);
+ std::shared_ptr<detail::ContextImpl> SecondContext = createContext(Device);
+
+ EXPECT_CALL(Mock.get(), olCreateProgram(_, _, _, _, _)).Times(2);
+
+ ol_symbol_handle_t FirstKernel = getKernel(FirstContext, Device, KernelName);
+ ol_symbol_handle_t SecondKernel =
+ getKernel(SecondContext, Device, KernelName);
+ EXPECT_NE(FirstKernel, nullptr);
+ EXPECT_NE(SecondKernel, nullptr);
+ EXPECT_NE(FirstKernel, SecondKernel);
+}
+
+// A repeated request within the same context must be served from the cache.
+TEST(ProgramCache, ProgramAndKernelAreCached) {
+ mock::MockWrapper Mock;
+ allowContextAndProgramLifetimeCalls(Mock.get());
+
+ const std::string KernelName = "kernel";
+ sycl::unittests::ScopedKernelRegistration Registration(KernelName);
+
+ const device Device;
+ std::shared_ptr<detail::ContextImpl> Context = createContext(Device);
+
+ EXPECT_CALL(Mock.get(), olCreateProgram(_, _, _, _, _)).Times(1);
+ EXPECT_CALL(Mock.get(), olGetSymbol(_, _, _, _)).Times(1);
+
+ ol_symbol_handle_t FirstKernel = getKernel(Context, Device, KernelName);
+ ol_symbol_handle_t SecondKernel = getKernel(Context, Device, KernelName);
+ EXPECT_EQ(FirstKernel, SecondKernel);
+}
+
+// Two images registered for the same device need two programs. The cache used
+// to be keyed by device alone, which handed out the first image's program for
+// kernels of the second one.
+TEST(ProgramCache, ProgramIsCreatedPerDeviceImage) {
+ mock::MockWrapper Mock;
+ allowContextAndProgramLifetimeCalls(Mock.get());
+
+ std::array<std::string, 2> KernelNames = {"image1kernel", "image2kernel"};
+ std::array<llvm::StringRef, 1> Image1Kernels = {KernelNames[0]};
+ std::array<llvm::StringRef, 1> Image2Kernels = {KernelNames[1]};
+
+ std::array<llvm::SmallString<0>, 2> Symbols;
+ llvm::offloading::sycl::writeSymbolTable(Image1Kernels, Symbols[0]);
+ llvm::offloading::sycl::writeSymbolTable(Image2Kernels, Symbols[1]);
+
+ llvm::SmallVector<llvm::object::OffloadBinary::OffloadingImage, 2> Images;
+ Images.push_back(sycl::unittests::createSYCLImage(Symbols[0]));
+ Images.push_back(sycl::unittests::createSYCLImage(Symbols[1]));
+ llvm::SmallString<0> Binary = llvm::object::OffloadBinary::write(Images);
+
+ detail::ProgramAndKernelManager &Manager =
+ detail::ProgramAndKernelManager::getInstance();
+ Manager.registerFatBin(Binary.data(), Binary.size());
+
+ const device Device;
+ std::shared_ptr<detail::ContextImpl> Context = createContext(Device);
+
+ EXPECT_CALL(Mock.get(), olCreateProgram(_, _, _, _, _)).Times(2);
+
+ ol_symbol_handle_t FirstKernel = getKernel(Context, Device, KernelNames[0]);
+ ol_symbol_handle_t SecondKernel = getKernel(Context, Device, KernelNames[1]);
+ EXPECT_NE(FirstKernel, SecondKernel);
+
+ Manager.unregisterFatBin(Binary.data(), Binary.size());
+}
+
+// liboffload does not reference-count contexts: a program tied to a context
+// that has already been destroyed is in an undefined state, so olDestroyProgram
+// must come first.
+TEST(ProgramCache, ProgramsAreDestroyedBeforeContext) {
+ mock::MockWrapper Mock;
+ allowContextAndProgramLifetimeCalls(Mock.get());
+
+ const std::string KernelName = "kernel";
+ sycl::unittests::ScopedKernelRegistration Registration(KernelName);
+
+ const device Device;
+ std::shared_ptr<detail::ContextImpl> Context = createContext(Device);
+ EXPECT_NE(getKernel(Context, Device, KernelName), nullptr);
+
+ {
+ InSequence Sequence;
+ EXPECT_CALL(Mock.get(), olDestroyProgram(_)).Times(1);
+ EXPECT_CALL(Mock.get(), olDestroyContext(_)).Times(1);
+ }
+
+ Context.reset();
+}
+
+// Programs are created from the image's memory and cache kernel names that
+// point into it, so unregistering the image must release them even though the
+// context that owns them stays alive.
+TEST(ProgramCache, ProgramsAreDestroyedOnImageUnregistration) {
+ mock::MockWrapper Mock;
+ allowContextAndProgramLifetimeCalls(Mock.get());
+
+ const std::string KernelName = "kernel";
+ std::array<llvm::StringRef, 1> KernelNames = {KernelName};
+ llvm::SmallString<0> Binary =
+ sycl::unittests::createSYCLDeviceBinary(KernelNames);
+
+ detail::ProgramAndKernelManager &Manager =
+ detail::ProgramAndKernelManager::getInstance();
+ Manager.registerFatBin(Binary.data(), Binary.size());
+
+ const device Device;
+ std::shared_ptr<detail::ContextImpl> Context = createContext(Device);
+ EXPECT_NE(getKernel(Context, Device, KernelName), nullptr);
+
+ EXPECT_CALL(Mock.get(), olDestroyProgram(_)).Times(1);
+ Manager.unregisterFatBin(Binary.data(), Binary.size());
+ // Qualified: the local Mock variable shadows ::testing::Mock here.
+ ::testing::Mock::VerifyAndClearExpectations(&Mock.get());
+
+ // Nothing is left for the context to release.
+ EXPECT_CALL(Mock.get(), olDestroyProgram(_)).Times(0);
+ EXPECT_CALL(Mock.get(), olDestroyContext(_)).Times(1);
+ Context.reset();
+}
diff --git a/libsycl/unittests/queue/queue.cpp b/libsycl/unittests/queue/queue.cpp
index c505f78ecbdd4..37174e571fd93 100644
--- a/libsycl/unittests/queue/queue.cpp
+++ b/libsycl/unittests/queue/queue.cpp
@@ -9,6 +9,7 @@
#include <mock/helpers.hpp>
#include <sycl/__impl/detail/config.hpp>
+#include <sycl/__impl/platform.hpp>
#include <sycl/__impl/queue.hpp>
#include <gmock/gmock.h>
@@ -28,3 +29,26 @@ TEST(Queue, CommonQueriesAndLifetime) {
EXPECT_EQ(Q.is_in_order(), false);
}
}
+
+TEST(Queue, ContextAndDeviceConstructor) {
+ mock::MockWrapper Mock;
+
+ const device Device;
+ const context Context = Device.get_platform().khr_get_default_context();
+ const auto Selector = [](const device &) { return 1; };
+ const async_handler AsyncHandler = [](exception_list) {};
+ EXPECT_CALL(Mock.get(), olCreateQueue(_, _, _)).Times(3);
+ EXPECT_CALL(Mock.get(), olDestroyQueue(_)).Times(3);
+
+ queue Queue(Context, Device);
+ EXPECT_EQ(Queue.get_context(), Context);
+ EXPECT_EQ(Queue.get_device(), Device);
+
+ queue SelectorQueue(Context, Selector);
+ EXPECT_EQ(SelectorQueue.get_context(), Context);
+ EXPECT_EQ(SelectorQueue.get_device(), Device);
+
+ queue AsyncSelectorQueue(Context, Selector, AsyncHandler);
+ EXPECT_EQ(AsyncSelectorQueue.get_context(), Context);
+ EXPECT_EQ(AsyncSelectorQueue.get_device(), Device);
+}
More information about the llvm-commits
mailing list