diff --git a/.github/workflows/sycl-linux-build.yml b/.github/workflows/sycl-linux-build.yml index 8971079f08f30..d55949f520f2e 100644 --- a/.github/workflows/sycl-linux-build.yml +++ b/.github/workflows/sycl-linux-build.yml @@ -171,7 +171,8 @@ jobs: --cmake-opt=-DCMAKE_CXX_COMPILER_LAUNCHER=ccache \ --cmake-opt="-DLLVM_INSTALL_UTILS=ON" \ --cmake-opt="-DNATIVECPU_USE_OCK=Off" \ - --cmake-opt="-DLLVM_EXPERIMENTAL_TARGETS_TO_BUILD=SPIRV" + --cmake-opt="-DLLVM_EXPERIMENTAL_TARGETS_TO_BUILD=SPIRV" \ + --cmake-opt="-DSYCL_INSTALL_DEVICE_CONFIG_FILE=ON" - name: Compile id: build run: cmake --build $GITHUB_WORKSPACE/build diff --git a/.github/workflows/sycl-windows-build.yml b/.github/workflows/sycl-windows-build.yml index 77ea70c23c8d1..3bba3b527ca5d 100644 --- a/.github/workflows/sycl-windows-build.yml +++ b/.github/workflows/sycl-windows-build.yml @@ -126,7 +126,8 @@ jobs: --cmake-opt="-DCMAKE_CXX_COMPILER_LAUNCHER=ccache" ^ --cmake-opt="-DCMAKE_C_COMPILER_LAUNCHER=ccache" ^ --cmake-opt="-DLLVM_INSTALL_UTILS=ON" ^ - --cmake-opt="-DLLVM_EXPERIMENTAL_TARGETS_TO_BUILD=SPIRV" + --cmake-opt="-DLLVM_EXPERIMENTAL_TARGETS_TO_BUILD=SPIRV" ^ + --cmake-opt="-DSYCL_INSTALL_DEVICE_CONFIG_FILE=ON" - name: Build id: build shell: bash diff --git a/llvm/include/llvm/SYCLLowerIR/CMakeLists.txt b/llvm/include/llvm/SYCLLowerIR/CMakeLists.txt index 86726f9b78a20..fee74421f102b 100644 --- a/llvm/include/llvm/SYCLLowerIR/CMakeLists.txt +++ b/llvm/include/llvm/SYCLLowerIR/CMakeLists.txt @@ -6,3 +6,7 @@ set(LLVM_TABLEGEN_PROJECT LLVM) set(LLVM_TARGET_DEFINITIONS DeviceConfigFile.td) tablegen(LLVM DeviceConfigFile.inc -gen-dynamic-tables) add_public_tablegen_target(DeviceConfigFile) +install(FILES "${CMAKE_CURRENT_SOURCE_DIR}/DeviceConfigFile.hpp" + "${CMAKE_CURRENT_BINARY_DIR}/DeviceConfigFile.inc" + DESTINATION include/llvm/SYCLLowerIR + COMPONENT DeviceConfigFile) diff --git a/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.hpp b/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.hpp index 6268ecc0b26f8..2e3b7c3ee064f 100644 --- a/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.hpp +++ b/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.hpp @@ -8,12 +8,9 @@ #include #include +#include #include -namespace llvm { -class StringRef; -} - namespace DeviceConfigFile { // This struct is used in DeviceConfigFile.td. Both the fields and the name of @@ -22,7 +19,7 @@ namespace DeviceConfigFile { // DeviceConfigFile.td. struct TargetInfo { bool maySupportOtherAspects; - std::vector aspects; + std::vector aspects; std::vector subGroupSizes; std::string aotToolchain; std::string aotToolchainOptions; diff --git a/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.td b/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.td index 8fb38aa01edfe..6d349a970245e 100644 --- a/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.td +++ b/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.td @@ -164,14 +164,18 @@ defvar IntelCpuAspects = [ AspectExt_oneapi_srgb, AspectExt_oneapi_native_assert, AspectExt_intel_legacy_image, AspectExt_oneapi_ballot_group, AspectExt_oneapi_fixed_size_group, AspectExt_oneapi_opportunistic_group, - AspectExt_oneapi_tangle_group, AspectExt_oneapi_private_alloca + AspectExt_oneapi_tangle_group, AspectExt_oneapi_private_alloca, + AspectOnline_compiler, AspectOnline_linker, AspectExt_intel_gpu_slices, + AspectExt_intel_gpu_subslices_per_slice, AspectExt_intel_gpu_eu_count_per_subslice, + AspectExt_intel_gpu_hw_threads_per_eu, AspectExt_intel_device_id, + AspectExt_oneapi_virtual_functions ] # AllUSMAspects; def : TargetInfo<"spir64", [], [], "", "", 1>; def : TargetInfo<"spir64_gen", [], [], "", "", 1>; def : TargetInfo<"spir64_x86_64", IntelCpuAspects, [4, 8, 16, 32, 64], "", "", 1>; def : TargetInfo<"spir64_fpga", [], [], "", "", 1>; -def : TargetInfo<"x86_64", [], [], "", "", 1>; +def : TargetInfo<"x86_64", IntelCpuAspects, [4, 8, 16, 32, 64], "", "", 1>; // Examples of how to use a combination of explicitly specified values + predefined lists //defvar AspectList = [AspectCpu] # AllUSMAspects; //def : TargetInfo<"Test", AspectList, []>; @@ -183,9 +187,11 @@ defvar Fp16Fp64Atomic64 = [AspectFp16, AspectFp64, AspectAtomic64]; defvar Fp16Atomic64 = [AspectFp16, AspectAtomic64]; defvar Sg8_16_32 = [8, 16, 32]; defvar Sg16_32 = [16, 32]; -defvar IntelBaseAspects = [AspectExt_intel_esimd]; +defvar IntelGPUBaseAspects = [AspectExt_intel_esimd, AspectExt_oneapi_ballot_group, + AspectExt_oneapi_fixed_size_group, AspectExt_oneapi_opportunistic_group, + AspectExt_oneapi_tangle_group]; class IntelTargetInfo Aspects, list subGroupSizesList> -: TargetInfo; +: TargetInfo; // Note: only the "canonical" target names are listed here - see // SYCL::gen::resolveGenDevice(). // @@ -249,8 +255,13 @@ class CudaTargetInfo aspectList, int subGroupSiz assert !eq(subGroupSize, 32), "sub-group size for Cuda must be equal to 32 and not " # subGroupSize # "."; } -defvar CudaMinAspects = !listconcat(AllUSMAspects, [AspectGpu, AspectFp64, AspectOnline_compiler, AspectOnline_linker, - AspectQueue_profiling, AspectExt_intel_pci_address, AspectExt_intel_max_mem_bandwidth, AspectExt_intel_memory_bus_width, +defvar CudaUSMAspects = [ + AspectUsm_device_allocations, AspectUsm_shared_allocations, + AspectUsm_atomic_shared_allocations +]; + +defvar CudaMinAspects = !listconcat(CudaUSMAspects, [AspectGpu, AspectFp64, AspectOnline_compiler, AspectOnline_linker, + AspectQueue_profiling, AspectExt_intel_pci_address, AspectExt_intel_memory_bus_width, AspectExt_intel_device_info_uuid, AspectExt_oneapi_native_assert, AspectExt_intel_free_memory, AspectExt_intel_device_id, AspectExt_intel_memory_clock_rate, AspectExt_oneapi_ballot_group, AspectExt_oneapi_fixed_size_group, AspectExt_oneapi_opportunistic_group, AspectExt_oneapi_graph, AspectExt_oneapi_limited_graph]); @@ -306,37 +317,43 @@ defvar HipSubgroupSizesRDNA = [32]; // gfxX10-gfx11 (encapsulates RDNA1..3), defvar HipSubgroupSizesCDNA = [64]; // gfx908, gfx90a (encapsulates CDNA1..2) defvar HipMinAspects = [AspectGpu, AspectFp64, AspectOnline_compiler, AspectOnline_linker, AspectQueue_profiling, - AspectExt_intel_pci_address, AspectExt_intel_max_mem_bandwidth, AspectExt_intel_device_id, + AspectExt_intel_pci_address, AspectExt_intel_device_id, AspectExt_intel_memory_clock_rate, AspectExt_intel_memory_bus_width, AspectExt_intel_free_memory]; +defvar HipUSMAspects = [ + AspectUsm_device_allocations, AspectUsm_host_allocations, + AspectUsm_shared_allocations, AspectUsm_atomic_host_allocations, + AspectUsm_atomic_shared_allocations +]; + // The following AMDGCN targets are ordered based on their ROCm driver support: // // Officially supported: -def : HipTargetInfo<"amd_gpu_gfx908", !listconcat(HipMinAspects, AllUSMAspects, +def : HipTargetInfo<"amd_gpu_gfx908", !listconcat(HipMinAspects, HipUSMAspects, [AspectExt_intel_device_info_uuid, AspectExt_oneapi_graph, AspectExt_oneapi_limited_graph]), HipSubgroupSizesCDNA>; -def : HipTargetInfo<"amd_gpu_gfx90a", !listconcat(HipMinAspects, AllUSMAspects, +def : HipTargetInfo<"amd_gpu_gfx90a", !listconcat(HipMinAspects, HipUSMAspects, [AspectExt_intel_device_info_uuid, AspectExt_oneapi_graph, AspectExt_oneapi_limited_graph, AspectExt_oneapi_native_assert]), HipSubgroupSizesCDNA>; // TODO: Need to verify whether device-side asserts (oneapi_native_assert) are // now working for the new CDNA3 gfx940, gfx941, gfx942 GPUs and fixed for the // other supported, gfx1030 and gfx1100, RDNA3 GPUs. -def : HipTargetInfo<"amd_gpu_gfx940", !listconcat(HipMinAspects, AllUSMAspects, +def : HipTargetInfo<"amd_gpu_gfx940", !listconcat(HipMinAspects, HipUSMAspects, [AspectExt_intel_device_info_uuid, AspectExt_oneapi_graph, AspectExt_oneapi_limited_graph]), HipSubgroupSizesCDNA>; -def : HipTargetInfo<"amd_gpu_gfx941", !listconcat(HipMinAspects, AllUSMAspects, +def : HipTargetInfo<"amd_gpu_gfx941", !listconcat(HipMinAspects, HipUSMAspects, [AspectExt_intel_device_info_uuid, AspectExt_oneapi_graph, AspectExt_oneapi_limited_graph]), HipSubgroupSizesCDNA>; -def : HipTargetInfo<"amd_gpu_gfx942", !listconcat(HipMinAspects, AllUSMAspects, +def : HipTargetInfo<"amd_gpu_gfx942", !listconcat(HipMinAspects, HipUSMAspects, [AspectExt_intel_device_info_uuid, AspectExt_oneapi_graph, AspectExt_oneapi_limited_graph]), HipSubgroupSizesCDNA>; -def : HipTargetInfo<"amd_gpu_gfx1030", !listconcat(HipMinAspects, AllUSMAspects, +def : HipTargetInfo<"amd_gpu_gfx1030", !listconcat(HipMinAspects, HipUSMAspects, [AspectExt_intel_device_info_uuid, AspectExt_oneapi_graph, AspectExt_oneapi_limited_graph]), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1100", !listconcat(HipMinAspects, AllUSMAspects, +def : HipTargetInfo<"amd_gpu_gfx1100", !listconcat(HipMinAspects, HipUSMAspects, [AspectExt_intel_device_info_uuid, AspectExt_oneapi_graph, AspectExt_oneapi_limited_graph]), HipSubgroupSizesRDNA>; // Deprecated support: -def : HipTargetInfo<"amd_gpu_gfx906", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesGCN5>; +def : HipTargetInfo<"amd_gpu_gfx906", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesGCN5>; // Unsupported (or unofficially supported): def : HipTargetInfo<"amd_gpu_gfx700", HipMinAspects, HipSubgroupSizesGCN2>; def : HipTargetInfo<"amd_gpu_gfx701", HipMinAspects, HipSubgroupSizesGCN2>; @@ -353,23 +370,23 @@ def : HipTargetInfo<"amd_gpu_gfx900", HipMinAspects, HipSubgroupSizesGCN5>; def : HipTargetInfo<"amd_gpu_gfx902", HipMinAspects, HipSubgroupSizesGCN5>; def : HipTargetInfo<"amd_gpu_gfx904", HipMinAspects, HipSubgroupSizesGCN5>; def : HipTargetInfo<"amd_gpu_gfx909", HipMinAspects, HipSubgroupSizesGCN5>; -def : HipTargetInfo<"amd_gpu_gfx90c", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesGCN5>; -def : HipTargetInfo<"amd_gpu_gfx1010", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1011", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1012", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1013", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1031", !listconcat(!listremove(HipMinAspects, [AspectExt_intel_free_memory]), AllUSMAspects), +def : HipTargetInfo<"amd_gpu_gfx90c", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesGCN5>; +def : HipTargetInfo<"amd_gpu_gfx1010", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1011", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1012", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1013", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1031", !listconcat(!listremove(HipMinAspects, [AspectExt_intel_free_memory]), HipUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1032", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1033", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1034", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1035", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1036", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1101", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1102", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1103", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1150", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; -def : HipTargetInfo<"amd_gpu_gfx1151", !listconcat(HipMinAspects, AllUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1032", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1033", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1034", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1035", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1036", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1101", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1102", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1103", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1150", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; +def : HipTargetInfo<"amd_gpu_gfx1151", !listconcat(HipMinAspects, HipUSMAspects), HipSubgroupSizesRDNA>; // TBA def : HipTargetInfo<"amd_gpu_gfx1200", [], []>; // RDNA 4 def : HipTargetInfo<"amd_gpu_gfx1201", [], []>; // RDNA 4 diff --git a/sycl/CMakeLists.txt b/sycl/CMakeLists.txt index cdc84bc122b57..998c2a37522ea 100644 --- a/sycl/CMakeLists.txt +++ b/sycl/CMakeLists.txt @@ -532,6 +532,11 @@ if("hip" IN_LIST SYCL_ENABLE_BACKENDS) list(APPEND SYCL_TOOLCHAIN_DEPLOY_COMPONENTS ur_adapter_hip) endif() +if(SYCL_INSTALL_DEVICE_CONFIG_FILE) + add_dependencies(sycl-toolchain DeviceConfigFile) + list(APPEND SYCL_TOOLCHAIN_DEPLOY_COMPONENTS DeviceConfigFile) +endif() + # Use it as fake dependency in order to force another command(s) to execute. add_custom_command(OUTPUT __force_it COMMAND "${CMAKE_COMMAND}" -E echo diff --git a/sycl/test-e2e/Basic/device_config_file_consistency.cpp b/sycl/test-e2e/Basic/device_config_file_consistency.cpp new file mode 100644 index 0000000000000..ba2fb9e8bdd4e --- /dev/null +++ b/sycl/test-e2e/Basic/device_config_file_consistency.cpp @@ -0,0 +1,139 @@ +// This test checks to see if every aspect and sub-group size declared in the +// device config file is supported by the device. Note this does not mean +// check that the device config file is exhaustive, only that the device +// supports everything it declares. However, this test does print out any +// aspects that are supported by the device but not declared in the device +// config file. + +// UNSUPPORTED: accelerator +// UNSUPPORTED-INTENDED: Accelerator is not supported by +// sycl_ext_oneapi_device_architecture. +// REQUIRES: device-config-file +// RUN: %{build} -o %t.out %device_config_file_include_flag +// RUN: %{run} %t.out +#include + +#include +#include + +#define __SYCL_ASPECT_DEPRECATED_ALIAS(ASPECT, ID, MESSAGE) \ + __SYCL_ASPECT_DEPRECATED(ASPECT, ID, MESSAGE) + +using namespace sycl; + +const char *getArchName(const device &Device) { + namespace syclex = sycl::ext::oneapi::experimental; + auto Arch = Device.get_info(); + switch (Arch) { +#define __SYCL_ARCHITECTURE(ARCH, VAL) \ + case syclex::architecture::ARCH: \ + return #ARCH; +#define __SYCL_ARCHITECTURE_ALIAS(ARCH, VAL) +#include +#undef __SYCL_ARCHITECTURE +#undef __SYCL_ARCHITECTURE_ALIAS + } + return "unknown"; +} + +// Checks if a container contains a specific element. +template +bool contains(const Container &C, const T &Elem) { + return std::find(C.begin(), C.end(), Elem) != C.end(); +} + +std::string_view getAspectName(aspect Asp) { + switch (Asp) { +#define __SYCL_ASPECT(ASPECT, ID) \ + case aspect::ASPECT: \ + return #ASPECT; +#include +#undef __SYCL_ASPECT + } + return "unknown"; +} + +aspect getAspectByName(std::string_view Name) { +#define __SYCL_ASPECT(ASPECT, ID) \ + if (Name == #ASPECT) \ + return aspect::ASPECT; +#include + throw std::invalid_argument("Unknown aspect name"); +} + +int main() { + // Get the device arch + queue Q; + auto Dev = Q.get_device(); + auto DeviceName = getArchName(Dev); + + auto TargetInfo = DeviceConfigFile::TargetTable.find(DeviceName); + if (TargetInfo == DeviceConfigFile::TargetTable.end()) { + std::cout << "No aspects found for device " << DeviceName << "\n"; + return 1; + } + + // Check aspects consistency + int NAspectInconsistencies = 0; + std::cout << "Checking consistency of aspects for device " << DeviceName + << "...\n"; + + auto SupportedAspects = Dev.get_info(); + auto DeviceConfigAspectNames = TargetInfo->second.aspects; + std::vector DeviceConfigAspects; + for (auto AspectName : DeviceConfigAspectNames) { + DeviceConfigAspects.push_back(getAspectByName(AspectName)); + } + + for (auto Asp : DeviceConfigAspects) { + if (!contains(SupportedAspects, Asp)) { + std::cout << "error: " << DeviceName << " does not support aspect " + << getAspectName(Asp) + << " but it is declared in the device config file\n"; + ++NAspectInconsistencies; + } + } + for (auto Asp : SupportedAspects) { + if (!contains(DeviceConfigAspects, Asp)) { + std::cout << "note: the device " << DeviceName << " supports aspect " + << getAspectName(Asp) + << " but it is not declared in the device config file\n"; + // Not necessarily an error, so we won't increment n_fail + } + } + + if (NAspectInconsistencies == 0) + std::cout << "All aspects are consistent\n"; + + // Check sub-group sizes consistency + int NSubGroupSizeInconsistencies = 0; + std::cout << "Checking consistency of sub-group sizes for device " + << DeviceName << "...\n"; + + auto SupportedSubGroupSizes = Dev.get_info(); + auto DeviceConfigSubGroupSizes = TargetInfo->second.subGroupSizes; + + for (auto Size : DeviceConfigSubGroupSizes) { + if (!contains(SupportedSubGroupSizes, Size)) { + std::cout << "error: " << DeviceName + << " does not support sub-group size " << Size + << " but it is declared in the device config file\n"; + ++NSubGroupSizeInconsistencies; + } + } + for (auto Size : SupportedSubGroupSizes) { + if (!contains(DeviceConfigSubGroupSizes, Size)) { + std::cout << "note: the device " << DeviceName + << " supports sub-group size " << Size + << " but it is not declared in the device config file\n"; + // Not necessarily an error, so we won't increment n_fail + } + } + + if (NSubGroupSizeInconsistencies == 0) + std::cout << "All sub-group sizes are consistent\n"; + + return NAspectInconsistencies + NSubGroupSizeInconsistencies; +} + +#undef __SYCL_ASPECT_DEPRECATED_ALIAS diff --git a/sycl/test-e2e/lit.cfg.py b/sycl/test-e2e/lit.cfg.py index 5b60d93387b7a..2ba0d1ade0243 100644 --- a/sycl/test-e2e/lit.cfg.py +++ b/sycl/test-e2e/lit.cfg.py @@ -683,6 +683,17 @@ def open_check_file(file_name): # Use short names for LIT rules. config.available_features.add("any-device-is-" + be) +if config.llvm_main_include_dir: + lit_config.note("Using device config file built from LLVM") + config.available_features.add("device-config-file") + config.substitutions.append( + ("%device_config_file_include_flag", f"-I {config.llvm_main_include_dir}") + ) +elif os.path.exists(f"{config.sycl_include}/llvm/SYCLLowerIR/DeviceConfigFile.hpp"): + lit_config.note("Using installed device config file") + config.available_features.add("device-config-file") + config.substitutions.append(("%device_config_file_include_flag", "")) + # That has to be executed last so that all device-independent features have been # discovered already. config.sycl_dev_features = {} diff --git a/sycl/test-e2e/lit.site.cfg.py.in b/sycl/test-e2e/lit.site.cfg.py.in index 00928dd9141fc..82e586b94f0c6 100644 --- a/sycl/test-e2e/lit.site.cfg.py.in +++ b/sycl/test-e2e/lit.site.cfg.py.in @@ -10,6 +10,7 @@ site.addsitedir("@CMAKE_CURRENT_SOURCE_DIR@") config.dpcpp_compiler = lit_config.params.get("dpcpp_compiler", "@SYCL_CXX_COMPILER@") config.dpcpp_root_dir= os.path.dirname(os.path.dirname(config.dpcpp_compiler)) +config.llvm_main_include_dir = "@LLVM_MAIN_INCLUDE_DIR@" config.llvm_tools_dir = os.path.join(config.dpcpp_root_dir, 'bin') config.lit_tools_dir = os.path.dirname("@TEST_SUITE_LIT@") config.dump_ir_supported = lit_config.params.get("dump_ir", ("@DUMP_IR_SUPPORTED@" if "@DUMP_IR_SUPPORTED@" else False)) diff --git a/sycl/test/basic_tests/device_config_file_aspects.cpp b/sycl/test/basic_tests/device_config_file_aspects.cpp index da08b87e60a2b..90c23d9ef593c 100644 --- a/sycl/test/basic_tests/device_config_file_aspects.cpp +++ b/sycl/test/basic_tests/device_config_file_aspects.cpp @@ -3,7 +3,6 @@ // #include -#include #include #include @@ -16,7 +15,7 @@ int main() { auto aspectsList = testAspects->second.aspects; #define __SYCL_ASPECT(ASPECT, ASPECT_VAL) \ - llvm::StringRef s##ASPECT(#ASPECT); \ + std::string_view s##ASPECT(#ASPECT); \ assert(std::find(aspectsList.begin(), aspectsList.end(), s##ASPECT) != \ aspectsList.end()); @@ -29,7 +28,7 @@ int main() { assert(testDeprecatedAspects != DeviceConfigFile::TargetTable.end()); auto deprecatedAspectsList = testDeprecatedAspects->second.aspects; #define __SYCL_ASPECT_DEPRECATED(ASPECT, ASPECT_VAL, MSG) \ - llvm::StringRef s##ASPECT(#ASPECT); \ + std::string_view s##ASPECT(#ASPECT); \ assert(std::find(deprecatedAspectsList.begin(), deprecatedAspectsList.end(), \ s##ASPECT) != deprecatedAspectsList.end());