From 9b74c13af4340c446919656ec9d430045d123ce7 Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Thu, 31 Jul 2025 06:55:30 -0700 Subject: [PATCH 01/12] [SYCL] Delete kernel name based info with the last image referencing it Prior to this patch, kernel name based info (e.g. kernel id or assert usage) was deleted whenever an image referencing it was removed. This is technically incorrect since multiple images can contain the same kernel name. --- .../program_manager/program_manager.cpp | 40 ++++++++++++++----- .../program_manager/program_manager.hpp | 6 +++ 2 files changed, 35 insertions(+), 11 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 6e40205692255..365ccba5b9a43 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -2042,6 +2042,9 @@ void ProgramManager::addImage(sycl_device_binary RawImg, } m_KernelIDs2BinImage.insert(std::make_pair(It->second, Img.get())); KernelIDs->push_back(It->second); + + // Keep track of image to kernel name reference count for cleanup. + m_BinImage2KernelNameRefCount[name]++; } cacheKernelUsesAssertInfo(*Img); @@ -2140,27 +2143,42 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { // Unmap the unique kernel IDs for the offload entries for (sycl_offload_entry EntriesIt = EntriesB; EntriesIt != EntriesE; EntriesIt = EntriesIt->Increment()) { - + const char *Name = EntriesIt->GetName(); // Drop entry for service kernel - if (std::strstr(EntriesIt->GetName(), "__sycl_service_kernel__")) { - m_ServiceKernels.erase(EntriesIt->GetName()); + if (std::strstr(Name, "__sycl_service_kernel__")) { + m_ServiceKernels.erase(Name); continue; } // Exported device functions won't have a kernel ID - if (m_ExportedSymbolImages.find(EntriesIt->GetName()) != - m_ExportedSymbolImages.end()) { + if (m_ExportedSymbolImages.find(Name) != m_ExportedSymbolImages.end()) { continue; } - // remove everything associated with this KernelName - m_KernelUsesAssert.erase(EntriesIt->GetName()); - m_KernelImplicitLocalArgPos.erase(EntriesIt->GetName()); + // Remove everything associated with this KernelName if this is the last + // image referencing it, otherwise remove just the ID -> Img mapping. + int &RefCount = m_BinImage2KernelNameRefCount[Name]; + assert(RefCount > 0); + + if (--RefCount == 0) { + m_KernelUsesAssert.erase(Name); + m_KernelImplicitLocalArgPos.erase(Name); + } - if (auto It = m_KernelName2KernelIDs.find(EntriesIt->GetName()); + if (auto It = m_KernelName2KernelIDs.find(Name); It != m_KernelName2KernelIDs.end()) { - m_KernelIDs2BinImage.erase(It->second); - m_KernelName2KernelIDs.erase(It); + if (RefCount == 0) { + m_KernelIDs2BinImage.erase(It->second); + m_KernelName2KernelIDs.erase(It); + } else { + auto ID2ImgIt = std::find_if( + m_KernelIDs2BinImage.begin(), m_KernelIDs2BinImage.end(), + [&](const auto &Pair) { + return Pair.first == It->second && Pair.second == Img; + }); + if (ID2ImgIt != m_KernelIDs2BinImage.end()) + m_KernelIDs2BinImage.erase(ID2ImgIt); + } } } diff --git a/sycl/source/detail/program_manager/program_manager.hpp b/sycl/source/detail/program_manager/program_manager.hpp index 87ce40d1a7e98..ab36f8f700f42 100644 --- a/sycl/source/detail/program_manager/program_manager.hpp +++ b/sycl/source/detail/program_manager/program_manager.hpp @@ -460,6 +460,12 @@ class ProgramManager { /// \ref Sync::getGlobalLock() while holding this mutex. std::mutex m_KernelIDsMutex; + /// Keeps track of binary image to kernel name reference count. + /// Used for checking if the last image referencing the kernel name + /// is removed in order to trigger cleanup of kernel name based information. + /// Access must be guarded by the m_KernelIDsMutex mutex. + std::unordered_map m_BinImage2KernelNameRefCount; + /// Caches all found service kernels to expedite future checks. A SYCL service /// kernel is a kernel that has not been defined by the user but is instead /// generated by the SYCL runtime. Service kernel name types must be declared From 0c50f1367eac915568c45dbc833875c20477f82c Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Mon, 4 Aug 2025 05:54:06 -0700 Subject: [PATCH 02/12] Improve m_KernelIDs2BinImage cleanup --- .../detail/program_manager/program_manager.cpp | 12 +++++++----- 1 file changed, 7 insertions(+), 5 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 365ccba5b9a43..a836c62cbce2e 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -2171,12 +2171,14 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { m_KernelIDs2BinImage.erase(It->second); m_KernelName2KernelIDs.erase(It); } else { - auto ID2ImgIt = std::find_if( - m_KernelIDs2BinImage.begin(), m_KernelIDs2BinImage.end(), - [&](const auto &Pair) { - return Pair.first == It->second && Pair.second == Img; + auto [RangeBegin, RangeEnd] = + m_KernelIDs2BinImage.equal_range(It->second); + + auto ID2ImgIt = + std::find_if(RangeBegin, RangeEnd, [&](const auto &Pair) { + return Pair.second == Img; }); - if (ID2ImgIt != m_KernelIDs2BinImage.end()) + if (ID2ImgIt != RangeEnd) m_KernelIDs2BinImage.erase(ID2ImgIt); } } From 338b1a416dddfb2198209fba2788585a8ed68468 Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Mon, 4 Aug 2025 07:09:44 -0700 Subject: [PATCH 03/12] Fix cleanup of exported symbol images map --- .../detail/program_manager/program_manager.cpp | 15 +++++++++++---- 1 file changed, 11 insertions(+), 4 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index a836c62cbce2e..f6b754e634d20 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -2178,8 +2178,8 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { std::find_if(RangeBegin, RangeEnd, [&](const auto &Pair) { return Pair.second == Img; }); - if (ID2ImgIt != RangeEnd) - m_KernelIDs2BinImage.erase(ID2ImgIt); + assert(ID2ImgIt != RangeEnd); + m_KernelIDs2BinImage.erase(ID2ImgIt); } } } @@ -2187,10 +2187,17 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { // Drop reverse mapping m_BinImg2KernelIDs.erase(Img); - // Unregister exported symbols (needs to happen after the ID unmap loop) + // Unregister exported symbol -> Img pair (needs to happen after the ID + // unmap loop) for (const sycl_device_binary_property &ESProp : Img->getExportedSymbols()) { - m_ExportedSymbolImages.erase(ESProp->Name); + auto [RangeBegin, RangeEnd] = + m_ExportedSymbolImages.equal_range(ESProp->Name); + auto It = std::find_if(RangeBegin, RangeEnd, [&](const auto &Pair) { + return Pair.second == Img; + }); + assert(It != RangeEnd); + m_ExportedSymbolImages.erase(It); } for (const sycl_device_binary_property &VFProp : From 2c2a0732c7b1904e3da981630025bc196877c76c Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Mon, 4 Aug 2025 08:37:55 -0700 Subject: [PATCH 04/12] Check presence in ID->Img map instead of asserting it --- sycl/source/detail/program_manager/program_manager.cpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index f6b754e634d20..875c9b17893fd 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -2178,8 +2178,8 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { std::find_if(RangeBegin, RangeEnd, [&](const auto &Pair) { return Pair.second == Img; }); - assert(ID2ImgIt != RangeEnd); - m_KernelIDs2BinImage.erase(ID2ImgIt); + if (ID2ImgIt != RangeEnd) + m_KernelIDs2BinImage.erase(ID2ImgIt); } } } From 151b5cab05cf4c2ae4bb50baa859d7cc1b21706d Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Tue, 5 Aug 2025 06:57:16 -0700 Subject: [PATCH 05/12] Fix the if vs assert part --- sycl/source/detail/program_manager/program_manager.cpp | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 875c9b17893fd..d8b7172514998 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -2178,8 +2178,8 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { std::find_if(RangeBegin, RangeEnd, [&](const auto &Pair) { return Pair.second == Img; }); - if (ID2ImgIt != RangeEnd) - m_KernelIDs2BinImage.erase(ID2ImgIt); + assert(ID2ImgIt != RangeEnd); + m_KernelIDs2BinImage.erase(ID2ImgIt); } } } @@ -2196,8 +2196,8 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { auto It = std::find_if(RangeBegin, RangeEnd, [&](const auto &Pair) { return Pair.second == Img; }); - assert(It != RangeEnd); - m_ExportedSymbolImages.erase(It); + if (It != RangeEnd) + m_ExportedSymbolImages.erase(It); } for (const sycl_device_binary_property &VFProp : From dcb478896200e1e3b70b89bf5c3032cc7325adac Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Tue, 5 Aug 2025 16:46:54 -0700 Subject: [PATCH 06/12] Debug kernel_lin.cpp test --- sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp | 1 + 1 file changed, 1 insertion(+) diff --git a/sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp b/sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp index d93a7880d404e..9d9f7be5fee57 100644 --- a/sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp +++ b/sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp @@ -4,4 +4,5 @@ // RUN: %clangxx -fPIC -shared -fsycl %{sycl_target_opts} %S/Inputs/kernel_module.cpp -o %t.so // RUN: %clangxx -DTEST_SHARED_LIB='"%t.so"' -ldl -fsycl %{sycl_target_opts} %S/Inputs/kernel_function.cpp -o %t.out +// RUN: %clangxx -DTEST_SHARED_LIB='"kernel_lin.cpp.tmp.so"' -ldl -fsycl %{sycl_target_opts} %S/Inputs/kernel_function.cpp -o %t2.out // RUN: %{run} %t.out From 23166af58f0ba7f89681460b9c8a06affb87c5ff Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Wed, 6 Aug 2025 08:11:22 -0700 Subject: [PATCH 07/12] Revert "Debug kernel_lin.cpp test" This reverts commit dcb478896200e1e3b70b89bf5c3032cc7325adac. --- sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp | 1 - 1 file changed, 1 deletion(-) diff --git a/sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp b/sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp index 9d9f7be5fee57..d93a7880d404e 100644 --- a/sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp +++ b/sycl/test-e2e/syclcompat/kernel/kernel_lin.cpp @@ -4,5 +4,4 @@ // RUN: %clangxx -fPIC -shared -fsycl %{sycl_target_opts} %S/Inputs/kernel_module.cpp -o %t.so // RUN: %clangxx -DTEST_SHARED_LIB='"%t.so"' -ldl -fsycl %{sycl_target_opts} %S/Inputs/kernel_function.cpp -o %t.out -// RUN: %clangxx -DTEST_SHARED_LIB='"kernel_lin.cpp.tmp.so"' -ldl -fsycl %{sycl_target_opts} %S/Inputs/kernel_function.cpp -o %t2.out // RUN: %{run} %t.out From 372a27185e187a446f7d1e7eb465e066e23955e8 Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Wed, 6 Aug 2025 08:14:33 -0700 Subject: [PATCH 08/12] Refcount cleanup --- .../detail/program_manager/program_manager.cpp | 18 +++++++++++------- .../detail/program_manager/program_manager.hpp | 2 +- 2 files changed, 12 insertions(+), 8 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index d8b7172514998..22dfd6501166e 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -2044,7 +2044,7 @@ void ProgramManager::addImage(sycl_device_binary RawImg, KernelIDs->push_back(It->second); // Keep track of image to kernel name reference count for cleanup. - m_BinImage2KernelNameRefCount[name]++; + m_KernelNameRefCount[name]++; } cacheKernelUsesAssertInfo(*Img); @@ -2157,13 +2157,11 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { // Remove everything associated with this KernelName if this is the last // image referencing it, otherwise remove just the ID -> Img mapping. - int &RefCount = m_BinImage2KernelNameRefCount[Name]; + auto RefCountIt = m_KernelNameRefCount.find(Name); + assert(RefCountIt != m_KernelNameRefCount.end()); + int &RefCount = RefCountIt->second; assert(RefCount > 0); - - if (--RefCount == 0) { - m_KernelUsesAssert.erase(Name); - m_KernelImplicitLocalArgPos.erase(Name); - } + --RefCount; if (auto It = m_KernelName2KernelIDs.find(Name); It != m_KernelName2KernelIDs.end()) { @@ -2182,6 +2180,12 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { m_KernelIDs2BinImage.erase(ID2ImgIt); } } + + if (RefCount == 0) { + m_KernelUsesAssert.erase(Name); + m_KernelImplicitLocalArgPos.erase(Name); + m_KernelNameRefCount.erase(RefCountIt); + } } // Drop reverse mapping diff --git a/sycl/source/detail/program_manager/program_manager.hpp b/sycl/source/detail/program_manager/program_manager.hpp index aa9f573ebc64b..eaea458f95e84 100644 --- a/sycl/source/detail/program_manager/program_manager.hpp +++ b/sycl/source/detail/program_manager/program_manager.hpp @@ -463,7 +463,7 @@ class ProgramManager { /// Used for checking if the last image referencing the kernel name /// is removed in order to trigger cleanup of kernel name based information. /// Access must be guarded by the m_KernelIDsMutex mutex. - std::unordered_map m_BinImage2KernelNameRefCount; + std::unordered_map m_KernelNameRefCount; /// Caches all found service kernels to expedite future checks. A SYCL service /// kernel is a kernel that has not been defined by the user but is instead From bd43afccb64eb32a9fffc546faa3ee69cfa82188 Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Wed, 6 Aug 2025 08:43:35 -0700 Subject: [PATCH 09/12] Extend cleanup unit tests --- sycl/unittests/program_manager/Cleanup.cpp | 15 +++++++++++++++ 1 file changed, 15 insertions(+) diff --git a/sycl/unittests/program_manager/Cleanup.cpp b/sycl/unittests/program_manager/Cleanup.cpp index 47be8b7d05645..ff9ce5552db69 100644 --- a/sycl/unittests/program_manager/Cleanup.cpp +++ b/sycl/unittests/program_manager/Cleanup.cpp @@ -61,6 +61,11 @@ class ProgramManagerExposed : public sycl::detail::ProgramManager { return NativePrograms; } + std::unordered_map & + getKernelNameRefCount() { + return m_KernelNameRefCount; + } + std::unordered_map> & @@ -291,6 +296,16 @@ void checkAllInvolvedContainers(ProgramManagerExposed &PM, size_t ExpectedCount, << Comment; } + { + EXPECT_EQ(PM.getKernelNameRefCount().size(), ExpectedCount) << Comment; + EXPECT_TRUE( + PM.getKernelNameRefCount().count(generateRefName("A", "Kernel")) > 0) + << Comment; + EXPECT_TRUE( + PM.getKernelNameRefCount().count(generateRefName("B", "Kernel")) > 0) + << Comment; + } + EXPECT_EQ(PM.getEliminatedKernelArgMask().size(), ExpectedCount) << Comment; { EXPECT_EQ(PM.getKernelUsesAssert().size(), ExpectedCount) << Comment; From b1f9365ff2add07ffac9f90612ffa9640476ebdd Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Fri, 8 Aug 2025 07:58:04 -0700 Subject: [PATCH 10/12] Add a unit test --- .../program_manager/program_manager.cpp | 45 ++-- sycl/unittests/program_manager/Cleanup.cpp | 200 ++++++++++-------- 2 files changed, 141 insertions(+), 104 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 22dfd6501166e..27764b73c2821 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -2118,6 +2118,18 @@ void ProgramManager::addImages(sycl_device_binaries DeviceBinary) { addImage(&(DeviceBinary->DeviceBinaries[I])); } +template +void removeFromMultimap(MultimapT &Map, const KeyT &Key, const ValT &Val, + bool AssertContains = true) { + auto [RangeBegin, RangeEnd] = Map.equal_range(Key); + auto It = std::find_if(RangeBegin, RangeEnd, + [&](const auto &Pair) { return Pair.second == Val; }); + if (!AssertContains && It == RangeEnd) + return; + assert(It != RangeEnd); + Map.erase(It); +} + void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { if (DeviceBinary->NumDeviceBinaries == 0) return; @@ -2146,7 +2158,7 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { const char *Name = EntriesIt->GetName(); // Drop entry for service kernel if (std::strstr(Name, "__sycl_service_kernel__")) { - m_ServiceKernels.erase(Name); + removeFromMultimap(m_ServiceKernels, Name, Img); continue; } @@ -2169,15 +2181,7 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { m_KernelIDs2BinImage.erase(It->second); m_KernelName2KernelIDs.erase(It); } else { - auto [RangeBegin, RangeEnd] = - m_KernelIDs2BinImage.equal_range(It->second); - - auto ID2ImgIt = - std::find_if(RangeBegin, RangeEnd, [&](const auto &Pair) { - return Pair.second == Img; - }); - assert(ID2ImgIt != RangeEnd); - m_KernelIDs2BinImage.erase(ID2ImgIt); + removeFromMultimap(m_KernelIDs2BinImage, It->second, Img); } } @@ -2195,20 +2199,23 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { // unmap loop) for (const sycl_device_binary_property &ESProp : Img->getExportedSymbols()) { - auto [RangeBegin, RangeEnd] = - m_ExportedSymbolImages.equal_range(ESProp->Name); - auto It = std::find_if(RangeBegin, RangeEnd, [&](const auto &Pair) { - return Pair.second == Img; - }); - if (It != RangeEnd) - m_ExportedSymbolImages.erase(It); + removeFromMultimap(m_ExportedSymbolImages, ESProp->Name, Img, + /*AssertContains*/ false); } for (const sycl_device_binary_property &VFProp : Img->getVirtualFunctions()) { std::string StrValue = DeviceBinaryProperty(VFProp).asCString(); - for (const auto &SetName : detail::split_string(StrValue, ',')) - m_VFSet2BinImage.erase(SetName); + for (const auto &SetName : detail::split_string(StrValue, ',')) { + auto It = m_VFSet2BinImage.find(SetName); + assert(It != m_VFSet2BinImage.end()); + auto &ImgSet = It->second; + auto ImgIt = ImgSet.find(Img); + assert(ImgIt != ImgSet.end()); + ImgSet.erase(ImgIt); + if (ImgSet.empty()) + m_VFSet2BinImage.erase(It); + } } m_DeviceGlobals.eraseEntries(Img); diff --git a/sycl/unittests/program_manager/Cleanup.cpp b/sycl/unittests/program_manager/Cleanup.cpp index ff9ce5552db69..4a39f06ee5250 100644 --- a/sycl/unittests/program_manager/Cleanup.cpp +++ b/sycl/unittests/program_manager/Cleanup.cpp @@ -137,6 +137,16 @@ std::string generateRefName(const std::string &ImageId, return FeatureName + "_" + ImageId; } +std::vector +generateRefNames(const std::vector &ImageIds, + const std::string &FeatureName) { + std::vector RefNames; + RefNames.reserve(ImageIds.size()); + for (const std::string &ImageId : ImageIds) + RefNames.push_back(generateRefName(ImageId, FeatureName)); + return RefNames; +} + sycl::ext::oneapi::experimental::device_global DeviceGlobalA; sycl::ext::oneapi::experimental::device_global DeviceGlobalB; sycl::ext::oneapi::experimental::device_global DeviceGlobalC; @@ -148,7 +158,8 @@ using PipeA = sycl::ext::intel::experimental::pipe; using PipeB = sycl::ext::intel::experimental::pipe; using PipeC = sycl::ext::intel::experimental::pipe; -sycl::unittest::MockDeviceImage generateImage(const std::string &ImageId) { +sycl::unittest::MockDeviceImage generateImage(const std::string &ImageId, + bool AddHostPipes = true) { sycl::unittest::MockPropertySet PropSet; std::initializer_list KernelNames{ @@ -186,11 +197,11 @@ sycl::unittest::MockDeviceImage generateImage(const std::string &ImageId) { std::vector{ sycl::unittest::makeDeviceGlobalInfo( generateRefName(ImageId, "DeviceGlobal"), sizeof(int), 0)}); - - PropSet.insert(__SYCL_PROPERTY_SET_SYCL_HOST_PIPES, - std::vector{ - sycl::unittest::makeHostPipeInfo( - generateRefName(ImageId, "HostPipe"), sizeof(int))}); + if (AddHostPipes) + PropSet.insert(__SYCL_PROPERTY_SET_SYCL_HOST_PIPES, + std::vector{ + sycl::unittest::makeHostPipeInfo( + generateRefName(ImageId, "HostPipe"), sizeof(int))}); std::vector Bin{0}; std::vector Entries = @@ -234,6 +245,11 @@ static std::array ImagesToKeep = { static std::array ImagesToRemove = { generateImage("C")}; +static std::array ImagesToKeepSameEntries = + {generateImage("A", /*AddHostPipe*/ false)}; +static std::array + ImagesToRemoveSameEntries = {generateImage("A", /*AddHostPipe*/ false)}; + static std::array ImagesToKeepKernelOnly = { generateImageKernelOnly("A"), generateImageKernelOnly("B")}; static std::array ImagesToRemoveKernelOnly = @@ -256,86 +272,75 @@ void convertAndAddImages( PM.addImages(&AllBinaries); } -void checkAllInvolvedContainers(ProgramManagerExposed &PM, size_t ExpectedCount, - const std::string &Comment) { - EXPECT_EQ(PM.getKernelID2BinImage().size(), ExpectedCount) << Comment; - { - EXPECT_EQ(PM.getKernelName2KernelID().size(), ExpectedCount) << Comment; - EXPECT_TRUE( - PM.getKernelName2KernelID().count(generateRefName("A", "Kernel")) > 0) - << Comment; - EXPECT_TRUE( - PM.getKernelName2KernelID().count(generateRefName("B", "Kernel")) > 0) - << Comment; - } - EXPECT_EQ(PM.getBinImage2KernelId().size(), ExpectedCount) << Comment; - { - EXPECT_EQ(PM.getServiceKernels().size(), ExpectedCount) << Comment; - EXPECT_TRUE(PM.getServiceKernels().count( - generateRefName("A", "__sycl_service_kernel__")) > 0) - << Comment; - EXPECT_TRUE(PM.getServiceKernels().count( - generateRefName("B", "__sycl_service_kernel__")) > 0) - << Comment; - } - { - EXPECT_EQ(PM.getExportedSymbolImages().size(), ExpectedCount) << Comment; - EXPECT_TRUE(PM.getExportedSymbolImages().count( - generateRefName("A", "Exported")) > 0) - << Comment; - EXPECT_TRUE(PM.getExportedSymbolImages().count( - generateRefName("B", "Exported")) > 0) - << Comment; - } - EXPECT_EQ(PM.getDeviceImages().size(), ExpectedCount) << Comment; - { - EXPECT_EQ(PM.getVFSet2BinImage().size(), ExpectedCount) << Comment; - EXPECT_TRUE(PM.getVFSet2BinImage().count(generateRefName("A", "VF")) > 0) - << Comment; - EXPECT_TRUE(PM.getVFSet2BinImage().count(generateRefName("B", "VF")) > 0) - << Comment; - } - - { - EXPECT_EQ(PM.getKernelNameRefCount().size(), ExpectedCount) << Comment; - EXPECT_TRUE( - PM.getKernelNameRefCount().count(generateRefName("A", "Kernel")) > 0) - << Comment; - EXPECT_TRUE( - PM.getKernelNameRefCount().count(generateRefName("B", "Kernel")) > 0) - << Comment; +template +void checkContainer(const T &Container, size_t ExpectedCount, + const std::vector &ExpectedEntries, + const std::string &Comment) { + EXPECT_EQ(Container.size(), ExpectedCount) << Comment; + for (const std::string &Entry : ExpectedEntries) { + EXPECT_TRUE(Container.count(Entry) > 0) << Comment; } +} - EXPECT_EQ(PM.getEliminatedKernelArgMask().size(), ExpectedCount) << Comment; - { - EXPECT_EQ(PM.getKernelUsesAssert().size(), ExpectedCount) << Comment; - EXPECT_TRUE(PM.getKernelUsesAssert().count(generateRefName("A", "Kernel")) > - 0) - << Comment; - EXPECT_TRUE(PM.getKernelUsesAssert().count(generateRefName("B", "Kernel")) > - 0) - << Comment; - } - EXPECT_EQ(PM.getKernelImplicitLocalArgPos().size(), ExpectedCount) << Comment; - - { - sycl::detail::DeviceGlobalMap &DeviceGlobalMap = PM.getDeviceGlobals(); - EXPECT_EQ(DeviceGlobalMap.size(), ExpectedCount) << Comment; - EXPECT_TRUE(DeviceGlobalMap.count(generateRefName("A", "DeviceGlobal")) > 0) - << Comment; - EXPECT_TRUE(DeviceGlobalMap.count(generateRefName("B", "DeviceGlobal")) > 0) - << Comment; - EXPECT_EQ(DeviceGlobalMap.getPointerMap().size(), ExpectedCount) << Comment; +void checkAllInvolvedContainers(ProgramManagerExposed &PM, + size_t ExpectedImgCount, + size_t ExpectedEntryCount, + const std::vector &ImgIds, + const std::string &CommentPostfix, + bool MultipleImgsPerEntryTestCase = false) { + EXPECT_EQ(PM.getKernelID2BinImage().size(), ExpectedImgCount) + << "KernelID2BinImg " + CommentPostfix; + checkContainer(PM.getKernelName2KernelID(), ExpectedEntryCount, + generateRefNames(ImgIds, "Kernel"), + "KernelName2KernelID " + CommentPostfix); + EXPECT_EQ(PM.getBinImage2KernelId().size(), ExpectedImgCount) + << CommentPostfix; + checkContainer(PM.getServiceKernels(), ExpectedImgCount, + generateRefNames(ImgIds, "__sycl_service_kernel__"), + "Service kernels " + CommentPostfix); + checkContainer(PM.getExportedSymbolImages(), ExpectedImgCount, + generateRefNames(ImgIds, "Exported"), + "Exported symbol images " + CommentPostfix); + EXPECT_EQ(PM.getDeviceImages().size(), ExpectedImgCount) + << "Device images " + CommentPostfix; + + checkContainer(PM.getVFSet2BinImage(), ExpectedEntryCount, + generateRefNames(ImgIds, "VF"), + "VFSet2BinImage " + CommentPostfix); + checkContainer(PM.getKernelNameRefCount(), ExpectedEntryCount, + generateRefNames(ImgIds, "Kernel"), + "Kernel name reference count " + CommentPostfix); + EXPECT_EQ(PM.getEliminatedKernelArgMask().size(), ExpectedImgCount) + << "Eliminated kernel arg mask " + CommentPostfix; + checkContainer(PM.getKernelUsesAssert(), ExpectedEntryCount, + generateRefNames(ImgIds, "Kernel"), + "KernelUsesAssert " + CommentPostfix); + EXPECT_EQ(PM.getKernelImplicitLocalArgPos().size(), ExpectedEntryCount) + << "Kernel implicit local arg pos " + CommentPostfix; + + if (!MultipleImgsPerEntryTestCase) { + // FIXME expected to fail for now, device globals cleanup seems to be + // purging all info for symbols associated with the removed image. + checkContainer(PM.getDeviceGlobals(), ExpectedEntryCount, + generateRefNames(ImgIds, "DeviceGlobal"), + "Device globals " + CommentPostfix); + + // The test case with the same entries in multiple images doesn't support + // host pipes since those are assumed to be unique. + checkContainer(PM.getHostPipes(), ExpectedEntryCount, + generateRefNames(ImgIds, "HostPipe"), + "Host pipes " + CommentPostfix); + EXPECT_EQ(PM.getPtrToHostPipe().size(), ExpectedEntryCount) + << "Pointer to host pipe " + CommentPostfix; } +} - { - EXPECT_EQ(PM.getHostPipes().size(), ExpectedCount) << Comment; - EXPECT_TRUE(PM.getHostPipes().count(generateRefName("A", "HostPipe")) > 0) - << Comment; - EXPECT_TRUE(PM.getHostPipes().count(generateRefName("B", "HostPipe")) > 0) - << Comment; - } - EXPECT_EQ(PM.getPtrToHostPipe().size(), ExpectedCount) << Comment; +void checkAllInvolvedContainers(ProgramManagerExposed &PM, size_t ExpectedCount, + const std::vector &ImgIds, + const std::string &CommentPostfix, + bool CheckHostPipes = false) { + checkAllInvolvedContainers(PM, ExpectedCount, ExpectedCount, ImgIds, + CommentPostfix, CheckHostPipes); } TEST(ImageRemoval, BaseContainers) { @@ -363,12 +368,37 @@ TEST(ImageRemoval, BaseContainers) { generateRefName("C", "HostPipe").c_str()); checkAllInvolvedContainers(PM, ImagesToRemove.size() + ImagesToKeep.size(), - "Check failed before removal"); + {"A", "B", "C"}, "check failed before removal"); + + PM.removeImages(&TestBinaries); + + checkAllInvolvedContainers(PM, ImagesToKeep.size(), {"A", "B"}, + "check failed after removal"); +} + +TEST(ImageRemoval, MultipleImagesPerEntry) { + ProgramManagerExposed PM; + + sycl_device_binary_struct NativeImages[ImagesToKeepSameEntries.size()]; + sycl_device_binaries_struct AllBinaries; + convertAndAddImages(PM, ImagesToKeepSameEntries, NativeImages, AllBinaries); + + sycl_device_binary_struct + NativeImagesForRemoval[ImagesToRemoveSameEntries.size()]; + sycl_device_binaries_struct TestBinaries; + convertAndAddImages(PM, ImagesToRemoveSameEntries, NativeImagesForRemoval, + TestBinaries); + + checkAllInvolvedContainers( + PM, ImagesToRemoveSameEntries.size() + ImagesToKeepSameEntries.size(), + /*ExpectedEntryCount*/ 1, {"A"}, "check failed before removal", + /*MultipleImgsPerEntryTestCase*/ true); PM.removeImages(&TestBinaries); - checkAllInvolvedContainers(PM, ImagesToKeep.size(), - "Check failed after removal"); + checkAllInvolvedContainers(PM, ImagesToKeepSameEntries.size(), {"A"}, + "check failed after removal", + /*MultipleImgsPerEntryTestCase*/ true); } TEST(ImageRemoval, NativePrograms) { From 92bd78bf329c72020fbb85fb53896d3525b3752d Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Tue, 12 Aug 2025 08:26:30 -0700 Subject: [PATCH 11/12] Apply comments --- .../program_manager/program_manager.cpp | 44 +++++++++---------- 1 file changed, 22 insertions(+), 22 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index f7e2f8477b583..c7979feea10bd 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -2119,8 +2119,8 @@ void ProgramManager::addImages(sycl_device_binaries DeviceBinary) { } template -void removeFromMultimap(MultimapT &Map, const KeyT &Key, const ValT &Val, - bool AssertContains = true) { +void removeFromMultimapByVal(MultimapT &Map, const KeyT &Key, const ValT &Val, + bool AssertContains = true) { auto [RangeBegin, RangeEnd] = Map.equal_range(Key); auto It = std::find_if(RangeBegin, RangeEnd, [&](const auto &Pair) { return Pair.second == Val; }); @@ -2158,7 +2158,7 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { const char *Name = EntriesIt->GetName(); // Drop entry for service kernel if (std::strstr(Name, "__sycl_service_kernel__")) { - removeFromMultimap(m_ServiceKernels, Name, Img); + removeFromMultimapByVal(m_ServiceKernels, Name, Img); continue; } @@ -2167,28 +2167,25 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { continue; } - // Remove everything associated with this KernelName if this is the last - // image referencing it, otherwise remove just the ID -> Img mapping. + auto Name2IDIt = m_KernelName2KernelIDs.find(Name); + if (Name2IDIt != m_KernelName2KernelIDs.end()) + removeFromMultimapByVal(m_KernelIDs2BinImage, Name2IDIt->second, Img); + auto RefCountIt = m_KernelNameRefCount.find(Name); assert(RefCountIt != m_KernelNameRefCount.end()); int &RefCount = RefCountIt->second; assert(RefCount > 0); - --RefCount; - - if (auto It = m_KernelName2KernelIDs.find(Name); - It != m_KernelName2KernelIDs.end()) { - if (RefCount == 0) { - m_KernelIDs2BinImage.erase(It->second); - m_KernelName2KernelIDs.erase(It); - } else { - removeFromMultimap(m_KernelIDs2BinImage, It->second, Img); - } - } - if (RefCount == 0) { + // Remove everything associated with this KernelName if this is the last + // image referencing it. + if (--RefCount == 0) { + // TODO aggregate all these maps into a single one since their entries + // share lifetime. m_KernelUsesAssert.erase(Name); m_KernelImplicitLocalArgPos.erase(Name); m_KernelNameRefCount.erase(RefCountIt); + if (Name2IDIt != m_KernelName2KernelIDs.end()) + m_KernelName2KernelIDs.erase(Name2IDIt); } } @@ -2199,20 +2196,23 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { // unmap loop) for (const sycl_device_binary_property &ESProp : Img->getExportedSymbols()) { - removeFromMultimap(m_ExportedSymbolImages, ESProp->Name, Img, - /*AssertContains*/ false); + removeFromMultimapByVal(m_ExportedSymbolImages, ESProp->Name, Img, + /*AssertContains*/ false); } for (const sycl_device_binary_property &VFProp : Img->getVirtualFunctions()) { std::string StrValue = DeviceBinaryProperty(VFProp).asCString(); - for (const auto &SetName : detail::split_string(StrValue, ',')) { - auto It = m_VFSet2BinImage.find(SetName); + // Unregister the image from all referenced virtual function sets. + for (const auto &VFSetName : detail::split_string(StrValue, ',')) { + auto It = m_VFSet2BinImage.find(VFSetName); assert(It != m_VFSet2BinImage.end()); - auto &ImgSet = It->second; + std::set &ImgSet = It->second; auto ImgIt = ImgSet.find(Img); assert(ImgIt != ImgSet.end()); ImgSet.erase(ImgIt); + // If no images referencing this virtual function set remain, drop + // it from the map. if (ImgSet.empty()) m_VFSet2BinImage.erase(It); } From 2370609810c42894526c5fe16015f0aa374b4149 Mon Sep 17 00:00:00 2001 From: Sergey Semenov Date: Tue, 12 Aug 2025 10:08:12 -0700 Subject: [PATCH 12/12] const char * -> string_view --- sycl/source/detail/program_manager/program_manager.cpp | 7 ++++--- 1 file changed, 4 insertions(+), 3 deletions(-) diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index c7979feea10bd..77f28a5131f8a 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -2155,15 +2155,16 @@ void ProgramManager::removeImages(sycl_device_binaries DeviceBinary) { // Unmap the unique kernel IDs for the offload entries for (sycl_offload_entry EntriesIt = EntriesB; EntriesIt != EntriesE; EntriesIt = EntriesIt->Increment()) { - const char *Name = EntriesIt->GetName(); + detail::KernelNameStrT Name = EntriesIt->GetName(); // Drop entry for service kernel - if (std::strstr(Name, "__sycl_service_kernel__")) { + if (Name.find("__sycl_service_kernel__") != std::string::npos) { removeFromMultimapByVal(m_ServiceKernels, Name, Img); continue; } // Exported device functions won't have a kernel ID - if (m_ExportedSymbolImages.find(Name) != m_ExportedSymbolImages.end()) { + if (m_ExportedSymbolImages.find(std::string(Name)) != + m_ExportedSymbolImages.end()) { continue; }