From 983d957a2947c5b5620325cc1b01e83ad09c3e43 Mon Sep 17 00:00:00 2001 From: Wenju He Date: Tue, 11 Aug 2026 03:28:51 +0200 Subject: [PATCH 1/3] [DeviceSanitizer] Move sanitizer predicates to new SanitizerUtils.h This is follow-up of review comment https://github.com/intel/llvm/pull/22901#discussion_r3752443031 isModuleUsingAsan/Msan/Tsan are sanitizer-specific, not generic SYCL utilities, so move them out of SYCLLowerIR/SYCLUtils.{h,cpp} into a new SanitizerUtils.{h,cpp}. Also add ASAN/MSAN/TSAN_KERNEL_METADATA_PREFIX constants there so the "__{Asan,Msan,Tsan}KernelMetadata" prefix strings have a single definition instead of being duplicated as literals across SanitizerPostOptimizer.cpp, SanitizerPostSplitProcessing.cpp, and clang's OffloadBundler.cpp. Co-Authored-By: Claude Sonnet 5 --- clang/lib/Driver/OffloadBundler.cpp | 10 +++--- llvm/include/llvm/SYCLLowerIR/SYCLUtils.h | 8 ----- .../include/llvm/SYCLLowerIR/SanitizerUtils.h | 29 +++++++++++++++ llvm/lib/SYCLLowerIR/CMakeLists.txt | 1 + llvm/lib/SYCLLowerIR/SYCLUtils.cpp | 19 ---------- .../SYCLLowerIR/SanitizerPostOptimizer.cpp | 8 ++--- llvm/lib/SYCLLowerIR/SanitizerUtils.cpp | 36 +++++++++++++++++++ .../SYCLPostLink/ComputeModuleRuntimeInfo.cpp | 1 + llvm/lib/SYCLPostLink/ModuleSplitter.cpp | 1 + .../SanitizerPostSplitProcessing.cpp | 6 ++-- .../lib/rtc/DeviceCompilation.cpp | 1 + 11 files changed, 83 insertions(+), 37 deletions(-) create mode 100644 llvm/include/llvm/SYCLLowerIR/SanitizerUtils.h create mode 100644 llvm/lib/SYCLLowerIR/SanitizerUtils.cpp diff --git a/clang/lib/Driver/OffloadBundler.cpp b/clang/lib/Driver/OffloadBundler.cpp index f51707bd88b0a..07fcb8f7dbd13 100644 --- a/clang/lib/Driver/OffloadBundler.cpp +++ b/clang/lib/Driver/OffloadBundler.cpp @@ -23,17 +23,18 @@ #include "llvm/ADT/StringExtras.h" #include "llvm/ADT/StringMap.h" #include "llvm/ADT/StringRef.h" +#include "llvm/BinaryFormat/Magic.h" #include "llvm/Bitcode/BitcodeWriter.h" #include "llvm/IR/Constants.h" #include "llvm/IR/LLVMContext.h" #include "llvm/IRReader/IRReader.h" -#include "llvm/BinaryFormat/Magic.h" #include "llvm/Object/Archive.h" #include "llvm/Object/ArchiveWriter.h" #include "llvm/Object/Binary.h" #include "llvm/Object/Error.h" #include "llvm/Object/ObjectFile.h" #include "llvm/Object/OffloadBundle.h" +#include "llvm/SYCLLowerIR/SanitizerUtils.h" #include "llvm/Support/Casting.h" #include "llvm/Support/Compiler.h" #include "llvm/Support/Compression.h" @@ -49,12 +50,12 @@ #include "llvm/Support/Path.h" #include "llvm/Support/Program.h" #include "llvm/Support/Signals.h" +#include "llvm/Support/SourceMgr.h" #include "llvm/Support/StringSaver.h" #include "llvm/Support/TargetSelect.h" #include "llvm/Support/Timer.h" #include "llvm/Support/WithColor.h" #include "llvm/Support/raw_ostream.h" -#include "llvm/Support/SourceMgr.h" #include "llvm/TargetParser/Host.h" #include "llvm/TargetParser/Triple.h" #include @@ -756,8 +757,9 @@ class ObjectFileHandler final : public FileHandler { Name == "__AsanDeviceGlobalMetadata" || Name == "__MsanDeviceGlobalMetadata" || Name == "__TsanDeviceGlobalMetadata" || - Name == "__AsanKernelMetadata" || Name == "__MsanKernelMetadata" || - Name == "__TsanKernelMetadata")) + Name == llvm::sycl::ASAN_KERNEL_METADATA_PREFIX || + Name == llvm::sycl::MSAN_KERNEL_METADATA_PREFIX || + Name == llvm::sycl::TSAN_KERNEL_METADATA_PREFIX)) continue; // Add symbol name with the target prefix to the buffer. diff --git a/llvm/include/llvm/SYCLLowerIR/SYCLUtils.h b/llvm/include/llvm/SYCLLowerIR/SYCLUtils.h index 6238b494e9124..c9ebcdae53f4b 100644 --- a/llvm/include/llvm/SYCLLowerIR/SYCLUtils.h +++ b/llvm/include/llvm/SYCLLowerIR/SYCLUtils.h @@ -19,15 +19,7 @@ #include namespace llvm { - -class Module; - namespace sycl { - -bool isModuleUsingAsan(const Module &M); -bool isModuleUsingMsan(const Module &M); -bool isModuleUsingTsan(const Module &M); - namespace utils { constexpr char ATTR_SYCL_MODULE_ID[] = "sycl-module-id"; constexpr char ATTR_SYCL_OPTLEVEL[] = "sycl-optlevel"; diff --git a/llvm/include/llvm/SYCLLowerIR/SanitizerUtils.h b/llvm/include/llvm/SYCLLowerIR/SanitizerUtils.h new file mode 100644 index 0000000000000..cc81e76af890a --- /dev/null +++ b/llvm/include/llvm/SYCLLowerIR/SanitizerUtils.h @@ -0,0 +1,29 @@ +//===------------ SanitizerUtils.h - sanitizer utility functions --------===// +// +// 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 +// +//===----------------------------------------------------------------------===// +// Utility functions for device sanitizers. +//===----------------------------------------------------------------------===// +#pragma once + +#include "llvm/ADT/StringRef.h" + +namespace llvm { + +class Module; + +namespace sycl { + +constexpr StringRef ASAN_KERNEL_METADATA_PREFIX = "__AsanKernelMetadata"; +constexpr StringRef MSAN_KERNEL_METADATA_PREFIX = "__MsanKernelMetadata"; +constexpr StringRef TSAN_KERNEL_METADATA_PREFIX = "__TsanKernelMetadata"; + +bool isModuleUsingAsan(const Module &M); +bool isModuleUsingMsan(const Module &M); +bool isModuleUsingTsan(const Module &M); + +} // namespace sycl +} // namespace llvm diff --git a/llvm/lib/SYCLLowerIR/CMakeLists.txt b/llvm/lib/SYCLLowerIR/CMakeLists.txt index badb3fd0972fd..72a49d6d5d494 100644 --- a/llvm/lib/SYCLLowerIR/CMakeLists.txt +++ b/llvm/lib/SYCLLowerIR/CMakeLists.txt @@ -64,6 +64,7 @@ add_llvm_component_library(LLVMSYCLLowerIR TargetHelpers.cpp SanitizerPostOptimizer.cpp + SanitizerUtils.cpp ADDITIONAL_HEADER_DIRS ${LLVM_MAIN_INCLUDE_DIR}/llvm/SYCLLowerIR diff --git a/llvm/lib/SYCLLowerIR/SYCLUtils.cpp b/llvm/lib/SYCLLowerIR/SYCLUtils.cpp index 86ecf2cb61ee0..84354acf7d416 100644 --- a/llvm/lib/SYCLLowerIR/SYCLUtils.cpp +++ b/llvm/lib/SYCLLowerIR/SYCLUtils.cpp @@ -14,25 +14,6 @@ namespace llvm { namespace sycl { - -bool isModuleUsingAsan(const Module &M) { - return any_of(M.globals(), [](const GlobalVariable &GV) { - return GV.getName().starts_with("__AsanKernelMetadata"); - }); -} - -bool isModuleUsingMsan(const Module &M) { - return any_of(M.globals(), [](const GlobalVariable &GV) { - return GV.getName().starts_with("__MsanKernelMetadata"); - }); -} - -bool isModuleUsingTsan(const Module &M) { - return any_of(M.globals(), [](const GlobalVariable &GV) { - return GV.getName().starts_with("__TsanKernelMetadata"); - }); -} - namespace utils { using namespace llvm::esimd; diff --git a/llvm/lib/SYCLLowerIR/SanitizerPostOptimizer.cpp b/llvm/lib/SYCLLowerIR/SanitizerPostOptimizer.cpp index c0318044d89b6..ad0e62ac7612a 100644 --- a/llvm/lib/SYCLLowerIR/SanitizerPostOptimizer.cpp +++ b/llvm/lib/SYCLLowerIR/SanitizerPostOptimizer.cpp @@ -13,7 +13,7 @@ //===----------------------------------------------------------------------===// #include "llvm/SYCLLowerIR/SanitizerPostOptimizer.h" -#include "llvm/SYCLLowerIR/SYCLUtils.h" +#include "llvm/SYCLLowerIR/SanitizerUtils.h" #include "llvm/IR/IRBuilder.h" #include "llvm/IR/InstVisitor.h" @@ -58,9 +58,9 @@ static bool FixSanitizerKernelMetadata(Module &M) { SmallVector KernelMetadatas; for (GlobalVariable &GV : M.globals()) { auto GVName = GV.getName(); - if (GVName.starts_with("__AsanKernelMetadata") || - GVName.starts_with("__MsanKernelMetadata") || - GVName.starts_with("__TsanKernelMetadata")) { + if (GVName.starts_with(sycl::ASAN_KERNEL_METADATA_PREFIX) || + GVName.starts_with(sycl::MSAN_KERNEL_METADATA_PREFIX) || + GVName.starts_with(sycl::TSAN_KERNEL_METADATA_PREFIX)) { KernelMetadatas.push_back(&GV); } } diff --git a/llvm/lib/SYCLLowerIR/SanitizerUtils.cpp b/llvm/lib/SYCLLowerIR/SanitizerUtils.cpp new file mode 100644 index 0000000000000..60421edc5e087 --- /dev/null +++ b/llvm/lib/SYCLLowerIR/SanitizerUtils.cpp @@ -0,0 +1,36 @@ +//===------------ SanitizerUtils.cpp - sanitizer utility functions ------===// +// +// 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 +// +//===----------------------------------------------------------------------===// +// Utility functions for device sanitizers. +//===----------------------------------------------------------------------===// +#include "llvm/SYCLLowerIR/SanitizerUtils.h" +#include "llvm/IR/GlobalVariable.h" +#include "llvm/IR/Module.h" + +namespace llvm { +namespace sycl { + +bool isModuleUsingAsan(const Module &M) { + return any_of(M.globals(), [](const GlobalVariable &GV) { + return GV.getName().starts_with(ASAN_KERNEL_METADATA_PREFIX); + }); +} + +bool isModuleUsingMsan(const Module &M) { + return any_of(M.globals(), [](const GlobalVariable &GV) { + return GV.getName().starts_with(MSAN_KERNEL_METADATA_PREFIX); + }); +} + +bool isModuleUsingTsan(const Module &M) { + return any_of(M.globals(), [](const GlobalVariable &GV) { + return GV.getName().starts_with(TSAN_KERNEL_METADATA_PREFIX); + }); +} + +} // namespace sycl +} // namespace llvm diff --git a/llvm/lib/SYCLPostLink/ComputeModuleRuntimeInfo.cpp b/llvm/lib/SYCLPostLink/ComputeModuleRuntimeInfo.cpp index 0edee0156eda5..05ec9a4a70d68 100644 --- a/llvm/lib/SYCLPostLink/ComputeModuleRuntimeInfo.cpp +++ b/llvm/lib/SYCLPostLink/ComputeModuleRuntimeInfo.cpp @@ -17,6 +17,7 @@ #include "llvm/SYCLLowerIR/LowerWGLocalMemory.h" #include "llvm/SYCLLowerIR/SYCLKernelParamOptInfo.h" #include "llvm/SYCLLowerIR/SYCLUtils.h" +#include "llvm/SYCLLowerIR/SanitizerUtils.h" #include "llvm/SYCLLowerIR/SpecConstants.h" #include "llvm/SYCLPostLink/ModuleSplitter.h" #include diff --git a/llvm/lib/SYCLPostLink/ModuleSplitter.cpp b/llvm/lib/SYCLPostLink/ModuleSplitter.cpp index 26c3c0e9eb6be..d4bd48d21688e 100644 --- a/llvm/lib/SYCLPostLink/ModuleSplitter.cpp +++ b/llvm/lib/SYCLPostLink/ModuleSplitter.cpp @@ -30,6 +30,7 @@ #include "llvm/SYCLLowerIR/SYCLJointMatrixTransform.h" #include "llvm/SYCLLowerIR/SYCLUtils.h" #include "llvm/SYCLLowerIR/SanitizerPostOptimizer.h" +#include "llvm/SYCLLowerIR/SanitizerUtils.h" #include "llvm/SYCLLowerIR/SpecConstants.h" #include "llvm/SYCLPostLink/ComputeModuleRuntimeInfo.h" #include "llvm/Support/CommandLine.h" diff --git a/llvm/lib/SYCLPostLink/SanitizerPostSplitProcessing.cpp b/llvm/lib/SYCLPostLink/SanitizerPostSplitProcessing.cpp index 2e0cd13994abd..57d441967c366 100644 --- a/llvm/lib/SYCLPostLink/SanitizerPostSplitProcessing.cpp +++ b/llvm/lib/SYCLPostLink/SanitizerPostSplitProcessing.cpp @@ -14,14 +14,16 @@ #include "llvm/IR/GlobalVariable.h" #include "llvm/IR/Instructions.h" #include "llvm/IR/Module.h" +#include "llvm/SYCLLowerIR/SanitizerUtils.h" using namespace llvm; namespace { bool createSanitizerModuleID(Module &M) { - constexpr StringRef Prefixes[] = { - "__AsanKernelMetadata", "__MsanKernelMetadata", "__TsanKernelMetadata"}; + constexpr StringRef Prefixes[] = {sycl::ASAN_KERNEL_METADATA_PREFIX, + sycl::MSAN_KERNEL_METADATA_PREFIX, + sycl::TSAN_KERNEL_METADATA_PREFIX}; SmallVector ModuleIDs; for (GlobalVariable &GV : M.globals()) { auto GVName = GV.getName(); diff --git a/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp b/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp index 6cfb99e9f3319..b7528e20923ee 100644 --- a/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp +++ b/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp @@ -45,6 +45,7 @@ #include #include #include +#include #include #include #include From 3f81615885927103580109716bb332e377bf04b4 Mon Sep 17 00:00:00 2001 From: Wenju He Date: Wed, 12 Aug 2026 02:36:23 +0200 Subject: [PATCH 2/3] change namespace to llvm::sycl::utils::, matching SYCLUtils.h --- clang/lib/Driver/OffloadBundler.cpp | 6 +++--- llvm/include/llvm/SYCLLowerIR/SanitizerUtils.h | 2 ++ llvm/lib/SYCLLowerIR/SanitizerPostOptimizer.cpp | 8 ++++---- llvm/lib/SYCLLowerIR/SanitizerUtils.cpp | 2 ++ llvm/lib/SYCLPostLink/ModuleSplitter.cpp | 4 ++-- llvm/lib/SYCLPostLink/SanitizerPostSplitProcessing.cpp | 6 +++--- sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp | 2 +- 7 files changed, 17 insertions(+), 13 deletions(-) diff --git a/clang/lib/Driver/OffloadBundler.cpp b/clang/lib/Driver/OffloadBundler.cpp index 07fcb8f7dbd13..7a32a0c97c88a 100644 --- a/clang/lib/Driver/OffloadBundler.cpp +++ b/clang/lib/Driver/OffloadBundler.cpp @@ -757,9 +757,9 @@ class ObjectFileHandler final : public FileHandler { Name == "__AsanDeviceGlobalMetadata" || Name == "__MsanDeviceGlobalMetadata" || Name == "__TsanDeviceGlobalMetadata" || - Name == llvm::sycl::ASAN_KERNEL_METADATA_PREFIX || - Name == llvm::sycl::MSAN_KERNEL_METADATA_PREFIX || - Name == llvm::sycl::TSAN_KERNEL_METADATA_PREFIX)) + Name == llvm::sycl::utils::ASAN_KERNEL_METADATA_PREFIX || + Name == llvm::sycl::utils::MSAN_KERNEL_METADATA_PREFIX || + Name == llvm::sycl::utils::TSAN_KERNEL_METADATA_PREFIX)) continue; // Add symbol name with the target prefix to the buffer. diff --git a/llvm/include/llvm/SYCLLowerIR/SanitizerUtils.h b/llvm/include/llvm/SYCLLowerIR/SanitizerUtils.h index cc81e76af890a..d74c5b470eeff 100644 --- a/llvm/include/llvm/SYCLLowerIR/SanitizerUtils.h +++ b/llvm/include/llvm/SYCLLowerIR/SanitizerUtils.h @@ -16,6 +16,7 @@ namespace llvm { class Module; namespace sycl { +namespace utils { constexpr StringRef ASAN_KERNEL_METADATA_PREFIX = "__AsanKernelMetadata"; constexpr StringRef MSAN_KERNEL_METADATA_PREFIX = "__MsanKernelMetadata"; @@ -25,5 +26,6 @@ bool isModuleUsingAsan(const Module &M); bool isModuleUsingMsan(const Module &M); bool isModuleUsingTsan(const Module &M); +} // namespace utils } // namespace sycl } // namespace llvm diff --git a/llvm/lib/SYCLLowerIR/SanitizerPostOptimizer.cpp b/llvm/lib/SYCLLowerIR/SanitizerPostOptimizer.cpp index ad0e62ac7612a..32fe93ea6095f 100644 --- a/llvm/lib/SYCLLowerIR/SanitizerPostOptimizer.cpp +++ b/llvm/lib/SYCLLowerIR/SanitizerPostOptimizer.cpp @@ -58,9 +58,9 @@ static bool FixSanitizerKernelMetadata(Module &M) { SmallVector KernelMetadatas; for (GlobalVariable &GV : M.globals()) { auto GVName = GV.getName(); - if (GVName.starts_with(sycl::ASAN_KERNEL_METADATA_PREFIX) || - GVName.starts_with(sycl::MSAN_KERNEL_METADATA_PREFIX) || - GVName.starts_with(sycl::TSAN_KERNEL_METADATA_PREFIX)) { + if (GVName.starts_with(sycl::utils::ASAN_KERNEL_METADATA_PREFIX) || + GVName.starts_with(sycl::utils::MSAN_KERNEL_METADATA_PREFIX) || + GVName.starts_with(sycl::utils::TSAN_KERNEL_METADATA_PREFIX)) { KernelMetadatas.push_back(&GV); } } @@ -131,7 +131,7 @@ PreservedAnalyses SanitizerPostOptimizerPass::run(Module &M, if (!FixSanitizerKernelMetadata(M)) return PreservedAnalyses::all(); - if (sycl::isModuleUsingMsan(M)) { + if (sycl::utils::isModuleUsingMsan(M)) { EliminateDeadCheck V; V.visit(M); V.eraseDeadCheck(); diff --git a/llvm/lib/SYCLLowerIR/SanitizerUtils.cpp b/llvm/lib/SYCLLowerIR/SanitizerUtils.cpp index 60421edc5e087..0c1214db88897 100644 --- a/llvm/lib/SYCLLowerIR/SanitizerUtils.cpp +++ b/llvm/lib/SYCLLowerIR/SanitizerUtils.cpp @@ -13,6 +13,7 @@ namespace llvm { namespace sycl { +namespace utils { bool isModuleUsingAsan(const Module &M) { return any_of(M.globals(), [](const GlobalVariable &GV) { @@ -32,5 +33,6 @@ bool isModuleUsingTsan(const Module &M) { }); } +} // namespace utils } // namespace sycl } // namespace llvm diff --git a/llvm/lib/SYCLPostLink/ModuleSplitter.cpp b/llvm/lib/SYCLPostLink/ModuleSplitter.cpp index d4bd48d21688e..9d174068b28c5 100644 --- a/llvm/lib/SYCLPostLink/ModuleSplitter.cpp +++ b/llvm/lib/SYCLPostLink/ModuleSplitter.cpp @@ -1216,8 +1216,8 @@ bool runPreSplitProcessingPipeline(Module &M) { MPM.addPass(RemoveDeviceGlobalFromLLVMCompilerUsed()); // Sanitizer specific passes. - if (sycl::isModuleUsingAsan(M) || sycl::isModuleUsingMsan(M) || - sycl::isModuleUsingTsan(M)) + if (sycl::utils::isModuleUsingAsan(M) || sycl::utils::isModuleUsingMsan(M) || + sycl::utils::isModuleUsingTsan(M)) MPM.addPass(SanitizerPostOptimizerPass()); // Transform Joint Matrix builtin calls to align them with SPIR-V friendly diff --git a/llvm/lib/SYCLPostLink/SanitizerPostSplitProcessing.cpp b/llvm/lib/SYCLPostLink/SanitizerPostSplitProcessing.cpp index 57d441967c366..3de2b95b84c5f 100644 --- a/llvm/lib/SYCLPostLink/SanitizerPostSplitProcessing.cpp +++ b/llvm/lib/SYCLPostLink/SanitizerPostSplitProcessing.cpp @@ -21,9 +21,9 @@ using namespace llvm; namespace { bool createSanitizerModuleID(Module &M) { - constexpr StringRef Prefixes[] = {sycl::ASAN_KERNEL_METADATA_PREFIX, - sycl::MSAN_KERNEL_METADATA_PREFIX, - sycl::TSAN_KERNEL_METADATA_PREFIX}; + constexpr StringRef Prefixes[] = {sycl::utils::ASAN_KERNEL_METADATA_PREFIX, + sycl::utils::MSAN_KERNEL_METADATA_PREFIX, + sycl::utils::TSAN_KERNEL_METADATA_PREFIX}; SmallVector ModuleIDs; for (GlobalVariable &GV : M.globals()) { auto GVName = GV.getName(); diff --git a/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp b/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp index b7528e20923ee..bbb36d4d6b11d 100644 --- a/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp +++ b/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp @@ -70,7 +70,7 @@ using namespace clang::driver; using namespace clang::options; using namespace llvm; using namespace llvm::opt; -using namespace llvm::sycl; +using namespace llvm::sycl::utils; using namespace llvm::module_split; using namespace llvm::util; using namespace llvm::vfs; From 84ad99bccc2eaa67f7e29ee97cf13802633a618f Mon Sep 17 00:00:00 2001 From: Wenju He Date: Wed, 12 Aug 2026 09:00:47 +0200 Subject: [PATCH 3/3] fix build --- llvm/lib/SYCLPostLink/ComputeModuleRuntimeInfo.cpp | 6 +++--- sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp | 7 ++++--- 2 files changed, 7 insertions(+), 6 deletions(-) diff --git a/llvm/lib/SYCLPostLink/ComputeModuleRuntimeInfo.cpp b/llvm/lib/SYCLPostLink/ComputeModuleRuntimeInfo.cpp index 05ec9a4a70d68..d0f25b03b4a8c 100644 --- a/llvm/lib/SYCLPostLink/ComputeModuleRuntimeInfo.cpp +++ b/llvm/lib/SYCLPostLink/ComputeModuleRuntimeInfo.cpp @@ -392,11 +392,11 @@ PropSetRegTy computeModuleProperties(const Module &M, } { - if (isModuleUsingAsan(M)) + if (utils::isModuleUsingAsan(M)) PropSet.add(PropSetRegTy::SYCL_MISC_PROP, "sanUsed", "asan"); - else if (isModuleUsingMsan(M)) + else if (utils::isModuleUsingMsan(M)) PropSet.add(PropSetRegTy::SYCL_MISC_PROP, "sanUsed", "msan"); - else if (isModuleUsingTsan(M)) + else if (utils::isModuleUsingTsan(M)) PropSet.add(PropSetRegTy::SYCL_MISC_PROP, "sanUsed", "tsan"); } diff --git a/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp b/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp index bbb36d4d6b11d..7290917868a68 100644 --- a/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp +++ b/sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp @@ -70,7 +70,7 @@ using namespace clang::driver; using namespace clang::options; using namespace llvm; using namespace llvm::opt; -using namespace llvm::sycl::utils; +using namespace llvm::sycl; using namespace llvm::module_split; using namespace llvm::util; using namespace llvm::vfs; @@ -1041,8 +1041,9 @@ jit_compiler::performPostLink(ModuleUPtr Module, // Otherwise: Port over the `removeSYCLKernelsConstRefArray` and // `removeDeviceGlobalFromCompilerUsed` methods. - assert(!(isModuleUsingAsan(*Module) || isModuleUsingMsan(*Module) || - isModuleUsingTsan(*Module))); + assert(!(utils::isModuleUsingAsan(*Module) || + utils::isModuleUsingMsan(*Module) || + utils::isModuleUsingTsan(*Module))); // Otherwise: Run `SanitizerKernelMetadataPass`. // Transform Joint Matrix builtin calls to align them with SPIR-V friendly