[llvm] [libsycl] Introduce Kernel name based cache (PR #225865)
via llvm-commits
llvm-commits at lists.llvm.org
Thu Oct 1 05:27:47 PDT 2026
https://github.com/Robertkq updated https://github.com/llvm/llvm-project/pull/225865
>From fe8551bb984ad9fd75cbe7be0d91d52faee17430 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Tue, 22 Sep 2026 17:57:06 +0300
Subject: [PATCH 1/5] Introduce DeviceKernelInfo cache mechanism
---
libsycl/src/CMakeLists.txt | 1 +
libsycl/src/detail/device_kernel_info.cpp | 34 +++++++++++++++++++++++
libsycl/src/detail/device_kernel_info.hpp | 23 +++++++++++++++
libsycl/src/detail/program_manager.cpp | 11 ++++++--
4 files changed, 67 insertions(+), 2 deletions(-)
create mode 100644 libsycl/src/detail/device_kernel_info.cpp
diff --git a/libsycl/src/CMakeLists.txt b/libsycl/src/CMakeLists.txt
index 870fc06b6334e..fea376266eca2 100644
--- a/libsycl/src/CMakeLists.txt
+++ b/libsycl/src/CMakeLists.txt
@@ -98,6 +98,7 @@ set(LIBSYCL_SOURCES
"detail/event_impl.cpp"
"detail/device_image_wrapper.cpp"
"detail/device_impl.cpp"
+ "detail/device_kernel_info.cpp"
"detail/global_objects.cpp"
"detail/platform_impl.cpp"
"detail/program_manager.cpp"
diff --git a/libsycl/src/detail/device_kernel_info.cpp b/libsycl/src/detail/device_kernel_info.cpp
new file mode 100644
index 0000000000000..4204e5f3a44c5
--- /dev/null
+++ b/libsycl/src/detail/device_kernel_info.cpp
@@ -0,0 +1,34 @@
+//===----------------------------------------------------------------------===//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+
+#include <detail/device_kernel_info.hpp>
+
+_LIBSYCL_BEGIN_NAMESPACE_SYCL
+
+namespace detail {
+
+ol_symbol_handle_t
+DeviceKernelInfo::tryGetCachedKernel(ContextImpl *Context,
+ ol_device_handle_t Device) {
+ CacheKeyT Key = {Context, Device};
+ std::lock_guard<std::mutex> Guard(MCacheMutex);
+ if (auto Result = MCache.find(Key); Result != MCache.end())
+ return Result->second;
+ return nullptr;
+}
+
+void DeviceKernelInfo::cacheKernel(ContextImpl *Context,
+ ol_device_handle_t Device,
+ ol_symbol_handle_t Kernel) {
+ CacheKeyT Key = {Context, Device};
+ std::lock_guard<std::mutex> Guard(MCacheMutex);
+ MCache.try_emplace(Key, Kernel);
+}
+} // namespace detail
+
+_LIBSYCL_END_NAMESPACE_SYCL
diff --git a/libsycl/src/detail/device_kernel_info.hpp b/libsycl/src/detail/device_kernel_info.hpp
index 7899487b05369..3de8a126b909b 100644
--- a/libsycl/src/detail/device_kernel_info.hpp
+++ b/libsycl/src/detail/device_kernel_info.hpp
@@ -20,12 +20,17 @@
#include <OffloadAPI.h>
+#include <cstddef>
+#include <mutex>
#include <string_view>
+#include <unordered_map>
+#include <utility>
_LIBSYCL_BEGIN_NAMESPACE_SYCL
namespace detail {
class DeviceImageManager;
+class ContextImpl;
// TODO: Pointers to instances of this class are supported to be stored in
// header function templates as a static variable to avoid repeated runtime
@@ -45,9 +50,27 @@ class DeviceKernelInfo {
/// \return the device image containing the device code of this kernel.
DeviceImageManager &getDeviceImage() const { return MDeviceImage; }
+ /// Returns a cache entry for pair key \p Context & \p Device if successful,
+ /// null otherwise
+ ol_symbol_handle_t tryGetCachedKernel(ContextImpl *Context,
+ ol_device_handle_t Device);
+
+ /// Adds \p Kernel to the cache by using the pair key \p Context & \p Device
+ void cacheKernel(ContextImpl *Context, ol_device_handle_t Device,
+ ol_symbol_handle_t Kernel);
+
private:
std::string_view MName;
DeviceImageManager &MDeviceImage;
+
+ using CacheKeyT = std::pair<ContextImpl *, ol_device_handle_t>;
+ struct CacheKeyHash {
+ std::size_t operator()(const CacheKeyT &Key) const noexcept {
+ return std::hash<ContextImpl *>{}(Key.first);
+ }
+ };
+ std::mutex MCacheMutex;
+ std::unordered_map<CacheKeyT, ol_symbol_handle_t, CacheKeyHash> MCache;
};
} // namespace detail
diff --git a/libsycl/src/detail/program_manager.cpp b/libsycl/src/detail/program_manager.cpp
index c594f8fe6542e..7c5d1af141616 100644
--- a/libsycl/src/detail/program_manager.cpp
+++ b/libsycl/src/detail/program_manager.cpp
@@ -169,6 +169,11 @@ ol_symbol_handle_t ProgramAndKernelManager::getOrCreateKernel(
DeviceImpl &Device) {
assert(Context && "Context can't be nullptr");
+ if (ol_symbol_handle_t CachedKernel =
+ KernelInfo.tryGetCachedKernel(Context.get(), Device.getOLHandle())) {
+ return CachedKernel;
+ }
+
std::lock_guard<std::mutex> KernelGuard(MDataCollectionMutex);
DeviceImageManager &DeviceImage = KernelInfo.getDeviceImage();
@@ -183,8 +188,10 @@ ol_symbol_handle_t ProgramAndKernelManager::getOrCreateKernel(
trackContext(Context);
// Lock order is MDataCollectionMutex -> ContextImpl::MProgramCacheMutex.
- return Context->getOrCreateKernel(DeviceImage, Device.getOLHandle(),
- KernelInfo.getName());
+ ol_symbol_handle_t Kernel = Context->getOrCreateKernel(
+ DeviceImage, Device.getOLHandle(), KernelInfo.getName());
+ KernelInfo.cacheKernel(Context.get(), Device.getOLHandle(), Kernel);
+ return Kernel;
}
bool ProgramAndKernelManager::hasCompatibleImage(const DeviceImpl &Device) {
>From a1683e84c245ab875d9dc2309aa035ad7d712fc3 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Tue, 22 Sep 2026 22:54:44 +0300
Subject: [PATCH 2/5] Drop DeviceKernelInfo cache entries when their context is
destroyed
---
libsycl/src/detail/context_impl.cpp | 14 ++++++++++++++
libsycl/src/detail/context_impl.hpp | 11 +++++++++++
libsycl/src/detail/device_kernel_info.cpp | 18 +++++++++++++++++-
libsycl/src/detail/device_kernel_info.hpp | 4 ++++
4 files changed, 46 insertions(+), 1 deletion(-)
diff --git a/libsycl/src/detail/context_impl.cpp b/libsycl/src/detail/context_impl.cpp
index 0525cee4535c7..82385d2ce67ec 100644
--- a/libsycl/src/detail/context_impl.cpp
+++ b/libsycl/src/detail/context_impl.cpp
@@ -7,6 +7,7 @@
//===----------------------------------------------------------------------===//
#include <detail/context_impl.hpp>
+#include <detail/device_kernel_info.hpp>
#include <detail/platform_impl.hpp>
_LIBSYCL_BEGIN_NAMESPACE_SYCL
@@ -40,6 +41,14 @@ ContextImpl::ContextImpl(std::vector<DeviceImpl *> &&DeviceList,
ContextImpl::~ContextImpl() {
assert(MOffloadContext && "Context must be created in ctor");
+ // Drop this context's entries from every DeviceKernelInfo cache that holds
+ // one, before the programs those entries point into are destroyed below.
+ {
+ std::lock_guard<std::mutex> Guard(MTrackedKernelInfosMutex);
+ for (DeviceKernelInfo *Info : MTrackedKernelInfos) {
+ Info->removeContext(this);
+ }
+ }
// 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
@@ -101,5 +110,10 @@ void ContextImpl::releaseAllPrograms() {
MPrograms.clear();
}
+void ContextImpl::trackKernelInfoCache(DeviceKernelInfo *Info) {
+ std::lock_guard<std::mutex> Guard(MTrackedKernelInfosMutex);
+ MTrackedKernelInfos.insert(Info);
+}
+
} // namespace detail
_LIBSYCL_END_NAMESPACE_SYCL
diff --git a/libsycl/src/detail/context_impl.hpp b/libsycl/src/detail/context_impl.hpp
index 2b01e911aa603..e14ead9d0e490 100644
--- a/libsycl/src/detail/context_impl.hpp
+++ b/libsycl/src/detail/context_impl.hpp
@@ -27,6 +27,7 @@
#include <mutex>
#include <string_view>
#include <unordered_map>
+#include <unordered_set>
_LIBSYCL_BEGIN_NAMESPACE_SYCL
@@ -36,6 +37,7 @@ namespace detail {
class PlatformImpl;
class DeviceImpl;
+class DeviceKernelInfo;
/// Context represents the runtime data structures and state required by a SYCL
/// backend API to interact with a group of devices associated with a platform.
@@ -114,6 +116,12 @@ class ContextImpl : public std::enable_shared_from_this<ContextImpl> {
/// This method is thread-safe.
void releaseAllPrograms();
+ /// Records that \p Info has (or is about to have) a cache entry keyed by
+ /// this context, so that the entry can be removed when this context is
+ /// destroyed. Safe to call more than once for the same \p Info.
+ /// This method is thread-safe.
+ void trackKernelInfoCache(DeviceKernelInfo *Info);
+
private:
const async_handler MAsyncHandler;
const std::vector<DeviceImpl *> MDevices;
@@ -124,6 +132,9 @@ class ContextImpl : public std::enable_shared_from_this<ContextImpl> {
using ProgramsByDeviceT =
std::unordered_map<ol_device_handle_t, ProgramWrapper>;
std::unordered_map<const DeviceImageManager *, ProgramsByDeviceT> MPrograms;
+
+ std::mutex MTrackedKernelInfosMutex;
+ std::unordered_set<DeviceKernelInfo *> MTrackedKernelInfos;
};
} // namespace detail
diff --git a/libsycl/src/detail/device_kernel_info.cpp b/libsycl/src/detail/device_kernel_info.cpp
index 4204e5f3a44c5..f4f0361383bc0 100644
--- a/libsycl/src/detail/device_kernel_info.cpp
+++ b/libsycl/src/detail/device_kernel_info.cpp
@@ -6,6 +6,7 @@
//
//===----------------------------------------------------------------------===//
+#include <detail/context_impl.hpp>
#include <detail/device_kernel_info.hpp>
_LIBSYCL_BEGIN_NAMESPACE_SYCL
@@ -26,8 +27,23 @@ void DeviceKernelInfo::cacheKernel(ContextImpl *Context,
ol_device_handle_t Device,
ol_symbol_handle_t Kernel) {
CacheKeyT Key = {Context, Device};
+ {
+ std::lock_guard<std::mutex> Guard(MCacheMutex);
+ MCache.try_emplace(Key, Kernel);
+ }
+ Context->trackKernelInfoCache(this);
+}
+
+void DeviceKernelInfo::removeContext(ContextImpl *Context) {
std::lock_guard<std::mutex> Guard(MCacheMutex);
- MCache.try_emplace(Key, Kernel);
+ for (auto It = MCache.begin(); It != MCache.end();) {
+ CacheKeyT Key = It->first;
+ if (Key.first == Context) {
+ It = MCache.erase(It);
+ } else {
+ ++It;
+ }
+ }
}
} // namespace detail
diff --git a/libsycl/src/detail/device_kernel_info.hpp b/libsycl/src/detail/device_kernel_info.hpp
index 3de8a126b909b..81edf8de240aa 100644
--- a/libsycl/src/detail/device_kernel_info.hpp
+++ b/libsycl/src/detail/device_kernel_info.hpp
@@ -59,6 +59,10 @@ class DeviceKernelInfo {
void cacheKernel(ContextImpl *Context, ol_device_handle_t Device,
ol_symbol_handle_t Kernel);
+ /// Removes every cache entry keyed by \p Context, regardless of device.
+ /// Called by ContextImpl when it is being destroyed.
+ void removeContext(ContextImpl *Context);
+
private:
std::string_view MName;
DeviceImageManager &MDeviceImage;
>From c50c4c94f7a4e22d70bc10d25be527034d023004 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Wed, 23 Sep 2026 00:06:16 +0300
Subject: [PATCH 3/5] Untrack DeviceKernelInfo entries at points near
termination, rename cache functions
---
libsycl/src/detail/context_impl.cpp | 7 ++++++-
libsycl/src/detail/context_impl.hpp | 6 ++++++
libsycl/src/detail/device_kernel_info.cpp | 8 ++++----
libsycl/src/detail/device_kernel_info.hpp | 6 +++---
libsycl/src/detail/program_manager.cpp | 18 ++++++++++++++++--
5 files changed, 35 insertions(+), 10 deletions(-)
diff --git a/libsycl/src/detail/context_impl.cpp b/libsycl/src/detail/context_impl.cpp
index 82385d2ce67ec..b0915792be182 100644
--- a/libsycl/src/detail/context_impl.cpp
+++ b/libsycl/src/detail/context_impl.cpp
@@ -46,7 +46,7 @@ ContextImpl::~ContextImpl() {
{
std::lock_guard<std::mutex> Guard(MTrackedKernelInfosMutex);
for (DeviceKernelInfo *Info : MTrackedKernelInfos) {
- Info->removeContext(this);
+ Info->removeCachedKernelsFor(this);
}
}
// liboffload does not reference-count contexts: every resource tied to a
@@ -115,5 +115,10 @@ void ContextImpl::trackKernelInfoCache(DeviceKernelInfo *Info) {
MTrackedKernelInfos.insert(Info);
}
+void ContextImpl::forgetKernelInfoCache(DeviceKernelInfo *Info) {
+ std::lock_guard<std::mutex> Guard(MTrackedKernelInfosMutex);
+ MTrackedKernelInfos.erase(Info);
+}
+
} // namespace detail
_LIBSYCL_END_NAMESPACE_SYCL
diff --git a/libsycl/src/detail/context_impl.hpp b/libsycl/src/detail/context_impl.hpp
index e14ead9d0e490..7236a9d238609 100644
--- a/libsycl/src/detail/context_impl.hpp
+++ b/libsycl/src/detail/context_impl.hpp
@@ -122,6 +122,12 @@ class ContextImpl : public std::enable_shared_from_this<ContextImpl> {
/// This method is thread-safe.
void trackKernelInfoCache(DeviceKernelInfo *Info);
+ /// Removes \p Info from the set of tracked caches, e.g. because \p Info is
+ /// about to be destroyed (its owning device image is being unregistered).
+ /// No-op if \p Info was not tracked.
+ /// This method is thread-safe.
+ void forgetKernelInfoCache(DeviceKernelInfo *Info);
+
private:
const async_handler MAsyncHandler;
const std::vector<DeviceImpl *> MDevices;
diff --git a/libsycl/src/detail/device_kernel_info.cpp b/libsycl/src/detail/device_kernel_info.cpp
index f4f0361383bc0..5dc8485487d2e 100644
--- a/libsycl/src/detail/device_kernel_info.cpp
+++ b/libsycl/src/detail/device_kernel_info.cpp
@@ -23,9 +23,9 @@ DeviceKernelInfo::tryGetCachedKernel(ContextImpl *Context,
return nullptr;
}
-void DeviceKernelInfo::cacheKernel(ContextImpl *Context,
- ol_device_handle_t Device,
- ol_symbol_handle_t Kernel) {
+void DeviceKernelInfo::addCachedKernel(ContextImpl *Context,
+ ol_device_handle_t Device,
+ ol_symbol_handle_t Kernel) {
CacheKeyT Key = {Context, Device};
{
std::lock_guard<std::mutex> Guard(MCacheMutex);
@@ -34,7 +34,7 @@ void DeviceKernelInfo::cacheKernel(ContextImpl *Context,
Context->trackKernelInfoCache(this);
}
-void DeviceKernelInfo::removeContext(ContextImpl *Context) {
+void DeviceKernelInfo::removeCachedKernelsFor(ContextImpl *Context) {
std::lock_guard<std::mutex> Guard(MCacheMutex);
for (auto It = MCache.begin(); It != MCache.end();) {
CacheKeyT Key = It->first;
diff --git a/libsycl/src/detail/device_kernel_info.hpp b/libsycl/src/detail/device_kernel_info.hpp
index 81edf8de240aa..340ff5140c45a 100644
--- a/libsycl/src/detail/device_kernel_info.hpp
+++ b/libsycl/src/detail/device_kernel_info.hpp
@@ -56,12 +56,12 @@ class DeviceKernelInfo {
ol_device_handle_t Device);
/// Adds \p Kernel to the cache by using the pair key \p Context & \p Device
- void cacheKernel(ContextImpl *Context, ol_device_handle_t Device,
- ol_symbol_handle_t Kernel);
+ void addCachedKernel(ContextImpl *Context, ol_device_handle_t Device,
+ ol_symbol_handle_t Kernel);
/// Removes every cache entry keyed by \p Context, regardless of device.
/// Called by ContextImpl when it is being destroyed.
- void removeContext(ContextImpl *Context);
+ void removeCachedKernelsFor(ContextImpl *Context);
private:
std::string_view MName;
diff --git a/libsycl/src/detail/program_manager.cpp b/libsycl/src/detail/program_manager.cpp
index 7c5d1af141616..c839efab3a634 100644
--- a/libsycl/src/detail/program_manager.cpp
+++ b/libsycl/src/detail/program_manager.cpp
@@ -38,8 +38,14 @@ void ProgramAndKernelManager::releaseResources() {
// 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())
+ if (std::shared_ptr<ContextImpl> Context = WeakContext.lock()) {
+ // Every DeviceKernelInfo below is about to be destroyed: make sure this
+ // context, if it outlives this call, does not keep pointers to them.
+ for (auto &[Name, Info] : MDeviceKernelInfoMap) {
+ Context->forgetKernelInfoCache(&Info);
+ }
Context->releaseAllPrograms();
+ }
}
MContextsWithPrograms.clear();
MDeviceKernelInfoMap.clear();
@@ -142,6 +148,14 @@ void ProgramAndKernelManager::unregisterFatBin(const void *BinaryStart,
llvm::offloading::sycl::forEachSymbol(Symbols, [&](llvm::StringRef Name) {
if (auto KernelIt = MDeviceKernelInfoMap.find(std::string_view(Name));
KernelIt != MDeviceKernelInfoMap.end()) {
+ // Remove this DeviceKernel info from every live context tracking it
+ DeviceKernelInfo *Info = &KernelIt->second;
+ for (const std::weak_ptr<ContextImpl> &WeakContext :
+ MContextsWithPrograms) {
+ if (std::shared_ptr<ContextImpl> Context = WeakContext.lock()) {
+ Context->forgetKernelInfoCache(Info);
+ }
+ }
// Clear kernel specific data by destroying its kernel info object.
MDeviceKernelInfoMap.erase(KernelIt);
}
@@ -190,7 +204,7 @@ ol_symbol_handle_t ProgramAndKernelManager::getOrCreateKernel(
// Lock order is MDataCollectionMutex -> ContextImpl::MProgramCacheMutex.
ol_symbol_handle_t Kernel = Context->getOrCreateKernel(
DeviceImage, Device.getOLHandle(), KernelInfo.getName());
- KernelInfo.cacheKernel(Context.get(), Device.getOLHandle(), Kernel);
+ KernelInfo.addCachedKernel(Context.get(), Device.getOLHandle(), Kernel);
return Kernel;
}
>From da45189cbb555b000aa8f8fc60877cd3739f5f35 Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Wed, 23 Sep 2026 20:18:38 +0300
Subject: [PATCH 4/5] Test DeviceKernelInfo's cache & restructed tests in
fixture
---
.../program_manager/program_cache.cpp | 93 +++++++++++++++----
1 file changed, 73 insertions(+), 20 deletions(-)
diff --git a/libsycl/unittests/program_manager/program_cache.cpp b/libsycl/unittests/program_manager/program_cache.cpp
index 9f2ba876ce533..2e20fecd38620 100644
--- a/libsycl/unittests/program_manager/program_cache.cpp
+++ b/libsycl/unittests/program_manager/program_cache.cpp
@@ -72,14 +72,79 @@ getKernel(const std::shared_ptr<detail::ContextImpl> &Context,
getKernelInfo(KernelName), Context, *detail::getSyclObjImpl(Device));
}
+class ProgramCacheTest : public testing::Test {
+protected:
+ void SetUp() override { allowContextAndProgramLifetimeCalls(Mock.get()); }
+
+ mock::MockWrapper Mock;
+};
+
} // namespace
+// A repeat request for the same (kernel, context, device) must be served
+// entirely by DeviceKernelInfo's own cache
+TEST_F(ProgramCacheTest, RepeatedKernelLookupSkipsCompatibilityCheck) {
+ 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(), olIsValidBinary(_, _, _, _)).Times(1);
+ EXPECT_CALL(Mock.get(), olCreateProgram(_, _, _, _, _)).Times(1);
+
+ ol_symbol_handle_t FirstKernel = getKernel(Context, Device, KernelName);
+ ol_symbol_handle_t SecondKernel = getKernel(Context, Device, KernelName);
+ EXPECT_EQ(FirstKernel, SecondKernel);
+}
+
+// A cache hit for one context must never be handed to a different context,
+// even for the exact same kernel and device.
+TEST_F(ProgramCacheTest, KernelCacheIsScopedPerContext) {
+ 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(), olIsValidBinary(_, _, _, _)).Times(2);
+ 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, SecondKernel);
+}
+
+// Destroying a context must remove its entries from every DeviceKernelInfo
+// cache it wrote into: otherwise, if a later context happens to be allocated
+// at the same address, it could be handed a symbol handle into an
+// already-destroyed program.
+TEST_F(ProgramCacheTest, ContextDestructionDropsCacheEntries) {
+ const std::string KernelName = "kernel";
+ sycl::unittests::ScopedKernelRegistration Registration(KernelName);
+
+ const device Device;
+ std::shared_ptr<detail::ContextImpl> Context = createContext(Device);
+ detail::ContextImpl *RawContext = Context.get();
+ ol_device_handle_t DeviceHandle =
+ detail::getSyclObjImpl(Device)->getOLHandle();
+
+ EXPECT_NE(getKernel(Context, Device, KernelName),
+ nullptr); // Populate the cache.
+
+ detail::DeviceKernelInfo &Info = getKernelInfo(KernelName);
+ EXPECT_NE(Info.tryGetCachedKernel(RawContext, DeviceHandle), nullptr);
+
+ Context.reset(); // Destroy the context, RawContext is dandling now.
+
+ EXPECT_EQ(Info.tryGetCachedKernel(RawContext, DeviceHandle), nullptr);
+}
+
// A program belongs to the context it was created in, so two contexts over the
// same device must not share one.
-TEST(ProgramCache, ProgramIsCreatedPerContext) {
- mock::MockWrapper Mock;
- allowContextAndProgramLifetimeCalls(Mock.get());
-
+TEST_F(ProgramCacheTest, ProgramIsCreatedPerContext) {
const std::string KernelName = "kernel";
sycl::unittests::ScopedKernelRegistration Registration(KernelName);
@@ -98,10 +163,7 @@ TEST(ProgramCache, ProgramIsCreatedPerContext) {
}
// A repeated request within the same context must be served from the cache.
-TEST(ProgramCache, ProgramAndKernelAreCached) {
- mock::MockWrapper Mock;
- allowContextAndProgramLifetimeCalls(Mock.get());
-
+TEST_F(ProgramCacheTest, ProgramAndKernelAreCached) {
const std::string KernelName = "kernel";
sycl::unittests::ScopedKernelRegistration Registration(KernelName);
@@ -119,10 +181,7 @@ TEST(ProgramCache, ProgramAndKernelAreCached) {
// 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());
-
+TEST_F(ProgramCacheTest, ProgramIsCreatedPerDeviceImage) {
std::array<std::string, 2> KernelNames = {"image1kernel", "image2kernel"};
std::array<llvm::StringRef, 1> Image1Kernels = {KernelNames[0]};
std::array<llvm::StringRef, 1> Image2Kernels = {KernelNames[1]};
@@ -155,10 +214,7 @@ TEST(ProgramCache, ProgramIsCreatedPerDeviceImage) {
// 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());
-
+TEST_F(ProgramCacheTest, ProgramsAreDestroyedBeforeContext) {
const std::string KernelName = "kernel";
sycl::unittests::ScopedKernelRegistration Registration(KernelName);
@@ -178,10 +234,7 @@ TEST(ProgramCache, ProgramsAreDestroyedBeforeContext) {
// 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());
-
+TEST_F(ProgramCacheTest, ProgramsAreDestroyedOnImageUnregistration) {
const std::string KernelName = "kernel";
std::array<llvm::StringRef, 1> KernelNames = {KernelName};
llvm::SmallString<0> Binary =
>From 45854d5b1708a6d7a522c021e0397c24455cbc1f Mon Sep 17 00:00:00 2001
From: Robertkq <robertvuia06 at gmail.com>
Date: Thu, 1 Oct 2026 15:27:24 +0300
Subject: [PATCH 5/5] code style & hash pair
---
libsycl/src/detail/context_impl.cpp | 3 +--
libsycl/src/detail/device_kernel_info.cpp | 5 ++---
libsycl/src/detail/device_kernel_info.hpp | 4 +++-
libsycl/src/detail/program_manager.cpp | 9 +++------
4 files changed, 9 insertions(+), 12 deletions(-)
diff --git a/libsycl/src/detail/context_impl.cpp b/libsycl/src/detail/context_impl.cpp
index b0915792be182..e66fb788d451d 100644
--- a/libsycl/src/detail/context_impl.cpp
+++ b/libsycl/src/detail/context_impl.cpp
@@ -45,9 +45,8 @@ ContextImpl::~ContextImpl() {
// one, before the programs those entries point into are destroyed below.
{
std::lock_guard<std::mutex> Guard(MTrackedKernelInfosMutex);
- for (DeviceKernelInfo *Info : MTrackedKernelInfos) {
+ for (DeviceKernelInfo *Info : MTrackedKernelInfos)
Info->removeCachedKernelsFor(this);
- }
}
// liboffload does not reference-count contexts: every resource tied to a
// context must be released before olDestroyContext, otherwise it is left in
diff --git a/libsycl/src/detail/device_kernel_info.cpp b/libsycl/src/detail/device_kernel_info.cpp
index 5dc8485487d2e..159b90c006fe7 100644
--- a/libsycl/src/detail/device_kernel_info.cpp
+++ b/libsycl/src/detail/device_kernel_info.cpp
@@ -38,11 +38,10 @@ void DeviceKernelInfo::removeCachedKernelsFor(ContextImpl *Context) {
std::lock_guard<std::mutex> Guard(MCacheMutex);
for (auto It = MCache.begin(); It != MCache.end();) {
CacheKeyT Key = It->first;
- if (Key.first == Context) {
+ if (Key.first == Context)
It = MCache.erase(It);
- } else {
+ else
++It;
- }
}
}
} // namespace detail
diff --git a/libsycl/src/detail/device_kernel_info.hpp b/libsycl/src/detail/device_kernel_info.hpp
index 340ff5140c45a..dd1605f3e6dbb 100644
--- a/libsycl/src/detail/device_kernel_info.hpp
+++ b/libsycl/src/detail/device_kernel_info.hpp
@@ -20,6 +20,8 @@
#include <OffloadAPI.h>
+#include <llvm/ADT/Hashing.h>
+
#include <cstddef>
#include <mutex>
#include <string_view>
@@ -70,7 +72,7 @@ class DeviceKernelInfo {
using CacheKeyT = std::pair<ContextImpl *, ol_device_handle_t>;
struct CacheKeyHash {
std::size_t operator()(const CacheKeyT &Key) const noexcept {
- return std::hash<ContextImpl *>{}(Key.first);
+ return llvm::hash_combine(Key.first, Key.second);
}
};
std::mutex MCacheMutex;
diff --git a/libsycl/src/detail/program_manager.cpp b/libsycl/src/detail/program_manager.cpp
index c839efab3a634..13030a4ab56d9 100644
--- a/libsycl/src/detail/program_manager.cpp
+++ b/libsycl/src/detail/program_manager.cpp
@@ -41,9 +41,8 @@ void ProgramAndKernelManager::releaseResources() {
if (std::shared_ptr<ContextImpl> Context = WeakContext.lock()) {
// Every DeviceKernelInfo below is about to be destroyed: make sure this
// context, if it outlives this call, does not keep pointers to them.
- for (auto &[Name, Info] : MDeviceKernelInfoMap) {
+ for (auto &[Name, Info] : MDeviceKernelInfoMap)
Context->forgetKernelInfoCache(&Info);
- }
Context->releaseAllPrograms();
}
}
@@ -152,9 +151,8 @@ void ProgramAndKernelManager::unregisterFatBin(const void *BinaryStart,
DeviceKernelInfo *Info = &KernelIt->second;
for (const std::weak_ptr<ContextImpl> &WeakContext :
MContextsWithPrograms) {
- if (std::shared_ptr<ContextImpl> Context = WeakContext.lock()) {
+ if (std::shared_ptr<ContextImpl> Context = WeakContext.lock())
Context->forgetKernelInfoCache(Info);
- }
}
// Clear kernel specific data by destroying its kernel info object.
MDeviceKernelInfoMap.erase(KernelIt);
@@ -184,9 +182,8 @@ ol_symbol_handle_t ProgramAndKernelManager::getOrCreateKernel(
assert(Context && "Context can't be nullptr");
if (ol_symbol_handle_t CachedKernel =
- KernelInfo.tryGetCachedKernel(Context.get(), Device.getOLHandle())) {
+ KernelInfo.tryGetCachedKernel(Context.get(), Device.getOLHandle()))
return CachedKernel;
- }
std::lock_guard<std::mutex> KernelGuard(MDataCollectionMutex);
More information about the llvm-commits
mailing list