diff --git a/clang/test/Driver/clang-offload-wrapper.c b/clang/test/Driver/clang-offload-wrapper.c index fdbb95614b215..af04576846b50 100644 --- a/clang/test/Driver/clang-offload-wrapper.c +++ b/clang/test/Driver/clang-offload-wrapper.c @@ -120,13 +120,17 @@ // CHECK-IR: declare void @__tgt_unregister_lib([[DESCTY]]*) // CHECK-IR: define internal void [[SYCL_REGFN]]() -// CHECK-IR: call void @__tgt_register_lib([[DESCTY]]* bitcast ([[SYCL_DESCTY]]* [[SYCL_DESC]] to [[DESCTY]]*)) +// CHECK-IR: call void @__sycl_register_lib([[SYCL_DESCTY]]* [[SYCL_DESC]]) // CHECK-IR: ret void +// CHECK-IR: declare void @__sycl_register_lib([[SYCL_DESCTY]]*) + // CHECK-IR: define internal void [[SYCL_UNREGFN]]() -// CHECK-IR: call void @__tgt_unregister_lib([[DESCTY]]* bitcast ([[SYCL_DESCTY]]* [[SYCL_DESC]] to [[DESCTY]]*)) +// CHECK-IR: call void @__sycl_unregister_lib([[SYCL_DESCTY]]* [[SYCL_DESC]]) // CHECK-IR: ret void +// CHECK-IR: declare void @__sycl_unregister_lib([[SYCL_DESCTY]]*) + // ------- // Check options' effects: -emit-reg-funcs, -desc-name // diff --git a/clang/tools/clang-offload-wrapper/ClangOffloadWrapper.cpp b/clang/tools/clang-offload-wrapper/ClangOffloadWrapper.cpp index 505b6581bdc1f..489b51ae37e8b 100644 --- a/clang/tools/clang-offload-wrapper/ClangOffloadWrapper.cpp +++ b/clang/tools/clang-offload-wrapper/ClangOffloadWrapper.cpp @@ -756,15 +756,18 @@ class BinaryWrapper { Func->setSection(".text.startup"); // Get RegFuncName function declaration. - auto *RegFuncTy = FunctionType::get(Type::getVoidTy(C), getBinDescPtrTy(), - /*isVarArg*/ false); + auto *RegFuncTy = FunctionType::get( + Type::getVoidTy(C), + Kind == OffloadKind::SYCL ? getSyclBinDescPtrTy() : getBinDescPtrTy(), + /*isVarArg=*/false); FunctionCallee RegFuncC = - M.getOrInsertFunction("__tgt_register_lib", RegFuncTy); + M.getOrInsertFunction(Kind == OffloadKind::SYCL ? "__sycl_register_lib" + : "__tgt_register_lib", + RegFuncTy); // Construct function body IRBuilder<> Builder(BasicBlock::Create(C, "entry", Func)); - Builder.CreateCall(RegFuncC, - Builder.CreatePointerCast(BinDesc, getBinDescPtrTy())); + Builder.CreateCall(RegFuncC, BinDesc); Builder.CreateRetVoid(); // Add this function to constructors. @@ -779,15 +782,18 @@ class BinaryWrapper { Func->setSection(".text.startup"); // Get UnregFuncName function declaration. - auto *UnRegFuncTy = FunctionType::get(Type::getVoidTy(C), getBinDescPtrTy(), - /*isVarArg*/ false); - FunctionCallee UnRegFuncC = - M.getOrInsertFunction("__tgt_unregister_lib", UnRegFuncTy); + auto *UnRegFuncTy = FunctionType::get( + Type::getVoidTy(C), + Kind == OffloadKind::SYCL ? getSyclBinDescPtrTy() : getBinDescPtrTy(), + /*isVarArg=*/false); + FunctionCallee UnRegFuncC = M.getOrInsertFunction( + Kind == OffloadKind::SYCL ? "__sycl_unregister_lib" + : "__tgt_unregister_lib", + UnRegFuncTy); // Construct function body IRBuilder<> Builder(BasicBlock::Create(C, "entry", Func)); - Builder.CreateCall(UnRegFuncC, - Builder.CreatePointerCast(BinDesc, getBinDescPtrTy())); + Builder.CreateCall(UnRegFuncC, BinDesc); Builder.CreateRetVoid(); // Add this function to global destructors. diff --git a/sycl/include/CL/sycl/detail/program_manager/program_manager.hpp b/sycl/include/CL/sycl/detail/program_manager/program_manager.hpp index 88d49fcbce8b4..884fae473d4fd 100644 --- a/sycl/include/CL/sycl/detail/program_manager/program_manager.hpp +++ b/sycl/include/CL/sycl/detail/program_manager/program_manager.hpp @@ -20,12 +20,12 @@ /// Executed as a part of current module's (.exe, .dll) static initialization. /// Registers device executable images with the runtime. -extern "C" void __tgt_register_lib(pi_device_binaries desc); +extern "C" void __sycl_register_lib(pi_device_binaries desc); /// Executed as a part of current module's (.exe, .dll) static /// de-initialization. /// Unregisters device executable images with the runtime. -extern "C" void __tgt_unregister_lib(pi_device_binaries desc); +extern "C" void __sycl_unregister_lib(pi_device_binaries desc); // +++ } diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp index 2c273dcfa46ad..3a40c5c114eda 100644 --- a/sycl/source/detail/program_manager/program_manager.cpp +++ b/sycl/source/detail/program_manager/program_manager.cpp @@ -831,11 +831,11 @@ void ProgramManager::dumpImage(const DeviceImage &Img, KernelSetId KSId) const { } // namespace sycl } // namespace cl -extern "C" void __tgt_register_lib(pi_device_binaries desc) { +extern "C" void __sycl_register_lib(pi_device_binaries desc) { cl::sycl::detail::ProgramManager::getInstance().addImages(desc); } // Executed as a part of current module's (.exe, .dll) static initialization -extern "C" void __tgt_unregister_lib(pi_device_binaries desc) { +extern "C" void __sycl_unregister_lib(pi_device_binaries desc) { // TODO implement the function } diff --git a/sycl/source/ld-version-script.txt b/sycl/source/ld-version-script.txt index df90e944bed00..f66b55fd09a9c 100644 --- a/sycl/source/ld-version-script.txt +++ b/sycl/source/ld-version-script.txt @@ -24,8 +24,8 @@ _Z20__spirv_ocl_prefetch*; /* Export offload image hooks */ - __tgt_register_lib; - __tgt_unregister_lib; + __sycl_register_lib; + __sycl_unregister_lib; local: *;