Skip to content

Commit 1fe4d83

Browse files
uditagarwal97uditagarwal97
committed
[Clang][SYCL] Discard pointers to free function descriptors when library is unloaded
1 parent b8f1bed commit 1fe4d83

File tree

10 files changed

+50
-5
lines changed

10 files changed

+50
-5
lines changed

‎clang/lib/Sema/SemaSYCL.cpp‎

Lines changed: 5 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -7313,6 +7313,11 @@ void SYCLIntegrationHeader::emit(raw_ostream &O) {
73137313
<< "sycl::detail::kernel_names, sycl::detail::kernel_args_sizes, "
73147314
<< KernelDescs.size() << ");\n";
73157315
O << " }\n";
7316+
O << " ~GlobalMapUpdater() {\n";
7317+
O << " sycl::detail::free_function_info_map::remove("
7318+
<< "sycl::detail::kernel_names, sycl::detail::kernel_args_sizes, "
7319+
<< KernelDescs.size() << ");\n";
7320+
O << " }\n";
73167321
O << "};\n";
73177322
O << "static GlobalMapUpdater updater;\n";
73187323
O << "} // namespace detail\n";

‎clang/test/CodeGenSYCL/free_function_default_template_arguments.cpp‎

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -1094,6 +1094,9 @@ namespace Testing::Tests {
10941094
// CHECK-NEXT: GlobalMapUpdater() {
10951095
// CHECK-NEXT: sycl::detail::free_function_info_map::add(sycl::detail::kernel_names, sycl::detail::kernel_args_sizes, 28);
10961096
// CHECK-NEXT: }
1097+
// CHECK-NEXT: ~GlobalMapUpdater() {
1098+
// CHECK-NEXT: sycl::detail::free_function_info_map::remove(sycl::detail::kernel_names, sycl::detail::kernel_args_sizes, 28);
1099+
// CHECK-NEXT: }
10971100
// CHECK-NEXT: };
10981101
// CHECK-NEXT: static GlobalMapUpdater updater;
10991102
// CHECK-NEXT: } // namespace detail

‎clang/test/CodeGenSYCL/free_function_int_header.cpp‎

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -1574,6 +1574,9 @@ void ff_24(int arg) {
15741574
// CHECK-NEXT: GlobalMapUpdater() {
15751575
// CHECK-NEXT: sycl::detail::free_function_info_map::add(sycl::detail::kernel_names, sycl::detail::kernel_args_sizes, 33);
15761576
// CHECK-NEXT: }
1577+
// CHECK-NEXT: ~GlobalMapUpdater() {
1578+
// CHECK-NEXT: sycl::detail::free_function_info_map::remove(sycl::detail::kernel_names, sycl::detail::kernel_args_sizes, 33);
1579+
// CHECK-NEXT: }
15771580
// CHECK-NEXT: };
15781581
// CHECK-NEXT: static GlobalMapUpdater updater;
15791582
// CHECK-NEXT: } // namespace detail

‎clang/test/CodeGenSYCL/free_function_int_header_rtc_mode.cpp‎

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -82,6 +82,9 @@ int main(){
8282
// CHECK-NORTC-NEXT: GlobalMapUpdater() {
8383
// CHECK-NORTC-NEXT: sycl::detail::free_function_info_map::add(sycl::detail::kernel_names, sycl::detail::kernel_args_sizes, 3);
8484
// CHECK-NORTC-NEXT: }
85+
// CHECK-NORTC-NEXT: ~GlobalMapUpdater() {
86+
// CHECK-NORTC-NEXT: sycl::detail::free_function_info_map::remove(sycl::detail::kernel_names, sycl::detail::kernel_args_sizes, 3);
87+
// CHECK-NORTC-NEXT: }
8588
// CHECK-NORTC-NEXT: };
8689
// CHECK-NORTC-NEXT: static GlobalMapUpdater updater;
8790
// CHECK-NORTC-NEXT: } // namespace detail

‎clang/test/SemaSYCL/Inputs/sycl/detail/kernel_global_info.hpp‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -16,6 +16,7 @@ namespace detail {
1616
namespace free_function_info_map {
1717

1818
__SYCL_EXPORT void add(const void *DeviceGlobalPtr, const char *UniqueId);
19+
__SYCL_EXPORT void remove(const void *DeviceGlobalPtr, const char *UniqueId);
1920

2021
} // namespace free_function_info_map
2122
} // namespace detail

‎sycl/include/sycl/detail/kernel_global_info.hpp‎

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -18,6 +18,8 @@ namespace free_function_info_map {
1818
__SYCL_EXPORT void add(const char *const *UniqueId,
1919
const unsigned *DeviceGlobalPtr, unsigned Size);
2020

21+
__SYCL_EXPORT void remove(const char *const *UniqueId,
22+
const unsigned *DeviceGlobalPtr, unsigned Size);
2123
} // namespace free_function_info_map
2224
} // namespace detail
2325
} // namespace _V1

‎sycl/source/detail/kernel_global_info.cpp‎

Lines changed: 11 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -23,6 +23,17 @@ __SYCL_EXPORT void add(const char *const *FreeFunctionNames,
2323
std::move(GlobalInfoToCopy));
2424
}
2525

26+
__SYCL_EXPORT void remove(const char *const *FreeFunctionNames,
27+
const unsigned *FreeFunctionNumArgs, unsigned Size) {
28+
std::unordered_map<std::string_view, unsigned> GlobalInfoToCopy;
29+
for (size_t i = 0; i < Size; ++i) {
30+
GlobalInfoToCopy[std::string_view{FreeFunctionNames[i]}] =
31+
FreeFunctionNumArgs[i];
32+
}
33+
detail::ProgramManager::getInstance().unRegisterKernelGlobalInfo(
34+
std::move(GlobalInfoToCopy));
35+
}
36+
2637
} // namespace detail::free_function_info_map
2738
} // namespace _V1
2839
} // namespace sycl

‎sycl/source/detail/program_manager/program_manager.cpp‎

Lines changed: 12 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -2417,12 +2417,22 @@ void ProgramManager::registerKernelGlobalInfo(
24172417
if (m_FreeFunctionKernelGlobalInfo.empty())
24182418
m_FreeFunctionKernelGlobalInfo = std::move(GlobalInfoToCopy);
24192419
else {
2420-
for (auto &GlobalInfo : GlobalInfoToCopy) {
2420+
for (auto &GlobalInfo : GlobalInfoToCopy)
24212421
m_FreeFunctionKernelGlobalInfo.insert(GlobalInfo);
2422-
}
24232422
}
24242423
}
24252424

2425+
// Remove entries from m_FreeFunctionKernelGlobalInfo that matches
2426+
// the ones in GlobalInfoToCopy. This function is called when a shared
2427+
// library consisting of SYCL kernels is unloaded.
2428+
void ProgramManager::unRegisterKernelGlobalInfo(
2429+
std::unordered_map<std::string_view, unsigned> &&GlobalInfoToCopy) {
2430+
std::lock_guard<std::mutex> Guard(MNativeProgramsMutex);
2431+
2432+
for (const auto &GlobalInfo : GlobalInfoToCopy)
2433+
m_FreeFunctionKernelGlobalInfo.erase(GlobalInfo.first);
2434+
}
2435+
24262436
std::optional<unsigned>
24272437
ProgramManager::getKernelGlobalInfoDesc(const char *UniqueId) {
24282438
std::lock_guard<std::mutex> Guard(MNativeProgramsMutex);

‎sycl/source/detail/program_manager/program_manager.hpp‎

Lines changed: 6 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -259,6 +259,12 @@ class ProgramManager {
259259
void registerKernelGlobalInfo(
260260
std::unordered_map<std::string_view, unsigned> &&GlobalInfoToCopy);
261261

262+
// The function removes kernel global descriptors from the
263+
// kernel global map when a shared library consisting o SYCL kernels
264+
// is unloaded.
265+
void unRegisterKernelGlobalInfo(
266+
std::unordered_map<std::string_view, unsigned> &&GlobalInfoToCopy);
267+
262268
// The function returns a pointer to the kernel global desc identified by
263269
// the unique ID from the kernel global map.
264270
std::optional<unsigned> getKernelGlobalInfoDesc(const char *UniqueId);

‎sycl/test/abi/sycl_symbols_linux.dump‎

Lines changed: 4 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -2985,12 +2985,12 @@ _ZN4sycl3_V121__isgreaterequal_implEdd
29852985
_ZN4sycl3_V121__isgreaterequal_implEff
29862986
_ZN4sycl3_V122accelerator_selector_vERKNS0_6deviceE
29872987
_ZN4sycl3_V128verifyUSMAllocatorPropertiesERKNS0_13property_listE
2988-
_ZN4sycl3_V139submit_kernel_direct_without_event_implILi1EEEvRKNS0_5queueERKNS0_8nd_rangeIXT_EEERNS0_6detail17HostKernelRefBaseEPNS9_16DeviceKernelInfoERKNS9_13code_locationEb
29892988
_ZN4sycl3_V136submit_kernel_direct_with_event_implILi1EEENS0_5eventERKNS0_5queueERKNS0_8nd_rangeIXT_EEERNS0_6detail17HostKernelRefBaseEPNSA_16DeviceKernelInfoERKNSA_13code_locationEb
2989+
_ZN4sycl3_V136submit_kernel_direct_with_event_implILi2EEENS0_5eventERKNS0_5queueERKNS0_8nd_rangeIXT_EEERNS0_6detail17HostKernelRefBaseEPNSA_16DeviceKernelInfoERKNSA_13code_locationEb
29902990
_ZN4sycl3_V136submit_kernel_direct_with_event_implILi3EEENS0_5eventERKNS0_5queueERKNS0_8nd_rangeIXT_EEERNS0_6detail17HostKernelRefBaseEPNSA_16DeviceKernelInfoERKNSA_13code_locationEb
2991+
_ZN4sycl3_V139submit_kernel_direct_without_event_implILi1EEEvRKNS0_5queueERKNS0_8nd_rangeIXT_EEERNS0_6detail17HostKernelRefBaseEPNS9_16DeviceKernelInfoERKNS9_13code_locationEb
29912992
_ZN4sycl3_V139submit_kernel_direct_without_event_implILi2EEEvRKNS0_5queueERKNS0_8nd_rangeIXT_EEERNS0_6detail17HostKernelRefBaseEPNS9_16DeviceKernelInfoERKNS9_13code_locationEb
29922993
_ZN4sycl3_V139submit_kernel_direct_without_event_implILi3EEEvRKNS0_5queueERKNS0_8nd_rangeIXT_EEERNS0_6detail17HostKernelRefBaseEPNS9_16DeviceKernelInfoERKNS9_13code_locationEb
2993-
_ZN4sycl3_V136submit_kernel_direct_with_event_implILi2EEENS0_5eventERKNS0_5queueERKNS0_8nd_rangeIXT_EEERNS0_6detail17HostKernelRefBaseEPNSA_16DeviceKernelInfoERKNSA_13code_locationEb
29942994
_ZN4sycl3_V13ext5intel12experimental9pipe_base13get_pipe_nameB5cxx11EPKv
29952995
_ZN4sycl3_V13ext5intel12experimental9pipe_base17wait_non_blockingERKNS0_5eventE
29962996
_ZN4sycl3_V13ext5intel12experimental9pipe_base18get_pipe_name_implEPKv
@@ -3344,6 +3344,7 @@ _ZN4sycl3_V16detail21LocalAccessorBaseHostC1ENS0_5rangeILi3EEEiiRKNS0_13property
33443344
_ZN4sycl3_V16detail21LocalAccessorBaseHostC2ENS0_5rangeILi3EEEiiRKNS0_13property_listE
33453345
_ZN4sycl3_V16detail22addHostAccessorAndWaitEPNS1_16AccessorImplHostE
33463346
_ZN4sycl3_V16detail22free_function_info_map3addEPKPKcPKjj
3347+
_ZN4sycl3_V16detail22free_function_info_map6removeEPKPKcPKjj
33473348
_ZN4sycl3_V16detail22get_kernel_bundle_implERKNS0_7contextERKSt6vectorINS0_6deviceESaIS6_EENS0_12bundle_stateE
33483349
_ZN4sycl3_V16detail22get_kernel_bundle_implERKNS0_7contextERKSt6vectorINS0_6deviceESaIS6_EENS0_12bundle_stateERKSt8functionIFbRKSt10shared_ptrINS1_17device_image_implEEEE
33493350
_ZN4sycl3_V16detail22get_kernel_bundle_implERKNS0_7contextERKSt6vectorINS0_6deviceESaIS6_EERKNS0_4spanIcLm18446744073709551615EEENS0_12bundle_stateE
@@ -3579,6 +3580,7 @@ _ZN4sycl3_V17handler18RangeRoundingTraceEv
35793580
_ZN4sycl3_V17handler18ext_oneapi_barrierERKSt6vectorINS0_5eventESaIS3_EE
35803581
_ZN4sycl3_V17handler18extractArgsAndReqsEv
35813582
_ZN4sycl3_V17handler19getMaxWorkGroups_v2Ev
3583+
_ZN4sycl3_V17handler19setDeviceKernelInfoEONS0_6kernelE
35823584
_ZN4sycl3_V17handler19supportsUSMMemcpy2DEv
35833585
_ZN4sycl3_V17handler19supportsUSMMemset2DEv
35843586
_ZN4sycl3_V17handler20DisableRangeRoundingEv
@@ -3601,7 +3603,6 @@ _ZN4sycl3_V17handler21setKernelWorkGroupMemEm
36013603
_ZN4sycl3_V17handler21setUserFacingNodeTypeENS0_3ext6oneapi12experimental9node_typeE
36023604
_ZN4sycl3_V17handler22ext_oneapi_fill2d_implEPvmPKvmmm
36033605
_ZN4sycl3_V17handler22memcpyFromDeviceGlobalEPvPKvbmm
3604-
_ZN4sycl3_V17handler19setDeviceKernelInfoEONS0_6kernelE
36053606
_ZN4sycl3_V17handler22setDeviceKernelInfoPtrEPNS0_6detail16DeviceKernelInfoE
36063607
_ZN4sycl3_V17handler22setHandlerKernelBundleENS0_6kernelE
36073608
_ZN4sycl3_V17handler22setHandlerKernelBundleERKSt10shared_ptrINS0_6detail18kernel_bundle_implEE

0 commit comments

Comments
 (0)