From c1a7ad666aeb7fc83c2f7f5d2424f4eef372a18b Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Tue, 10 Dec 2024 03:48:45 -0800 Subject: [PATCH 01/17] introduce SPIR-V Backend support in LIT tests and CI: a new LIT feature 'spirv-backend', substitutions in e2e are aware about the new LIT feature, new records in the sycl-nightly workflow to use the new LIT feature --- .github/workflows/sycl-nightly.yml | 56 ++++++++++++++++++++++++++++++ sycl/test-e2e/README.md | 6 ++++ sycl/test-e2e/format.py | 10 +++++- sycl/test-e2e/lit.cfg.py | 2 ++ 4 files changed, 73 insertions(+), 1 deletion(-) diff --git a/.github/workflows/sycl-nightly.yml b/.github/workflows/sycl-nightly.yml index 5c6b6ad89f57d..cb4c4222dba1d 100644 --- a/.github/workflows/sycl-nightly.yml +++ b/.github/workflows/sycl-nightly.yml @@ -87,6 +87,62 @@ jobs: image_options: -u 1001 target_devices: opencl:cpu tests_selector: e2e + + - name: SYCL-CTS on OCL CPU + runner: '["Linux", "gen12"]' + image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest + image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN + target_devices: opencl:cpu + tests_selector: cts + + - name: SYCL-CTS on L0 gen12 + runner: '["Linux", "gen12"]' + image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest + image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN + target_devices: level_zero:gpu + tests_selector: cts + + - name: Intel L0 GPU w/ LLVM SPIR-V Backend + runner: '["Linux", "gen12"]' + image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest + image_options: -u 1001 --device=/dev/dri -v /dev/dri/by-path:/dev/dri/by-path --privileged --cap-add SYS_ADMIN + target_devices: level_zero:gpu + reset_intel_gpu: true + tests_selector: e2e + extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True + + - name: Intel OCL GPU w/ LLVM SPIR-V Backend + runner: '["Linux", "gen12"]' + image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest + image_options: -u 1001 --device=/dev/dri -v /dev/dri/by-path:/dev/dri/by-path --privileged --cap-add SYS_ADMIN + target_devices: opencl:gpu + reset_intel_gpu: true + tests_selector: e2e + extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True + + - name: OCL CPU (Intel/GEN12) w/ LLVM SPIR-V Backend + runner: '["Linux", "gen12"]' + image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest + image_options: -u 1001 --privileged --cap-add SYS_ADMIN + target_devices: opencl:cpu + tests_selector: e2e + extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True + + - name: SYCL-CTS on OCL CPU w/ LLVM SPIR-V Backend + runner: '["Linux", "gen12"]' + image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest + image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN + target_devices: opencl:cpu + tests_selector: cts + extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True + + - name: SYCL-CTS on L0 gen12 w/ LLVM SPIR-V Backend + runner: '["Linux", "gen12"]' + image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest + image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN + target_devices: level_zero:gpu + tests_selector: cts + extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True uses: ./.github/workflows/sycl-linux-run-tests.yml with: name: ${{ matrix.name }} diff --git a/sycl/test-e2e/README.md b/sycl/test-e2e/README.md index 396aa9ef7341a..508c35ef98afb 100644 --- a/sycl/test-e2e/README.md +++ b/sycl/test-e2e/README.md @@ -256,6 +256,12 @@ place. No new tests should use these features: * **gpu-intel-pvc** - Intel GPU PVC availability; * **gpu-intel-pvc-vg** - Intel GPU PVC-VG availability; +### Use the LLVM SPIR-V Backend to generate SPIR-V code + +It's possible to use the LLVM SPIR-V Backend instead of llvm-spirv tool to +convert LLVM IR to SPIR-V. This feature must be set manually by passing the +`spirv-backend` argument to `llvm-lit`. + ### llvm-lit parameters Following options can be passed to llvm-lit tool through --param option to diff --git a/sycl/test-e2e/format.py b/sycl/test-e2e/format.py index a69810145507d..662a85d145675 100644 --- a/sycl/test-e2e/format.py +++ b/sycl/test-e2e/format.py @@ -175,13 +175,21 @@ def execute(self, test, litConfig): substitutions = lit.TestRunner.getDefaultSubstitutions(test, tmpDir, tmpBase) substitutions.append(("%{sycl_triple}", format(",".join(triples)))) + # -fsycl-use-spirv-backend-for-spirv-gen is needed + # to support "spirv-backend" feature + fsycl_use_spirv_backend = ( + "-fsycl-use-spirv-backend-for-spirv-gen" + if "spirv-backend" in test.config.available_features + else "" + ) # -fsycl-targets is needed for CUDA/HIP, so just use it be default so # -that new tests by default would runnable there (unless they have # -other restrictions). substitutions.append( ( "%{build}", - "%clangxx -fsycl -fsycl-targets=%{sycl_triple} %verbose_print %s", + "%clangxx -fsycl -fsycl-targets=%{sycl_triple} %verbose_print %s" + + fsycl_use_spirv_backend, ) ) if platform.system() == "Windows": diff --git a/sycl/test-e2e/lit.cfg.py b/sycl/test-e2e/lit.cfg.py index d723d3452b7d6..5550722ebab19 100644 --- a/sycl/test-e2e/lit.cfg.py +++ b/sycl/test-e2e/lit.cfg.py @@ -225,6 +225,8 @@ def check_igc_tag_and_add_feature(): if lit_config.params.get("enable-perf-tests", False): config.available_features.add("enable-perf-tests") +if lit_config.params.get("spirv-backend", False): + config.available_features.add("spirv-backend") # Use this to make sure that any dynamic checks below are done in the build # directory and not where the sources are located. This is important for the From 0644c4d40692c1e444a3572941005ab415653548 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Tue, 10 Dec 2024 04:30:34 -0800 Subject: [PATCH 02/17] fix --param value --- .github/workflows/sycl-nightly.yml | 10 +++++----- 1 file changed, 5 insertions(+), 5 deletions(-) diff --git a/.github/workflows/sycl-nightly.yml b/.github/workflows/sycl-nightly.yml index cb4c4222dba1d..92f0f5ffa0c44 100644 --- a/.github/workflows/sycl-nightly.yml +++ b/.github/workflows/sycl-nightly.yml @@ -109,7 +109,7 @@ jobs: target_devices: level_zero:gpu reset_intel_gpu: true tests_selector: e2e - extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True + extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True - name: Intel OCL GPU w/ LLVM SPIR-V Backend runner: '["Linux", "gen12"]' @@ -118,7 +118,7 @@ jobs: target_devices: opencl:gpu reset_intel_gpu: true tests_selector: e2e - extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True + extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True - name: OCL CPU (Intel/GEN12) w/ LLVM SPIR-V Backend runner: '["Linux", "gen12"]' @@ -126,7 +126,7 @@ jobs: image_options: -u 1001 --privileged --cap-add SYS_ADMIN target_devices: opencl:cpu tests_selector: e2e - extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True + extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True - name: SYCL-CTS on OCL CPU w/ LLVM SPIR-V Backend runner: '["Linux", "gen12"]' @@ -134,7 +134,7 @@ jobs: image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN target_devices: opencl:cpu tests_selector: cts - extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True + extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True - name: SYCL-CTS on L0 gen12 w/ LLVM SPIR-V Backend runner: '["Linux", "gen12"]' @@ -142,7 +142,7 @@ jobs: image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN target_devices: level_zero:gpu tests_selector: cts - extra_lit_opts: --param spirv-backend --param gpu-intel-gen12=True + extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True uses: ./.github/workflows/sycl-linux-run-tests.yml with: name: ${{ matrix.name }} From 8a54a5d047d62ffcf5ee9c251fe7b2ca50dc46d8 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Tue, 10 Dec 2024 07:17:35 -0800 Subject: [PATCH 03/17] Update a command to execute SPIRV-V Backend in the toolchain; Update unsupported test cases and mark temporarily failing with a known issue with XFAIL/XFAIL-TRACKER; fix how the 'build' substitution is updated for spirv-backend LIT feature --- clang/lib/Driver/ToolChains/Clang.cpp | 38 +++++-------------- sycl/test-e2e/AddressSanitizer/lit.local.cfg | 3 ++ .../DeviceGlobal/multi_device_images.cpp | 3 ++ .../out-of-bounds/local/multiple_source.cpp | 4 ++ sycl/test-e2e/Basic/backend_info.cpp | 2 +- .../Basic/fpga_tests/fpga_io_pipes.cpp | 2 +- sycl/test-e2e/Basic/multisource.cpp | 3 ++ .../free_function_kernels.cpp | 3 ++ .../NewOffloadDriver/math_device_lib.cpp | 3 ++ sycl/test-e2e/ESIMD/lit.local.cfg | 3 ++ sycl/test-e2e/InvokeSimd/lit.local.cfg | 3 ++ .../LLVMIntrinsicLowering/bitreverse.cpp | 3 ++ .../sub_byte_bitreverse.cpp | 3 ++ .../test-e2e/NewOffloadDriver/multisource.cpp | 3 ++ .../sycl-external-with-optional-features.cpp | 3 ++ .../NonUniformGroups/is_fixed_topology.cpp | 2 +- .../NonUniformGroups/is_user_constructed.cpp | 2 +- .../sycl-external-with-optional-features.cpp | 3 ++ .../Regression/DAE-separate-compile.cpp | 3 ++ sycl/test-e2e/SeparateCompile/same-kernel.cpp | 3 ++ .../sycl-external-within-staticlib.cpp | 3 ++ .../SeparateCompile/sycl-external.cpp | 3 ++ sycl/test-e2e/format.py | 2 +- 23 files changed, 67 insertions(+), 33 deletions(-) diff --git a/clang/lib/Driver/ToolChains/Clang.cpp b/clang/lib/Driver/ToolChains/Clang.cpp index 832733eb2bda8..bfadd7b1619e0 100644 --- a/clang/lib/Driver/ToolChains/Clang.cpp +++ b/clang/lib/Driver/ToolChains/Clang.cpp @@ -10702,38 +10702,20 @@ static void getSPIRVBackendOpts(const llvm::opt::ArgList &TCArgs, ArgStringList &BackendArgs) { BackendArgs.push_back(TCArgs.MakeArgString("-filetype=obj")); BackendArgs.push_back( - TCArgs.MakeArgString("-mtriple=spirv64-unknown-unknown")); - // TODO: Optimization level is currently forced to -O0 due to some testing - // issues. Update optimization level after testing issues are resolved. - BackendArgs.push_back(TCArgs.MakeArgString("-O0")); + TCArgs.MakeArgString("-mtriple=spirv64v1.6-unknown-unknown")); BackendArgs.push_back( TCArgs.MakeArgString("--avoid-spirv-capabilities=Shader")); BackendArgs.push_back( TCArgs.MakeArgString("--translator-compatibility-mode")); - - // TODO: There is some overlap between the lists of extensions in SPIR-V - // backend and SPIR-V Trnaslator). We will try to combine them when SPIR-V - // backdn is ready. - std::string ExtArg("--spirv-ext="); - std::string DefaultExtArg = - "+SPV_EXT_shader_atomic_float_add,+SPV_EXT_shader_atomic_float_min_max" - ",+SPV_KHR_no_integer_wrap_decoration,+SPV_KHR_float_controls" - ",+SPV_KHR_expect_assume,+SPV_KHR_linkonce_odr"; - std::string INTELExtArg = ",+SPV_INTEL_subgroups,+SPV_INTEL_function_pointers" - ",+SPV_INTEL_arbitrary_precision_integers" - ",+SPV_INTEL_variable_length_array"; - ExtArg = ExtArg + DefaultExtArg + INTELExtArg; - - // Other args - ExtArg += ",+SPV_INTEL_bfloat16_conversion" - ",+SPV_KHR_uniform_group_instructions" - ",+SPV_INTEL_optnone" - ",+SPV_KHR_subgroup_rotate" - ",+SPV_INTEL_usm_storage_classes" - ",+SPV_EXT_shader_atomic_float16_add" - ",+SPV_KHR_bit_instructions"; - - BackendArgs.push_back(TCArgs.MakeArgString(ExtArg)); + // TODO: A list of SPIR-V extensions that are supported by the SPIR-V backend + // is growing. Let's postpone the decision on which extensions to enable until + // - the list is stable, and + // - we decide on a mapping of user requested extensions into backend's ones. + // Meanwhile we enable all the SPIR-V backend extensions. + BackendArgs.push_back(TCArgs.MakeArgString("--spirv-ext=all")); + // TODO: + // - handle -Xspirv-translator option to avoid "argument unused during compilation" error + // - handle --spirv-ext=+ and --spirv-ext=- options } // Utility function to gather all llvm-spirv options. diff --git a/sycl/test-e2e/AddressSanitizer/lit.local.cfg b/sycl/test-e2e/AddressSanitizer/lit.local.cfg index d768697d07f6d..73ea3c1824e1e 100644 --- a/sycl/test-e2e/AddressSanitizer/lit.local.cfg +++ b/sycl/test-e2e/AddressSanitizer/lit.local.cfg @@ -10,3 +10,6 @@ config.substitutions.append( # https://github.com/intel/llvm/issues/15953 config.unsupported_features += ['gpu-intel-gen12'] + +# CMPLRLLVM-64052 +config.unsupported_features += ['spirv-backend'] diff --git a/sycl/test-e2e/AddressSanitizer/out-of-bounds/DeviceGlobal/multi_device_images.cpp b/sycl/test-e2e/AddressSanitizer/out-of-bounds/DeviceGlobal/multi_device_images.cpp index 5cab931adc4ab..05b044a5ead5d 100644 --- a/sycl/test-e2e/AddressSanitizer/out-of-bounds/DeviceGlobal/multi_device_images.cpp +++ b/sycl/test-e2e/AddressSanitizer/out-of-bounds/DeviceGlobal/multi_device_images.cpp @@ -4,6 +4,9 @@ // RUN: %clangxx -fsycl %device_asan_flags -O2 -g %t1.o %t2.o -o %t.out // RUN: %{run} not %t.out 2>&1 | FileCheck %s +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #include #include diff --git a/sycl/test-e2e/AddressSanitizer/out-of-bounds/local/multiple_source.cpp b/sycl/test-e2e/AddressSanitizer/out-of-bounds/local/multiple_source.cpp index abb152ab4719a..797eb4de216be 100644 --- a/sycl/test-e2e/AddressSanitizer/out-of-bounds/local/multiple_source.cpp +++ b/sycl/test-e2e/AddressSanitizer/out-of-bounds/local/multiple_source.cpp @@ -3,6 +3,10 @@ // RUN: %{build} %device_asan_flags -O2 -g -DUSER_CODE_2 -c -o %t2.o // RUN: %clangxx -fsycl %device_asan_flags -O2 -g %t1.o %t2.o -o %t.out // RUN: %{run} not %t.out 2>&1 | FileCheck %s + +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #include constexpr std::size_t N = 4; diff --git a/sycl/test-e2e/Basic/backend_info.cpp b/sycl/test-e2e/Basic/backend_info.cpp index f61b1809a6e46..d5fa958f45b18 100644 --- a/sycl/test-e2e/Basic/backend_info.cpp +++ b/sycl/test-e2e/Basic/backend_info.cpp @@ -1,7 +1,7 @@ // RUN: %{build} -o %t.out // RUN: %{run} %t.out // -// RUN: %{build} -DTEST_ERRORS -D_GLIBCXX_USE_CXX11_ABI=0 -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note +// RUN: %{build} -DTEST_ERRORS -D_GLIBCXX_USE_CXX11_ABI=0 -fsyntax-only -Wno-error=unused-command-line-argument -Xclang -verify -Xclang -verify-ignore-unexpected=note //==--- backend_info.cpp - SYCL backend info test---------------------------==// // diff --git a/sycl/test-e2e/Basic/fpga_tests/fpga_io_pipes.cpp b/sycl/test-e2e/Basic/fpga_tests/fpga_io_pipes.cpp index 1eb50966b76ee..83d7ed7b1684b 100644 --- a/sycl/test-e2e/Basic/fpga_tests/fpga_io_pipes.cpp +++ b/sycl/test-e2e/Basic/fpga_tests/fpga_io_pipes.cpp @@ -1,4 +1,4 @@ -// RUN: %{build} -fsyntax-only -o %t.out +// RUN: %{build} -fsyntax-only -Wno-error=unused-command-line-argument -o %t.out // TODO: launch the test if the feature is supported. //==------------ fpga_io_pipes.cpp - SYCL FPGA pipes test ------------------==// // diff --git a/sycl/test-e2e/Basic/multisource.cpp b/sycl/test-e2e/Basic/multisource.cpp index 23c95ce2eddd3..e09e3126db132 100644 --- a/sycl/test-e2e/Basic/multisource.cpp +++ b/sycl/test-e2e/Basic/multisource.cpp @@ -19,6 +19,9 @@ // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %t.init.o %t.calc.o %t.main.o -Wno-unused-command-line-argument -o %t2.fat // RUN: %{run} %t2.fat +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #include #include diff --git a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/free_function_kernels.cpp b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/free_function_kernels.cpp index 74758a837cd46..7c43170d5181f 100644 --- a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/free_function_kernels.cpp +++ b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/free_function_kernels.cpp @@ -11,6 +11,9 @@ // XFAIL: hip_amd // XFAIL-TRACKER: https://github.com/intel/llvm/issues/15742 +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16319 + #include #include #include diff --git a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/math_device_lib.cpp b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/math_device_lib.cpp index 949063d62b0c1..09723cf3d725b 100644 --- a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/math_device_lib.cpp +++ b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/math_device_lib.cpp @@ -7,6 +7,9 @@ // RUN: %{build} --offload-new-driver -fsycl-allow-device-image-dependencies -fsycl-device-lib-jit-link %{mathflags} -o %t.out // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16319 + #include #include diff --git a/sycl/test-e2e/ESIMD/lit.local.cfg b/sycl/test-e2e/ESIMD/lit.local.cfg index 41191cc92a6c5..d4e58ef498ecd 100644 --- a/sycl/test-e2e/ESIMD/lit.local.cfg +++ b/sycl/test-e2e/ESIMD/lit.local.cfg @@ -21,3 +21,6 @@ for substitution in config.substitutions: original_clangxx=substitution[1] config.substitutions.insert(0, ("%clangxx", original_clangxx+" -Wno-error=deprecated-declarations")) + +# At the moment SPIR-V Backend has no plans to suport this feature. +config.unsupported_features += ['spirv-backend'] diff --git a/sycl/test-e2e/InvokeSimd/lit.local.cfg b/sycl/test-e2e/InvokeSimd/lit.local.cfg index 7f24b7421e028..054df935c362b 100644 --- a/sycl/test-e2e/InvokeSimd/lit.local.cfg +++ b/sycl/test-e2e/InvokeSimd/lit.local.cfg @@ -6,3 +6,6 @@ config.required_features += ['gpu'] # TODO: enable on Windows once driver is ready. if platform.system() != "Linux": config.unsupported = True + +# At the moment SPIR-V Backend has no plans to suport this feature. +config.unsupported_features += ['spirv-backend'] diff --git a/sycl/test-e2e/LLVMIntrinsicLowering/bitreverse.cpp b/sycl/test-e2e/LLVMIntrinsicLowering/bitreverse.cpp index 6217ac66ade30..e1d3a76c70b4d 100644 --- a/sycl/test-e2e/LLVMIntrinsicLowering/bitreverse.cpp +++ b/sycl/test-e2e/LLVMIntrinsicLowering/bitreverse.cpp @@ -2,6 +2,9 @@ // UNSUPPORTED: hip || cuda +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16318, CMPLRLLVM-62187 + // Make dump directory. // RUN: rm -rf %t.spvdir && mkdir %t.spvdir diff --git a/sycl/test-e2e/LLVMIntrinsicLowering/sub_byte_bitreverse.cpp b/sycl/test-e2e/LLVMIntrinsicLowering/sub_byte_bitreverse.cpp index fe22a32a49d2a..976d30b6be49c 100644 --- a/sycl/test-e2e/LLVMIntrinsicLowering/sub_byte_bitreverse.cpp +++ b/sycl/test-e2e/LLVMIntrinsicLowering/sub_byte_bitreverse.cpp @@ -6,6 +6,9 @@ // XFAIL: gpu // XFAIL-TRACKER: https://github.com/intel/intel-graphics-compiler/issues/330 +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16318, CMPLRLLVM-62187 + // Make dump directory. // RUN: rm -rf %t.spvdir && mkdir %t.spvdir diff --git a/sycl/test-e2e/NewOffloadDriver/multisource.cpp b/sycl/test-e2e/NewOffloadDriver/multisource.cpp index cf9f518c89995..90fb502f56247 100644 --- a/sycl/test-e2e/NewOffloadDriver/multisource.cpp +++ b/sycl/test-e2e/NewOffloadDriver/multisource.cpp @@ -39,6 +39,9 @@ // RUN: %clangxx -Wno-error=unused-command-line-argument -fsycl -fsycl-targets=%{sycl_triple} --offload-new-driver %t.main.o %t.a -o %t4.fat // RUN: %{run} %t4.fat +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #include #include diff --git a/sycl/test-e2e/NewOffloadDriver/sycl-external-with-optional-features.cpp b/sycl/test-e2e/NewOffloadDriver/sycl-external-with-optional-features.cpp index 144466f673bba..2c48c99d9cda7 100644 --- a/sycl/test-e2e/NewOffloadDriver/sycl-external-with-optional-features.cpp +++ b/sycl/test-e2e/NewOffloadDriver/sycl-external-with-optional-features.cpp @@ -6,6 +6,9 @@ // XFAIL: cuda // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16413 +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #ifdef SOURCE1 #include #include diff --git a/sycl/test-e2e/NonUniformGroups/is_fixed_topology.cpp b/sycl/test-e2e/NonUniformGroups/is_fixed_topology.cpp index e8183ce41d3d2..480e9d3763bd5 100644 --- a/sycl/test-e2e/NonUniformGroups/is_fixed_topology.cpp +++ b/sycl/test-e2e/NonUniformGroups/is_fixed_topology.cpp @@ -1,4 +1,4 @@ -// RUN: %{build} -fsyntax-only -o %t.out +// RUN: %{build} -fsyntax-only -Wno-error=unused-command-line-argument -o %t.out #include namespace syclex = sycl::ext::oneapi::experimental; diff --git a/sycl/test-e2e/NonUniformGroups/is_user_constructed.cpp b/sycl/test-e2e/NonUniformGroups/is_user_constructed.cpp index c3ea8b59dc54c..35c4017779b78 100644 --- a/sycl/test-e2e/NonUniformGroups/is_user_constructed.cpp +++ b/sycl/test-e2e/NonUniformGroups/is_user_constructed.cpp @@ -1,5 +1,5 @@ // FIXME: Move to sycl/test. -// RUN: %{build} -fsyntax-only -o %t.out +// RUN: %{build} -fsyntax-only -Wno-error=unused-command-line-argument -o %t.out #include #include diff --git a/sycl/test-e2e/OptionalKernelFeatures/sycl-external-with-optional-features.cpp b/sycl/test-e2e/OptionalKernelFeatures/sycl-external-with-optional-features.cpp index 5a04ea4ed55df..f4b9264ade953 100644 --- a/sycl/test-e2e/OptionalKernelFeatures/sycl-external-with-optional-features.cpp +++ b/sycl/test-e2e/OptionalKernelFeatures/sycl-external-with-optional-features.cpp @@ -3,6 +3,9 @@ // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %t1.o %t2.o -Wno-unused-command-line-argument -o %t.exe // RUN: %{run} %t.exe +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #ifdef SOURCE1 #include #include diff --git a/sycl/test-e2e/Regression/DAE-separate-compile.cpp b/sycl/test-e2e/Regression/DAE-separate-compile.cpp index 64e19ec8f90ad..90069369dc0d2 100644 --- a/sycl/test-e2e/Regression/DAE-separate-compile.cpp +++ b/sycl/test-e2e/Regression/DAE-separate-compile.cpp @@ -8,6 +8,9 @@ // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %t.o %O0 -Wno-unused-command-line-argument -o %t.out // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + // Failing on HIP AMD, enable after fixed // UNSUPPORTED: hip_amd diff --git a/sycl/test-e2e/SeparateCompile/same-kernel.cpp b/sycl/test-e2e/SeparateCompile/same-kernel.cpp index 8bdadbe7ad62b..a8ce30833e1dd 100644 --- a/sycl/test-e2e/SeparateCompile/same-kernel.cpp +++ b/sycl/test-e2e/SeparateCompile/same-kernel.cpp @@ -15,6 +15,9 @@ // RUN: %clangxx %t-same-kernel-a.o %t-same-kernel-b.o -Wno-unused-command-line-argument -o %t-same-kernel.exe -fsycl -fsycl-targets=%{sycl_triple} // RUN: %{run} %t-same-kernel.exe +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #include using namespace sycl; diff --git a/sycl/test-e2e/SeparateCompile/sycl-external-within-staticlib.cpp b/sycl/test-e2e/SeparateCompile/sycl-external-within-staticlib.cpp index 219634f47646c..4572174dd2638 100644 --- a/sycl/test-e2e/SeparateCompile/sycl-external-within-staticlib.cpp +++ b/sycl/test-e2e/SeparateCompile/sycl-external-within-staticlib.cpp @@ -16,6 +16,9 @@ // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} -O3 %t3.o %t_repacked.a -Wno-unused-command-line-argument -o %t2.exe // RUN: %{run} %t2.exe +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #include #include diff --git a/sycl/test-e2e/SeparateCompile/sycl-external.cpp b/sycl/test-e2e/SeparateCompile/sycl-external.cpp index 37facb7ecfc57..b8b9ca6b743ce 100644 --- a/sycl/test-e2e/SeparateCompile/sycl-external.cpp +++ b/sycl/test-e2e/SeparateCompile/sycl-external.cpp @@ -12,6 +12,9 @@ // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %t2.o %t.a -Wno-unused-command-line-argument -o %t2.exe // RUN: %{run} %t2.exe +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #include #include diff --git a/sycl/test-e2e/format.py b/sycl/test-e2e/format.py index 662a85d145675..aaabb02c22f94 100644 --- a/sycl/test-e2e/format.py +++ b/sycl/test-e2e/format.py @@ -178,7 +178,7 @@ def execute(self, test, litConfig): # -fsycl-use-spirv-backend-for-spirv-gen is needed # to support "spirv-backend" feature fsycl_use_spirv_backend = ( - "-fsycl-use-spirv-backend-for-spirv-gen" + " -fsycl-use-spirv-backend-for-spirv-gen" if "spirv-backend" in test.config.available_features else "" ) From 213bcdd56e50c2d43488d72611b4a11111566988 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Tue, 10 Dec 2024 08:25:11 -0800 Subject: [PATCH 04/17] Postpone the question of running SYCL CTS to a next PR --- .github/workflows/sycl-nightly.yml | 16 ---------------- 1 file changed, 16 deletions(-) diff --git a/.github/workflows/sycl-nightly.yml b/.github/workflows/sycl-nightly.yml index 92f0f5ffa0c44..0af64822e74dd 100644 --- a/.github/workflows/sycl-nightly.yml +++ b/.github/workflows/sycl-nightly.yml @@ -127,22 +127,6 @@ jobs: target_devices: opencl:cpu tests_selector: e2e extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True - - - name: SYCL-CTS on OCL CPU w/ LLVM SPIR-V Backend - runner: '["Linux", "gen12"]' - image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest - image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN - target_devices: opencl:cpu - tests_selector: cts - extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True - - - name: SYCL-CTS on L0 gen12 w/ LLVM SPIR-V Backend - runner: '["Linux", "gen12"]' - image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest - image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN - target_devices: level_zero:gpu - tests_selector: cts - extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True uses: ./.github/workflows/sycl-linux-run-tests.yml with: name: ${{ matrix.name }} From 275ad48783a2f76a10448810f79b39532d248a00 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Tue, 7 Jan 2025 07:47:58 -0800 Subject: [PATCH 05/17] move 'build-only' tests into sycl/test --- .github/workflows/sycl-nightly.yml | 14 -------------- .../Basic => test}/fpga_tests/fpga_io_pipes.cpp | 2 +- .../non-uniform-groups}/is_fixed_topology.cpp | 2 +- .../non-uniform-groups}/is_user_constructed.cpp | 3 +-- 4 files changed, 3 insertions(+), 18 deletions(-) rename sycl/{test-e2e/Basic => test}/fpga_tests/fpga_io_pipes.cpp (98%) rename sycl/{test-e2e/NonUniformGroups => test/non-uniform-groups}/is_fixed_topology.cpp (88%) rename sycl/{test-e2e/NonUniformGroups => test/non-uniform-groups}/is_user_constructed.cpp (87%) diff --git a/.github/workflows/sycl-nightly.yml b/.github/workflows/sycl-nightly.yml index 0af64822e74dd..f2ced79c42f75 100644 --- a/.github/workflows/sycl-nightly.yml +++ b/.github/workflows/sycl-nightly.yml @@ -88,20 +88,6 @@ jobs: target_devices: opencl:cpu tests_selector: e2e - - name: SYCL-CTS on OCL CPU - runner: '["Linux", "gen12"]' - image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest - image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN - target_devices: opencl:cpu - tests_selector: cts - - - name: SYCL-CTS on L0 gen12 - runner: '["Linux", "gen12"]' - image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest - image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN - target_devices: level_zero:gpu - tests_selector: cts - - name: Intel L0 GPU w/ LLVM SPIR-V Backend runner: '["Linux", "gen12"]' image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest diff --git a/sycl/test-e2e/Basic/fpga_tests/fpga_io_pipes.cpp b/sycl/test/fpga_tests/fpga_io_pipes.cpp similarity index 98% rename from sycl/test-e2e/Basic/fpga_tests/fpga_io_pipes.cpp rename to sycl/test/fpga_tests/fpga_io_pipes.cpp index 83d7ed7b1684b..1eb50966b76ee 100644 --- a/sycl/test-e2e/Basic/fpga_tests/fpga_io_pipes.cpp +++ b/sycl/test/fpga_tests/fpga_io_pipes.cpp @@ -1,4 +1,4 @@ -// RUN: %{build} -fsyntax-only -Wno-error=unused-command-line-argument -o %t.out +// RUN: %{build} -fsyntax-only -o %t.out // TODO: launch the test if the feature is supported. //==------------ fpga_io_pipes.cpp - SYCL FPGA pipes test ------------------==// // diff --git a/sycl/test-e2e/NonUniformGroups/is_fixed_topology.cpp b/sycl/test/non-uniform-groups/is_fixed_topology.cpp similarity index 88% rename from sycl/test-e2e/NonUniformGroups/is_fixed_topology.cpp rename to sycl/test/non-uniform-groups/is_fixed_topology.cpp index 480e9d3763bd5..e8183ce41d3d2 100644 --- a/sycl/test-e2e/NonUniformGroups/is_fixed_topology.cpp +++ b/sycl/test/non-uniform-groups/is_fixed_topology.cpp @@ -1,4 +1,4 @@ -// RUN: %{build} -fsyntax-only -Wno-error=unused-command-line-argument -o %t.out +// RUN: %{build} -fsyntax-only -o %t.out #include namespace syclex = sycl::ext::oneapi::experimental; diff --git a/sycl/test-e2e/NonUniformGroups/is_user_constructed.cpp b/sycl/test/non-uniform-groups/is_user_constructed.cpp similarity index 87% rename from sycl/test-e2e/NonUniformGroups/is_user_constructed.cpp rename to sycl/test/non-uniform-groups/is_user_constructed.cpp index 35c4017779b78..44fb8b816aec2 100644 --- a/sycl/test-e2e/NonUniformGroups/is_user_constructed.cpp +++ b/sycl/test/non-uniform-groups/is_user_constructed.cpp @@ -1,5 +1,4 @@ -// FIXME: Move to sycl/test. -// RUN: %{build} -fsyntax-only -Wno-error=unused-command-line-argument -o %t.out +// RUN: %{build} -fsyntax-only -o %t.out #include #include From ad767a06c72577f5fe1aab12934439c6e999c469 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Wed, 8 Jan 2025 03:35:39 -0800 Subject: [PATCH 06/17] Update test cases with XFAIL/UNSUPPORTED after intel/llvm and llvm-project updates --- sycl/test-e2e/Basic/multisource_spv_obj.cpp | 3 +++ sycl/test-e2e/Compression/compression.cpp | 3 +++ sycl/test-e2e/Config/env_vars.cpp | 3 +++ .../DeviceImageDependencies/NewOffloadDriver/dynamic.cpp | 3 +++ .../DeviceImageDependencies/NewOffloadDriver/objects.cpp | 3 +++ .../NewOffloadDriver/singleDynamicLibrary.cpp | 3 +++ sycl/test-e2e/DeviceLib/ITTAnnotations/atomic.cpp | 3 +++ sycl/test-e2e/DeviceLib/ITTAnnotations/barrier.cpp | 3 +++ sycl/test-e2e/DeviceLib/imf_simd_emulate_test.cpp | 4 ++++ .../GroupAlgorithm/SYCL2020/reduce_over_group_size.cpp | 3 +++ sycl/test-e2e/HierPar/hier_par_wgscope_O0.cpp | 3 +++ sycl/test-e2e/MemorySanitizer/check_buffer.cpp | 3 +++ sycl/test-e2e/MemorySanitizer/check_call.cpp | 3 +++ sycl/test-e2e/MemorySanitizer/check_divide.cpp | 3 +++ sycl/test-e2e/MemorySanitizer/check_large_access.cpp | 3 +++ sycl/test-e2e/NonUniformGroups/ballot_group_algorithms.cpp | 4 ++++ .../test-e2e/NonUniformGroups/fixed_size_group_algorithms.cpp | 4 ++++ .../is_compatible/is_compatible_with_aspects.cpp | 3 +++ .../PropagateOptionsToBackend/sycl-opt-level-opencl.cpp | 3 +++ .../Regression/default-constructed-local-accessor.cpp | 3 +++ sycl/test-e2e/Regression/unoptimized_stream.cpp | 3 +++ sycl/test-e2e/Scheduler/DataMovement.cpp | 3 +++ sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp | 2 +- 23 files changed, 70 insertions(+), 1 deletion(-) diff --git a/sycl/test-e2e/Basic/multisource_spv_obj.cpp b/sycl/test-e2e/Basic/multisource_spv_obj.cpp index 0f097ce3cd5db..5369426ad99c0 100644 --- a/sycl/test-e2e/Basic/multisource_spv_obj.cpp +++ b/sycl/test-e2e/Basic/multisource_spv_obj.cpp @@ -28,6 +28,9 @@ // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %t.init.o %t.calc.o %t.main.o -Wno-unused-command-line-argument -o %t3.fat // RUN: %{run} %t3.fat +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64059 + #include #include diff --git a/sycl/test-e2e/Compression/compression.cpp b/sycl/test-e2e/Compression/compression.cpp index 26147407f73c7..3c7d765f2e12f 100644 --- a/sycl/test-e2e/Compression/compression.cpp +++ b/sycl/test-e2e/Compression/compression.cpp @@ -6,3 +6,6 @@ // RUN: %{run} %t_not_compress.out // RUN: %{run} %t_compress.out // RUN: not diff %t_not_compress.out %t_compress.out + +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 diff --git a/sycl/test-e2e/Config/env_vars.cpp b/sycl/test-e2e/Config/env_vars.cpp index 7316ed5095dad..4d243f1841d51 100644 --- a/sycl/test-e2e/Config/env_vars.cpp +++ b/sycl/test-e2e/Config/env_vars.cpp @@ -13,6 +13,9 @@ // RUN: %if cpu %{ env SYCL_PROGRAM_COMPILE_OPTIONS="-enable-link-options -cl-denorms-are-zero" SHOULD_CRASH=1 %{run} %t.out %} // RUN: %if cpu %{ env SYCL_PROGRAM_APPEND_COMPILE_OPTIONS="-enable-link-options -cl-denorms-are-zero" SHOULD_CRASH=1 %{run} %t.out %} +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include "../helpers.hpp" #include #include diff --git a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/dynamic.cpp b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/dynamic.cpp index ff75d442cb892..4df006681987f 100644 --- a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/dynamic.cpp +++ b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/dynamic.cpp @@ -19,6 +19,9 @@ // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16319 + #include "a.hpp" #include #include diff --git a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/objects.cpp b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/objects.cpp index 9e71e88a35422..976cbfcaf7488 100644 --- a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/objects.cpp +++ b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/objects.cpp @@ -10,6 +10,9 @@ // RUN: %{build} --offload-new-driver -fsycl-allow-device-image-dependencies %t_a.o %t_b.o %t_c.o %t_d.o -I %S/Inputs -o %t.out // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16319 + #include "a.hpp" #include #include diff --git a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/singleDynamicLibrary.cpp b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/singleDynamicLibrary.cpp index cde5f7cdbaad9..cf1da88594b3c 100644 --- a/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/singleDynamicLibrary.cpp +++ b/sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/singleDynamicLibrary.cpp @@ -20,6 +20,9 @@ // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16319 + #include "wrapper.hpp" int main() { return (wrapper()); } diff --git a/sycl/test-e2e/DeviceLib/ITTAnnotations/atomic.cpp b/sycl/test-e2e/DeviceLib/ITTAnnotations/atomic.cpp index 272c21885365f..5f6ae8e73fae5 100644 --- a/sycl/test-e2e/DeviceLib/ITTAnnotations/atomic.cpp +++ b/sycl/test-e2e/DeviceLib/ITTAnnotations/atomic.cpp @@ -1,6 +1,9 @@ // RUN: %{build} -Wno-error=unused-command-line-argument -fsycl-instrument-device-code -o %t.out // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include "sycl/atomic_ref.hpp" #include "sycl/detail/core.hpp" diff --git a/sycl/test-e2e/DeviceLib/ITTAnnotations/barrier.cpp b/sycl/test-e2e/DeviceLib/ITTAnnotations/barrier.cpp index 6597b06576744..99b4294a3fab9 100644 --- a/sycl/test-e2e/DeviceLib/ITTAnnotations/barrier.cpp +++ b/sycl/test-e2e/DeviceLib/ITTAnnotations/barrier.cpp @@ -1,6 +1,9 @@ // RUN: %{build} -Wno-error=unused-command-line-argument -fsycl-instrument-device-code -o %t.out // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include "sycl/detail/core.hpp" #include diff --git a/sycl/test-e2e/DeviceLib/imf_simd_emulate_test.cpp b/sycl/test-e2e/DeviceLib/imf_simd_emulate_test.cpp index 241e8a15c681e..36293d56e595a 100644 --- a/sycl/test-e2e/DeviceLib/imf_simd_emulate_test.cpp +++ b/sycl/test-e2e/DeviceLib/imf_simd_emulate_test.cpp @@ -5,6 +5,10 @@ // RUN: %{run} %t2.out // // UNSUPPORTED: cuda || hip + +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include "imf_utils.hpp" #include #include diff --git a/sycl/test-e2e/GroupAlgorithm/SYCL2020/reduce_over_group_size.cpp b/sycl/test-e2e/GroupAlgorithm/SYCL2020/reduce_over_group_size.cpp index a9dc4e352838f..12029762fe364 100644 --- a/sycl/test-e2e/GroupAlgorithm/SYCL2020/reduce_over_group_size.cpp +++ b/sycl/test-e2e/GroupAlgorithm/SYCL2020/reduce_over_group_size.cpp @@ -8,6 +8,9 @@ // RUN: %{build} %O0 -o %t_O0.out // RUN: %{run} %t_O0.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + /* test performs a lattice reduction. sycl::vec is sensitive to .get_size() vs .size() in SYCL headers diff --git a/sycl/test-e2e/HierPar/hier_par_wgscope_O0.cpp b/sycl/test-e2e/HierPar/hier_par_wgscope_O0.cpp index 59a03f18feb4c..662b38334b122 100644 --- a/sycl/test-e2e/HierPar/hier_par_wgscope_O0.cpp +++ b/sycl/test-e2e/HierPar/hier_par_wgscope_O0.cpp @@ -9,6 +9,9 @@ // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + // This test checks correctness of hierarchical kernel execution when there is // code and data in the work group scope, and when the test is compiled with // -O0 switch. diff --git a/sycl/test-e2e/MemorySanitizer/check_buffer.cpp b/sycl/test-e2e/MemorySanitizer/check_buffer.cpp index df4c6ef7bfc5a..02781c560ea85 100644 --- a/sycl/test-e2e/MemorySanitizer/check_buffer.cpp +++ b/sycl/test-e2e/MemorySanitizer/check_buffer.cpp @@ -6,6 +6,9 @@ // RUN: %{build} %device_msan_flags -O2 -g -o %t3.out // RUN: %{run} not %t3.out 2>&1 | FileCheck %s +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include __attribute__((noinline)) long long foo(int data1, long long data2) { diff --git a/sycl/test-e2e/MemorySanitizer/check_call.cpp b/sycl/test-e2e/MemorySanitizer/check_call.cpp index aa2a608027876..1d271bcb79ed2 100644 --- a/sycl/test-e2e/MemorySanitizer/check_call.cpp +++ b/sycl/test-e2e/MemorySanitizer/check_call.cpp @@ -6,6 +6,9 @@ // RUN: %{build} %device_msan_flags -O2 -g -o %t3.out // RUN: %{run} not %t3.out 2>&1 | FileCheck %s +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include #include diff --git a/sycl/test-e2e/MemorySanitizer/check_divide.cpp b/sycl/test-e2e/MemorySanitizer/check_divide.cpp index 99d0d588cbbac..512d93b8891c3 100644 --- a/sycl/test-e2e/MemorySanitizer/check_divide.cpp +++ b/sycl/test-e2e/MemorySanitizer/check_divide.cpp @@ -6,6 +6,9 @@ // RUN: %{build} %device_msan_flags -O2 -g -o %t3.out // RUN: %{run} not %t3.out 2>&1 | FileCheck %s +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include #include diff --git a/sycl/test-e2e/MemorySanitizer/check_large_access.cpp b/sycl/test-e2e/MemorySanitizer/check_large_access.cpp index cb06f1b9ace7a..bf9b48f9b4601 100644 --- a/sycl/test-e2e/MemorySanitizer/check_large_access.cpp +++ b/sycl/test-e2e/MemorySanitizer/check_large_access.cpp @@ -4,6 +4,9 @@ // RUN: %{build} %device_msan_flags -O2 -g -o %t3.out // RUN: %{run} not %t3.out 2>&1 | FileCheck %s +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include "sycl/detail/core.hpp" #include diff --git a/sycl/test-e2e/NonUniformGroups/ballot_group_algorithms.cpp b/sycl/test-e2e/NonUniformGroups/ballot_group_algorithms.cpp index 14e976edc8492..da9154ae602d0 100644 --- a/sycl/test-e2e/NonUniformGroups/ballot_group_algorithms.cpp +++ b/sycl/test-e2e/NonUniformGroups/ballot_group_algorithms.cpp @@ -13,6 +13,10 @@ // UNSUPPORTED: cuda // UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/12995 +// UNSUPPORTED: spirv-backend +// UNSUPPORTED-TRACKER: CMPLRLLVM-64702 +// The test is disabled for spirv-backend while we investigate the root cause. + #include #include #include diff --git a/sycl/test-e2e/NonUniformGroups/fixed_size_group_algorithms.cpp b/sycl/test-e2e/NonUniformGroups/fixed_size_group_algorithms.cpp index dfed10e9f587e..07898d0146fb9 100644 --- a/sycl/test-e2e/NonUniformGroups/fixed_size_group_algorithms.cpp +++ b/sycl/test-e2e/NonUniformGroups/fixed_size_group_algorithms.cpp @@ -9,6 +9,10 @@ // REQUIRES: sg-32 // REQUIRES: aspect-ext_oneapi_fixed_size_group +// UNSUPPORTED: spirv-backend +// UNSUPPORTED-TRACKER: CMPLRLLVM-64702 +// The test is disabled for spirv-backend while we investigate the root cause. + #include #include #include diff --git a/sycl/test-e2e/OptionalKernelFeatures/is_compatible/is_compatible_with_aspects.cpp b/sycl/test-e2e/OptionalKernelFeatures/is_compatible/is_compatible_with_aspects.cpp index e65ef55201089..1afdf9975cf99 100644 --- a/sycl/test-e2e/OptionalKernelFeatures/is_compatible/is_compatible_with_aspects.cpp +++ b/sycl/test-e2e/OptionalKernelFeatures/is_compatible/is_compatible_with_aspects.cpp @@ -4,6 +4,9 @@ // RUN: %{build} -Wno-error=incorrect-sub-group-size %O0 -o %t.out // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include #include diff --git a/sycl/test-e2e/PropagateOptionsToBackend/sycl-opt-level-opencl.cpp b/sycl/test-e2e/PropagateOptionsToBackend/sycl-opt-level-opencl.cpp index 78283f5b7cf62..db98542e76f68 100644 --- a/sycl/test-e2e/PropagateOptionsToBackend/sycl-opt-level-opencl.cpp +++ b/sycl/test-e2e/PropagateOptionsToBackend/sycl-opt-level-opencl.cpp @@ -22,6 +22,9 @@ // -O2 | /* no option */ // -O3 | /* no option */ +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include int main() { diff --git a/sycl/test-e2e/Regression/default-constructed-local-accessor.cpp b/sycl/test-e2e/Regression/default-constructed-local-accessor.cpp index 07950111aa345..b9fb35a901f48 100644 --- a/sycl/test-e2e/Regression/default-constructed-local-accessor.cpp +++ b/sycl/test-e2e/Regression/default-constructed-local-accessor.cpp @@ -3,6 +3,9 @@ // RUN: %{build} -o %t.out %O0 // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include using namespace sycl; diff --git a/sycl/test-e2e/Regression/unoptimized_stream.cpp b/sycl/test-e2e/Regression/unoptimized_stream.cpp index b107947ce5ea7..39b19a11549d3 100644 --- a/sycl/test-e2e/Regression/unoptimized_stream.cpp +++ b/sycl/test-e2e/Regression/unoptimized_stream.cpp @@ -1,6 +1,9 @@ // RUN: %{build} %O0 -o %t.out // RUN: %{run} %t.out +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 + #include #include diff --git a/sycl/test-e2e/Scheduler/DataMovement.cpp b/sycl/test-e2e/Scheduler/DataMovement.cpp index c06f11fef7c59..ce7f2eda9f760 100644 --- a/sycl/test-e2e/Scheduler/DataMovement.cpp +++ b/sycl/test-e2e/Scheduler/DataMovement.cpp @@ -1,6 +1,9 @@ // RUN: %{build} -o %t.out %debug_option // RUN: %{run} %t.out // +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/llvm/llvm-project/issues/122075 +// //==-------------------------- DataMovement.cpp ----------------------------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. diff --git a/sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp b/sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp index f01a25d4179f4..f2b139946b5a9 100644 --- a/sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp +++ b/sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp @@ -1,7 +1,7 @@ // REQUIRES: aspect-usm_shared_allocations // // On CPU it segfaults within the kernel that performs virtual function call. -// XFAIL: cpu +// XFAIL: cpu && !spirv-backend // XFAIL-TRACKER: https://github.com/intel/llvm/issues/15080 // UNSUPPORTED: gpu // UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/15068 From f3a2bad722454b2ca3dae2a51990435863e3de16 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Wed, 8 Jan 2025 08:40:39 -0800 Subject: [PATCH 07/17] clang-format --- clang/lib/Driver/ToolChains/Clang.cpp | 3 ++- 1 file changed, 2 insertions(+), 1 deletion(-) diff --git a/clang/lib/Driver/ToolChains/Clang.cpp b/clang/lib/Driver/ToolChains/Clang.cpp index bfadd7b1619e0..33abaf2919c53 100644 --- a/clang/lib/Driver/ToolChains/Clang.cpp +++ b/clang/lib/Driver/ToolChains/Clang.cpp @@ -10714,7 +10714,8 @@ static void getSPIRVBackendOpts(const llvm::opt::ArgList &TCArgs, // Meanwhile we enable all the SPIR-V backend extensions. BackendArgs.push_back(TCArgs.MakeArgString("--spirv-ext=all")); // TODO: - // - handle -Xspirv-translator option to avoid "argument unused during compilation" error + // - handle -Xspirv-translator option to avoid "argument unused during + // compilation" error // - handle --spirv-ext=+ and --spirv-ext=- options } From c52b5caae6ab39d2e6e0a640bddcf79b0902806b Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Wed, 8 Jan 2025 10:01:35 -0800 Subject: [PATCH 08/17] %{build} -> %clangxx in sycl/test/ --- sycl/test/fpga_tests/fpga_io_pipes.cpp | 2 +- sycl/test/non-uniform-groups/is_fixed_topology.cpp | 2 +- sycl/test/non-uniform-groups/is_user_constructed.cpp | 2 +- 3 files changed, 3 insertions(+), 3 deletions(-) diff --git a/sycl/test/fpga_tests/fpga_io_pipes.cpp b/sycl/test/fpga_tests/fpga_io_pipes.cpp index 1eb50966b76ee..29736615f0947 100644 --- a/sycl/test/fpga_tests/fpga_io_pipes.cpp +++ b/sycl/test/fpga_tests/fpga_io_pipes.cpp @@ -1,4 +1,4 @@ -// RUN: %{build} -fsyntax-only -o %t.out +// RUN: %clangxx -fsyntax-only -o %t.out // TODO: launch the test if the feature is supported. //==------------ fpga_io_pipes.cpp - SYCL FPGA pipes test ------------------==// // diff --git a/sycl/test/non-uniform-groups/is_fixed_topology.cpp b/sycl/test/non-uniform-groups/is_fixed_topology.cpp index e8183ce41d3d2..d6363037a2af3 100644 --- a/sycl/test/non-uniform-groups/is_fixed_topology.cpp +++ b/sycl/test/non-uniform-groups/is_fixed_topology.cpp @@ -1,4 +1,4 @@ -// RUN: %{build} -fsyntax-only -o %t.out +// RUN: %clangxx -fsyntax-only -o %t.out #include namespace syclex = sycl::ext::oneapi::experimental; diff --git a/sycl/test/non-uniform-groups/is_user_constructed.cpp b/sycl/test/non-uniform-groups/is_user_constructed.cpp index 44fb8b816aec2..a2f918ba8b181 100644 --- a/sycl/test/non-uniform-groups/is_user_constructed.cpp +++ b/sycl/test/non-uniform-groups/is_user_constructed.cpp @@ -1,4 +1,4 @@ -// RUN: %{build} -fsyntax-only -o %t.out +// RUN: %clangxx -fsyntax-only -o %t.out #include #include From 5bd7b827988a5052e7dfa30583e87d8252bddd56 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Thu, 9 Jan 2025 01:27:48 -0800 Subject: [PATCH 09/17] Fix RUN lines in tests, move FPGA header to tests --- sycl/test/fpga_tests/fpga_io_pipes.cpp | 2 +- sycl/{test-e2e/Basic => test}/fpga_tests/io_pipe_def.h | 0 sycl/test/non-uniform-groups/is_fixed_topology.cpp | 2 +- sycl/test/non-uniform-groups/is_user_constructed.cpp | 2 +- 4 files changed, 3 insertions(+), 3 deletions(-) rename sycl/{test-e2e/Basic => test}/fpga_tests/io_pipe_def.h (100%) diff --git a/sycl/test/fpga_tests/fpga_io_pipes.cpp b/sycl/test/fpga_tests/fpga_io_pipes.cpp index 29736615f0947..0513386f97dd1 100644 --- a/sycl/test/fpga_tests/fpga_io_pipes.cpp +++ b/sycl/test/fpga_tests/fpga_io_pipes.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsyntax-only -o %t.out +// RUN: %clangxx -fsycl -fsyntax-only %s // TODO: launch the test if the feature is supported. //==------------ fpga_io_pipes.cpp - SYCL FPGA pipes test ------------------==// // diff --git a/sycl/test-e2e/Basic/fpga_tests/io_pipe_def.h b/sycl/test/fpga_tests/io_pipe_def.h similarity index 100% rename from sycl/test-e2e/Basic/fpga_tests/io_pipe_def.h rename to sycl/test/fpga_tests/io_pipe_def.h diff --git a/sycl/test/non-uniform-groups/is_fixed_topology.cpp b/sycl/test/non-uniform-groups/is_fixed_topology.cpp index d6363037a2af3..f92dc5b560ea6 100644 --- a/sycl/test/non-uniform-groups/is_fixed_topology.cpp +++ b/sycl/test/non-uniform-groups/is_fixed_topology.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsyntax-only -o %t.out +// RUN: %clangxx -fsycl -fsyntax-only %s #include namespace syclex = sycl::ext::oneapi::experimental; diff --git a/sycl/test/non-uniform-groups/is_user_constructed.cpp b/sycl/test/non-uniform-groups/is_user_constructed.cpp index a2f918ba8b181..d5b7df40662fb 100644 --- a/sycl/test/non-uniform-groups/is_user_constructed.cpp +++ b/sycl/test/non-uniform-groups/is_user_constructed.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsyntax-only -o %t.out +// RUN: %clangxx -fsycl -fsyntax-only %s #include #include From f3ec0ec627faf42274003c8c91f2a1c3a78387ce Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Thu, 9 Jan 2025 02:53:30 -0800 Subject: [PATCH 10/17] Fix clang test for using SPIR-V backend for SYCL offloading --- clang/test/Driver/sycl-spirv-backend.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/clang/test/Driver/sycl-spirv-backend.cpp b/clang/test/Driver/sycl-spirv-backend.cpp index db159f6deafb5..7697c1055b3d2 100644 --- a/clang/test/Driver/sycl-spirv-backend.cpp +++ b/clang/test/Driver/sycl-spirv-backend.cpp @@ -3,4 +3,4 @@ /// // RUN: %clangxx -fsycl -fsycl-use-spirv-backend-for-spirv-gen -### %s 2>&1 | FileCheck %s -// CHECK: llc{{.*}} "-filetype=obj" "-mtriple=spirv64-unknown-unknown" "-O0" "--avoid-spirv-capabilities=Shader" "--translator-compatibility-mode" "--spirv-ext= +// CHECK: llc{{.*}} "-filetype=obj" "-mtriple=spirv64{{[^-]*}}-unknown-unknown" "--avoid-spirv-capabilities=Shader" "--translator-compatibility-mode" "--spirv-ext= From e3871b9dd063596695f785680b0091d40f59d0c7 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Thu, 9 Jan 2025 04:34:28 -0800 Subject: [PATCH 11/17] add SPIR-V Backend testing into SYCL Pre Commit on Linux --- .github/workflows/sycl-linux-precommit.yml | 18 ++++++++++++++++++ 1 file changed, 18 insertions(+) diff --git a/.github/workflows/sycl-linux-precommit.yml b/.github/workflows/sycl-linux-precommit.yml index 9d1825067a661..6dd17a2819fa4 100644 --- a/.github/workflows/sycl-linux-precommit.yml +++ b/.github/workflows/sycl-linux-precommit.yml @@ -127,6 +127,24 @@ jobs: fail-fast: false matrix: include: + - name: E2E tests on Intel PVC/L0 GPU w/ LLVM SPIR-V Backend + runner: '["Linux", "pvc"]' + image: ghcr.io/intel/llvm/ubuntu2404_intel_drivers:latest + image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN + target_devices: level_zero:gpu + extra_lit_opts: -j 50 --param spirv-backend=True + - name: E2E tests on Intel PVC/OCL GPU w/ LLVM SPIR-V Backend + runner: '["Linux", "pvc"]' + image: ghcr.io/intel/llvm/ubuntu2404_intel_drivers:latest + image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN + target_devices: opencl:gpu + extra_lit_opts: -j 50 --param spirv-backend=True + - name: E2E tests on Intel GEN12/OCL CPU w/ LLVM SPIR-V Backend + runner: '["Linux", "gen12"]' + image: ghcr.io/intel/llvm/ubuntu2404_intel_drivers:latest + image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN + target_devices: opencl:cpu + extra_lit_opts: -j 50 --param spirv-backend=True - name: NVIDIA/CUDA runner: '["Linux", "cuda"]' image: ghcr.io/intel/llvm/ubuntu2404_build:latest From fcdf61955374dfc3eeda8ebc602462ec076d7b75 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Thu, 9 Jan 2025 08:57:37 -0800 Subject: [PATCH 12/17] update XFAIL & UNSUPPORTED using actual versions of SPIR-V Backend and run-time drivers in intel/llvm CI --- sycl/test-e2e/AddressCast/dynamic_address_cast.cpp | 4 ++++ sycl/test-e2e/BFloat16/bfloat16_builtins.cpp | 5 +++++ .../device_architecture_comparison_on_device_aot.cpp | 2 +- sycl/test-e2e/DeviceCodeSplit/grf.cpp | 2 +- sycl/test-e2e/DeviceLib/imf_fp16_trivial_test.cpp | 4 ++++ sycl/test-e2e/DeviceLib/imf_fp64_rounding_test.cpp | 4 ++++ sycl/test-e2e/DeviceLib/imf_half_type_cast.cpp | 4 ++++ .../SYCL2020/group_sort/group_and_joint_sort.cpp | 5 +++++ .../KernelAndProgram/kernel-bundle-merge-options-env.cpp | 4 ++++ .../KernelAndProgram/kernel-bundle-merge-options.cpp | 4 ++++ .../KernelAndProgram/persistent-cache-multi-device.cpp | 4 ++++ sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp | 5 +++++ sycl/test-e2e/Matrix/lit.local.cfg | 4 ++++ .../ProgramManager/multi_device_bundle/build_twice.cpp | 4 ++++ .../multi_device_bundle/device_libs_and_caching.cpp | 4 ++++ .../PropagateOptionsToBackend/sycl-opt-level-level-zero.cpp | 4 ++++ sycl/test-e2e/Reduction/reduction_nd_N_vars.cpp | 4 ++++ sycl/test-e2e/Reduction/reduction_span.cpp | 5 +++++ sycl/test-e2e/Reduction/reduction_span_pack.cpp | 4 ++++ sycl/test-e2e/Regression/group_load_fortified.cpp | 4 ++++ sycl/test-e2e/Regression/multithread_write_accessor.cpp | 2 +- sycl/test-e2e/SubGroup/sub_group_as.cpp | 4 ++++ sycl/test-e2e/SubGroup/sub_group_as_vec.cpp | 4 ++++ sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp | 2 +- sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp | 2 +- 25 files changed, 89 insertions(+), 5 deletions(-) create mode 100644 sycl/test-e2e/Matrix/lit.local.cfg diff --git a/sycl/test-e2e/AddressCast/dynamic_address_cast.cpp b/sycl/test-e2e/AddressCast/dynamic_address_cast.cpp index 441fe486564b3..616cb801306b1 100644 --- a/sycl/test-e2e/AddressCast/dynamic_address_cast.cpp +++ b/sycl/test-e2e/AddressCast/dynamic_address_cast.cpp @@ -13,6 +13,10 @@ // RUN: %{build} -o %t.out // RUN: %{run} %t.out +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && gpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include #include diff --git a/sycl/test-e2e/BFloat16/bfloat16_builtins.cpp b/sycl/test-e2e/BFloat16/bfloat16_builtins.cpp index 9c69e0cd7bf71..74e9f327d4393 100644 --- a/sycl/test-e2e/BFloat16/bfloat16_builtins.cpp +++ b/sycl/test-e2e/BFloat16/bfloat16_builtins.cpp @@ -12,6 +12,11 @@ // RUN: %if preview-breaking-changes-supported %{ %clangxx -fsycl -fpreview-breaking-changes -fsycl-targets=%{sycl_triple} %if any-device-is-cuda %{ -Xsycl-target-backend=nvptx64-nvidia-cuda --cuda-gpu-arch=sm_80 %} %s -o %t2.out %{mathflags} %} // RUN: %if preview-breaking-changes-supported %{ %{run} %t2.out %} +// Flaky timeout on CPU. Enable when fixed. +// Depends on SPIR-V Backend & run-time drivers version. +// UNSUPPORTED: spirv-backend && cpu +// UNSUPPORTED-TRACKER: CMPLRLLVM-64705 + #include "bfloat16_builtins.hpp" int main() { diff --git a/sycl/test-e2e/DeviceArchitecture/device_architecture_comparison_on_device_aot.cpp b/sycl/test-e2e/DeviceArchitecture/device_architecture_comparison_on_device_aot.cpp index 88f55d00aa903..3db3bde191ce5 100644 --- a/sycl/test-e2e/DeviceArchitecture/device_architecture_comparison_on_device_aot.cpp +++ b/sycl/test-e2e/DeviceArchitecture/device_architecture_comparison_on_device_aot.cpp @@ -1,6 +1,6 @@ // REQUIRES: arch-intel_gpu_pvc, ocloc -// XFAIL: arch-intel_gpu_pvc +// XFAIL: arch-intel_gpu_pvc && !spirv-backend // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16401 // RUN: %clangxx -fsycl -fsycl-targets=intel_gpu_pvc %s -o %t.out diff --git a/sycl/test-e2e/DeviceCodeSplit/grf.cpp b/sycl/test-e2e/DeviceCodeSplit/grf.cpp index 1e5b085d207d6..f750c2350b0e1 100644 --- a/sycl/test-e2e/DeviceCodeSplit/grf.cpp +++ b/sycl/test-e2e/DeviceCodeSplit/grf.cpp @@ -14,7 +14,7 @@ // compiler option // REQUIRES: arch-intel_gpu_pvc -// XFAIL: arch-intel_gpu_pvc +// XFAIL: arch-intel_gpu_pvc && !(spirv-backend && opencl) // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16401 // RUN: %{build} -Wno-error=deprecated-declarations -o %t1.out diff --git a/sycl/test-e2e/DeviceLib/imf_fp16_trivial_test.cpp b/sycl/test-e2e/DeviceLib/imf_fp16_trivial_test.cpp index 24ea52f391695..5006933ca3c73 100644 --- a/sycl/test-e2e/DeviceLib/imf_fp16_trivial_test.cpp +++ b/sycl/test-e2e/DeviceLib/imf_fp16_trivial_test.cpp @@ -13,6 +13,10 @@ // Windows doesn't yet have full shutdown(). // UNSUPPORTED: ze_debug && windows +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && gpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include "imf_utils.hpp" #include #include diff --git a/sycl/test-e2e/DeviceLib/imf_fp64_rounding_test.cpp b/sycl/test-e2e/DeviceLib/imf_fp64_rounding_test.cpp index 4952d959988a9..82d925588850a 100644 --- a/sycl/test-e2e/DeviceLib/imf_fp64_rounding_test.cpp +++ b/sycl/test-e2e/DeviceLib/imf_fp64_rounding_test.cpp @@ -7,6 +7,10 @@ // // UNSUPPORTED: cuda || hip +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include "imf_utils.hpp" #include diff --git a/sycl/test-e2e/DeviceLib/imf_half_type_cast.cpp b/sycl/test-e2e/DeviceLib/imf_half_type_cast.cpp index 2491def1cff18..3b613f1bae300 100644 --- a/sycl/test-e2e/DeviceLib/imf_half_type_cast.cpp +++ b/sycl/test-e2e/DeviceLib/imf_half_type_cast.cpp @@ -12,6 +12,10 @@ // Windows doesn't yet have full shutdown(). // UNSUPPORTED: ze_debug && windows +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include "imf_utils.hpp" #include diff --git a/sycl/test-e2e/GroupAlgorithm/SYCL2020/group_sort/group_and_joint_sort.cpp b/sycl/test-e2e/GroupAlgorithm/SYCL2020/group_sort/group_and_joint_sort.cpp index a5ca0006b79ed..ab7ff36dd7c19 100644 --- a/sycl/test-e2e/GroupAlgorithm/SYCL2020/group_sort/group_and_joint_sort.cpp +++ b/sycl/test-e2e/GroupAlgorithm/SYCL2020/group_sort/group_and_joint_sort.cpp @@ -2,6 +2,11 @@ // RUN: %{build} -fsycl-device-code-split=per_kernel -o %t.out // RUN: %{run} %t.out +// Timeout on CPU. Enable when fixed. +// Depends on SPIR-V Backend & run-time drivers version. +// UNSUPPORTED: spirv-backend && cpu +// UNSUPPORTED-TRACKER: CMPLRLLVM-64705 + // The test verifies sort API extension. // Currently it checks the following combinations: // For number of elements {18, 64} diff --git a/sycl/test-e2e/KernelAndProgram/kernel-bundle-merge-options-env.cpp b/sycl/test-e2e/KernelAndProgram/kernel-bundle-merge-options-env.cpp index aec55eceddfe6..79b593270412f 100644 --- a/sycl/test-e2e/KernelAndProgram/kernel-bundle-merge-options-env.cpp +++ b/sycl/test-e2e/KernelAndProgram/kernel-bundle-merge-options-env.cpp @@ -7,6 +7,10 @@ // RUN: env SYCL_UR_TRACE=2 SYCL_PROGRAM_COMPILE_OPTIONS=-DENV_COMPILE_OPTS SYCL_PROGRAM_LINK_OPTIONS=-DENV_LINK_OPTS SYCL_PROGRAM_APPEND_COMPILE_OPTIONS=-DENV_APPEND_COMPILE_OPTS SYCL_PROGRAM_APPEND_LINK_OPTIONS=-DENV_APPEND_LINK_OPTS %{run} %t2.out | FileCheck %s // UNSUPPORTED: hip +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && gpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include "kernel-bundle-merge-options.hpp" // CHECK: <--- urProgramBuild{{.*}}{{[^bar]*}}-DENV_COMPILE_OPTS -DENV_APPEND_COMPILE_OPTS{{[^bar]*}}-DENV_LINK_OPTS -DENV_APPEND_LINK_OPTS{{[^bar]*}} diff --git a/sycl/test-e2e/KernelAndProgram/kernel-bundle-merge-options.cpp b/sycl/test-e2e/KernelAndProgram/kernel-bundle-merge-options.cpp index 10037fe75f6b4..b713f4d9064d8 100644 --- a/sycl/test-e2e/KernelAndProgram/kernel-bundle-merge-options.cpp +++ b/sycl/test-e2e/KernelAndProgram/kernel-bundle-merge-options.cpp @@ -3,6 +3,10 @@ // RUN: env SYCL_UR_TRACE=2 %{run} %t.out | FileCheck %s // UNSUPPORTED: hip +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && gpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + // Note that the UR call might be urProgramBuild OR urProgramBuildExp . // The same is true for Compile and Link. // We want the first match. Don't put parentheses after. diff --git a/sycl/test-e2e/KernelAndProgram/persistent-cache-multi-device.cpp b/sycl/test-e2e/KernelAndProgram/persistent-cache-multi-device.cpp index 48a9bdf74d8cf..131c78e40026c 100644 --- a/sycl/test-e2e/KernelAndProgram/persistent-cache-multi-device.cpp +++ b/sycl/test-e2e/KernelAndProgram/persistent-cache-multi-device.cpp @@ -5,6 +5,10 @@ // RUN: env NEOReadDebugKeys=1 CreateMultipleRootDevices=3 SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_TRACE=1 SYCL_CACHE_DIR=%t/cache_dir env -u XDG_CACHE_HOME env -u HOME %{run} %t.out 2>&1 | FileCheck %s --check-prefixes=CHECK-BUILD // RUN: env NEOReadDebugKeys=1 CreateMultipleRootDevices=3 SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_TRACE=1 SYCL_CACHE_DIR=%t/cache_dir env -u XDG_CACHE_HOME env -u HOME %{run} %t.out 2>&1 | FileCheck %s --check-prefixes=CHECK-CACHE +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64705 + // Test checks that persistent cache works correctly with multiple devices. #include diff --git a/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp b/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp index d8b020e971d4a..c35d6c408265f 100644 --- a/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp +++ b/sycl/test-e2e/KernelCompiler/kernel_compiler_sycl.cpp @@ -9,6 +9,11 @@ // REQUIRES: (opencl || level_zero) // UNSUPPORTED: accelerator +// Flaky timeout on CPU. Enable when fixed. +// Depends on SPIR-V Backend & run-time drivers version. +// UNSUPPORTED: spirv-backend && cpu +// UNSUPPORTED-TRACKER: CMPLRLLVM-64705 + // -- Test the kernel_compiler with SYCL source. // RUN: %{build} -o %t.out diff --git a/sycl/test-e2e/Matrix/lit.local.cfg b/sycl/test-e2e/Matrix/lit.local.cfg new file mode 100644 index 0000000000000..fe942ed141799 --- /dev/null +++ b/sycl/test-e2e/Matrix/lit.local.cfg @@ -0,0 +1,4 @@ +# SPIR-V Backend is in process of developing support for this feature. +# At the moment support of the feature depends on SPIR-V Backend & run-time +# drivers version, so we temporarily mark it as unsupported (CMPLRLLVM-64705). +config.unsupported_features += ['spirv-backend'] diff --git a/sycl/test-e2e/ProgramManager/multi_device_bundle/build_twice.cpp b/sycl/test-e2e/ProgramManager/multi_device_bundle/build_twice.cpp index d4aaea78c7173..12906f2e10539 100644 --- a/sycl/test-e2e/ProgramManager/multi_device_bundle/build_twice.cpp +++ b/sycl/test-e2e/ProgramManager/multi_device_bundle/build_twice.cpp @@ -6,6 +6,10 @@ // RUN: %{build} -o %t.out // RUN: env NEOReadDebugKeys=1 CreateMultipleRootDevices=3 SYCL_UR_TRACE=2 %{run} %t.out | FileCheck %s +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include #include diff --git a/sycl/test-e2e/ProgramManager/multi_device_bundle/device_libs_and_caching.cpp b/sycl/test-e2e/ProgramManager/multi_device_bundle/device_libs_and_caching.cpp index 1534b2330f552..2a54744ebdef5 100644 --- a/sycl/test-e2e/ProgramManager/multi_device_bundle/device_libs_and_caching.cpp +++ b/sycl/test-e2e/ProgramManager/multi_device_bundle/device_libs_and_caching.cpp @@ -29,6 +29,10 @@ // Check the case when in-memory caching of the programs is disabled. // RUN: env SYCL_CACHE_IN_MEM=0 NEOReadDebugKeys=1 CreateMultipleRootDevices=4 %{run} %t.out +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include #include #include diff --git a/sycl/test-e2e/PropagateOptionsToBackend/sycl-opt-level-level-zero.cpp b/sycl/test-e2e/PropagateOptionsToBackend/sycl-opt-level-level-zero.cpp index 735f25e77c528..1be87a4a26749 100644 --- a/sycl/test-e2e/PropagateOptionsToBackend/sycl-opt-level-level-zero.cpp +++ b/sycl/test-e2e/PropagateOptionsToBackend/sycl-opt-level-level-zero.cpp @@ -9,6 +9,10 @@ // RUN: %{build} -O3 -o %t3.out // RUN: env SYCL_UR_TRACE=2 %{run} %t3.out 2>&1 | FileCheck %s --check-prefixes=CHECK1 +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend +// XFAIL-TRACKER: CMPLRLLVM-64705 + // This test verifies the propagation of front-end compiler optimization // option to the backend. // API call in device code: diff --git a/sycl/test-e2e/Reduction/reduction_nd_N_vars.cpp b/sycl/test-e2e/Reduction/reduction_nd_N_vars.cpp index 4ea1fe420e31c..ad65c42fccf70 100644 --- a/sycl/test-e2e/Reduction/reduction_nd_N_vars.cpp +++ b/sycl/test-e2e/Reduction/reduction_nd_N_vars.cpp @@ -4,6 +4,10 @@ // Windows doesn't yet have full shutdown(). // UNSUPPORTED: ze_debug && windows +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && gpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + // This test checks handling of parallel_for() accepting nd_range and // two or more reductions. diff --git a/sycl/test-e2e/Reduction/reduction_span.cpp b/sycl/test-e2e/Reduction/reduction_span.cpp index 668399f00a212..2d07c8d3e1039 100644 --- a/sycl/test-e2e/Reduction/reduction_span.cpp +++ b/sycl/test-e2e/Reduction/reduction_span.cpp @@ -5,6 +5,11 @@ // UNSUPPORTED: ze_debug && windows // This test performs basic checks of reductions initialized with a sycl::span +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && cpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + + #include #include diff --git a/sycl/test-e2e/Reduction/reduction_span_pack.cpp b/sycl/test-e2e/Reduction/reduction_span_pack.cpp index 023e78fe5e85d..a5063282e0e74 100644 --- a/sycl/test-e2e/Reduction/reduction_span_pack.cpp +++ b/sycl/test-e2e/Reduction/reduction_span_pack.cpp @@ -6,6 +6,10 @@ // This test performs basic checks of reductions initialized with a pack // containing at least one sycl::span +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && cpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include #include diff --git a/sycl/test-e2e/Regression/group_load_fortified.cpp b/sycl/test-e2e/Regression/group_load_fortified.cpp index ffa067221752f..c7b1287548a38 100644 --- a/sycl/test-e2e/Regression/group_load_fortified.cpp +++ b/sycl/test-e2e/Regression/group_load_fortified.cpp @@ -2,6 +2,10 @@ // RUN: %{build} -D_FORTIFY_SOURCE=2 -o %t.out // RUN: %{run} %t.out +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && cpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + // Checks that group_load runs even when the source code is fortified. This // failed at one point due to the use of std::memcpy in the implementation, // which would hold an assert in device code when fortified, which would fail diff --git a/sycl/test-e2e/Regression/multithread_write_accessor.cpp b/sycl/test-e2e/Regression/multithread_write_accessor.cpp index 87299ed3e4d5c..053a6ddca5114 100644 --- a/sycl/test-e2e/Regression/multithread_write_accessor.cpp +++ b/sycl/test-e2e/Regression/multithread_write_accessor.cpp @@ -1,7 +1,7 @@ // RUN: %{build} -o %t.out %threads_lib // RUN: %{run} %t.out -// XFAIL: arch-intel_gpu_pvc +// XFAIL: arch-intel_gpu_pvc && !spirv-backend // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16401 #include diff --git a/sycl/test-e2e/SubGroup/sub_group_as.cpp b/sycl/test-e2e/SubGroup/sub_group_as.cpp index 67b6102ff8300..2f48cc7489a56 100644 --- a/sycl/test-e2e/SubGroup/sub_group_as.cpp +++ b/sycl/test-e2e/SubGroup/sub_group_as.cpp @@ -4,6 +4,10 @@ // RUN: %{build} -DUSE_DEPRECATED_LOCAL_ACC -o %t2.out -Wno-deprecated-declarations // RUN: %{run} %t2.out +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && gpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include #include #include diff --git a/sycl/test-e2e/SubGroup/sub_group_as_vec.cpp b/sycl/test-e2e/SubGroup/sub_group_as_vec.cpp index 210c147d6a776..d58bc16d6310d 100644 --- a/sycl/test-e2e/SubGroup/sub_group_as_vec.cpp +++ b/sycl/test-e2e/SubGroup/sub_group_as_vec.cpp @@ -4,6 +4,10 @@ // RUN: %{build} -DUSE_DEPRECATED_LOCAL_ACC -o %t2.out -Wno-deprecated-declarations // RUN: %{run} %t2.out +// Depends on SPIR-V Backend & run-time drivers version. +// XFAIL: spirv-backend && gpu +// XFAIL-TRACKER: CMPLRLLVM-64705 + #include "helper.hpp" #include #include diff --git a/sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp b/sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp index f2b139946b5a9..f01a25d4179f4 100644 --- a/sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp +++ b/sycl/test-e2e/VirtualFunctions/misc/group-barrier.cpp @@ -1,7 +1,7 @@ // REQUIRES: aspect-usm_shared_allocations // // On CPU it segfaults within the kernel that performs virtual function call. -// XFAIL: cpu && !spirv-backend +// XFAIL: cpu // XFAIL-TRACKER: https://github.com/intel/llvm/issues/15080 // UNSUPPORTED: gpu // UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/15068 diff --git a/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp b/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp index 033f5c99d74e1..27664c05f4936 100644 --- a/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp +++ b/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp @@ -26,7 +26,7 @@ // UNSUPPORTED: linux && opencl && (gpu-intel-gen12 || gpu-intel-dg2) // UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/15275 -// XFAIL: arch-intel_gpu_pvc +// XFAIL: arch-intel_gpu_pvc && !(spirv-backend && opencl) // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16401 #include From 7e20e3dab0d49dae589417364b625a712e835421 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Fri, 10 Jan 2025 01:33:48 -0800 Subject: [PATCH 13/17] fix XFAIL; clang-format --- sycl/test-e2e/DeviceCodeSplit/grf.cpp | 2 +- sycl/test-e2e/Reduction/reduction_span.cpp | 1 - sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp | 2 +- 3 files changed, 2 insertions(+), 3 deletions(-) diff --git a/sycl/test-e2e/DeviceCodeSplit/grf.cpp b/sycl/test-e2e/DeviceCodeSplit/grf.cpp index f750c2350b0e1..1e5b085d207d6 100644 --- a/sycl/test-e2e/DeviceCodeSplit/grf.cpp +++ b/sycl/test-e2e/DeviceCodeSplit/grf.cpp @@ -14,7 +14,7 @@ // compiler option // REQUIRES: arch-intel_gpu_pvc -// XFAIL: arch-intel_gpu_pvc && !(spirv-backend && opencl) +// XFAIL: arch-intel_gpu_pvc // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16401 // RUN: %{build} -Wno-error=deprecated-declarations -o %t1.out diff --git a/sycl/test-e2e/Reduction/reduction_span.cpp b/sycl/test-e2e/Reduction/reduction_span.cpp index 2d07c8d3e1039..fdb70ec8f5734 100644 --- a/sycl/test-e2e/Reduction/reduction_span.cpp +++ b/sycl/test-e2e/Reduction/reduction_span.cpp @@ -9,7 +9,6 @@ // XFAIL: spirv-backend && cpu // XFAIL-TRACKER: CMPLRLLVM-64705 - #include #include diff --git a/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp b/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp index 27664c05f4936..033f5c99d74e1 100644 --- a/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp +++ b/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp @@ -26,7 +26,7 @@ // UNSUPPORTED: linux && opencl && (gpu-intel-gen12 || gpu-intel-dg2) // UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/15275 -// XFAIL: arch-intel_gpu_pvc && !(spirv-backend && opencl) +// XFAIL: arch-intel_gpu_pvc // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16401 #include From f83e98f2fc1b25260ccca2a9cd6d075bd45f3dd4 Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Fri, 10 Jan 2025 02:25:43 -0800 Subject: [PATCH 14/17] UNSUPPORT flaky tests --- sycl/test-e2e/DeviceCodeSplit/grf.cpp | 4 ++++ sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp | 4 ++++ 2 files changed, 8 insertions(+) diff --git a/sycl/test-e2e/DeviceCodeSplit/grf.cpp b/sycl/test-e2e/DeviceCodeSplit/grf.cpp index 1e5b085d207d6..aa25f5b0fde77 100644 --- a/sycl/test-e2e/DeviceCodeSplit/grf.cpp +++ b/sycl/test-e2e/DeviceCodeSplit/grf.cpp @@ -17,6 +17,10 @@ // XFAIL: arch-intel_gpu_pvc // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16401 +// Flaky pass/fail behaviour. +// UNSUPPORTED: spirv-backend +// UNSUPPORTED-TRACKER: CMPLRLLVM-64705 + // RUN: %{build} -Wno-error=deprecated-declarations -o %t1.out // RUN: env SYCL_UR_TRACE=2 %{run} %t1.out 2>&1 | FileCheck %s --check-prefixes=CHECK,CHECK-NO-VAR // RUN: env SYCL_PROGRAM_COMPILE_OPTIONS="-g" SYCL_UR_TRACE=2 %{run} %t1.out 2>&1 | FileCheck %s --check-prefixes=CHECK,CHECK-WITH-VAR diff --git a/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp b/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp index 033f5c99d74e1..fdf3a10fdc586 100644 --- a/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp +++ b/sycl/test-e2e/syclcompat/launch/launch_policy_lmem.cpp @@ -29,6 +29,10 @@ // XFAIL: arch-intel_gpu_pvc // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16401 +// Flaky pass/fail behaviour. +// UNSUPPORTED: spirv-backend +// UNSUPPORTED-TRACKER: CMPLRLLVM-64705 + #include #include #include From 92e3ac7b6592ed50fceed710a5e9caa935c29f2d Mon Sep 17 00:00:00 2001 From: "Levytskyy, Vyacheslav" Date: Mon, 13 Jan 2025 02:57:03 -0800 Subject: [PATCH 15/17] revert chanches in CI and Clang/Driver --- .github/workflows/sycl-linux-precommit.yml | 18 ---------- .github/workflows/sycl-nightly.yml | 26 --------------- clang/lib/Driver/ToolChains/Clang.cpp | 39 ++++++++++++++++------ clang/test/Driver/sycl-spirv-backend.cpp | 2 +- sycl/test-e2e/README.md | 6 ---- sycl/test-e2e/format.py | 10 +----- sycl/test-e2e/lit.cfg.py | 2 -- 7 files changed, 30 insertions(+), 73 deletions(-) diff --git a/.github/workflows/sycl-linux-precommit.yml b/.github/workflows/sycl-linux-precommit.yml index 6dd17a2819fa4..9d1825067a661 100644 --- a/.github/workflows/sycl-linux-precommit.yml +++ b/.github/workflows/sycl-linux-precommit.yml @@ -127,24 +127,6 @@ jobs: fail-fast: false matrix: include: - - name: E2E tests on Intel PVC/L0 GPU w/ LLVM SPIR-V Backend - runner: '["Linux", "pvc"]' - image: ghcr.io/intel/llvm/ubuntu2404_intel_drivers:latest - image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN - target_devices: level_zero:gpu - extra_lit_opts: -j 50 --param spirv-backend=True - - name: E2E tests on Intel PVC/OCL GPU w/ LLVM SPIR-V Backend - runner: '["Linux", "pvc"]' - image: ghcr.io/intel/llvm/ubuntu2404_intel_drivers:latest - image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN - target_devices: opencl:gpu - extra_lit_opts: -j 50 --param spirv-backend=True - - name: E2E tests on Intel GEN12/OCL CPU w/ LLVM SPIR-V Backend - runner: '["Linux", "gen12"]' - image: ghcr.io/intel/llvm/ubuntu2404_intel_drivers:latest - image_options: -u 1001 --device=/dev/dri --privileged --cap-add SYS_ADMIN - target_devices: opencl:cpu - extra_lit_opts: -j 50 --param spirv-backend=True - name: NVIDIA/CUDA runner: '["Linux", "cuda"]' image: ghcr.io/intel/llvm/ubuntu2404_build:latest diff --git a/.github/workflows/sycl-nightly.yml b/.github/workflows/sycl-nightly.yml index f2ced79c42f75..5c6b6ad89f57d 100644 --- a/.github/workflows/sycl-nightly.yml +++ b/.github/workflows/sycl-nightly.yml @@ -87,32 +87,6 @@ jobs: image_options: -u 1001 target_devices: opencl:cpu tests_selector: e2e - - - name: Intel L0 GPU w/ LLVM SPIR-V Backend - runner: '["Linux", "gen12"]' - image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest - image_options: -u 1001 --device=/dev/dri -v /dev/dri/by-path:/dev/dri/by-path --privileged --cap-add SYS_ADMIN - target_devices: level_zero:gpu - reset_intel_gpu: true - tests_selector: e2e - extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True - - - name: Intel OCL GPU w/ LLVM SPIR-V Backend - runner: '["Linux", "gen12"]' - image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest - image_options: -u 1001 --device=/dev/dri -v /dev/dri/by-path:/dev/dri/by-path --privileged --cap-add SYS_ADMIN - target_devices: opencl:gpu - reset_intel_gpu: true - tests_selector: e2e - extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True - - - name: OCL CPU (Intel/GEN12) w/ LLVM SPIR-V Backend - runner: '["Linux", "gen12"]' - image: ghcr.io/intel/llvm/ubuntu2204_intel_drivers:latest - image_options: -u 1001 --privileged --cap-add SYS_ADMIN - target_devices: opencl:cpu - tests_selector: e2e - extra_lit_opts: --param spirv-backend=True --param gpu-intel-gen12=True uses: ./.github/workflows/sycl-linux-run-tests.yml with: name: ${{ matrix.name }} diff --git a/clang/lib/Driver/ToolChains/Clang.cpp b/clang/lib/Driver/ToolChains/Clang.cpp index 33abaf2919c53..832733eb2bda8 100644 --- a/clang/lib/Driver/ToolChains/Clang.cpp +++ b/clang/lib/Driver/ToolChains/Clang.cpp @@ -10702,21 +10702,38 @@ static void getSPIRVBackendOpts(const llvm::opt::ArgList &TCArgs, ArgStringList &BackendArgs) { BackendArgs.push_back(TCArgs.MakeArgString("-filetype=obj")); BackendArgs.push_back( - TCArgs.MakeArgString("-mtriple=spirv64v1.6-unknown-unknown")); + TCArgs.MakeArgString("-mtriple=spirv64-unknown-unknown")); + // TODO: Optimization level is currently forced to -O0 due to some testing + // issues. Update optimization level after testing issues are resolved. + BackendArgs.push_back(TCArgs.MakeArgString("-O0")); BackendArgs.push_back( TCArgs.MakeArgString("--avoid-spirv-capabilities=Shader")); BackendArgs.push_back( TCArgs.MakeArgString("--translator-compatibility-mode")); - // TODO: A list of SPIR-V extensions that are supported by the SPIR-V backend - // is growing. Let's postpone the decision on which extensions to enable until - // - the list is stable, and - // - we decide on a mapping of user requested extensions into backend's ones. - // Meanwhile we enable all the SPIR-V backend extensions. - BackendArgs.push_back(TCArgs.MakeArgString("--spirv-ext=all")); - // TODO: - // - handle -Xspirv-translator option to avoid "argument unused during - // compilation" error - // - handle --spirv-ext=+ and --spirv-ext=- options + + // TODO: There is some overlap between the lists of extensions in SPIR-V + // backend and SPIR-V Trnaslator). We will try to combine them when SPIR-V + // backdn is ready. + std::string ExtArg("--spirv-ext="); + std::string DefaultExtArg = + "+SPV_EXT_shader_atomic_float_add,+SPV_EXT_shader_atomic_float_min_max" + ",+SPV_KHR_no_integer_wrap_decoration,+SPV_KHR_float_controls" + ",+SPV_KHR_expect_assume,+SPV_KHR_linkonce_odr"; + std::string INTELExtArg = ",+SPV_INTEL_subgroups,+SPV_INTEL_function_pointers" + ",+SPV_INTEL_arbitrary_precision_integers" + ",+SPV_INTEL_variable_length_array"; + ExtArg = ExtArg + DefaultExtArg + INTELExtArg; + + // Other args + ExtArg += ",+SPV_INTEL_bfloat16_conversion" + ",+SPV_KHR_uniform_group_instructions" + ",+SPV_INTEL_optnone" + ",+SPV_KHR_subgroup_rotate" + ",+SPV_INTEL_usm_storage_classes" + ",+SPV_EXT_shader_atomic_float16_add" + ",+SPV_KHR_bit_instructions"; + + BackendArgs.push_back(TCArgs.MakeArgString(ExtArg)); } // Utility function to gather all llvm-spirv options. diff --git a/clang/test/Driver/sycl-spirv-backend.cpp b/clang/test/Driver/sycl-spirv-backend.cpp index 7697c1055b3d2..db159f6deafb5 100644 --- a/clang/test/Driver/sycl-spirv-backend.cpp +++ b/clang/test/Driver/sycl-spirv-backend.cpp @@ -3,4 +3,4 @@ /// // RUN: %clangxx -fsycl -fsycl-use-spirv-backend-for-spirv-gen -### %s 2>&1 | FileCheck %s -// CHECK: llc{{.*}} "-filetype=obj" "-mtriple=spirv64{{[^-]*}}-unknown-unknown" "--avoid-spirv-capabilities=Shader" "--translator-compatibility-mode" "--spirv-ext= +// CHECK: llc{{.*}} "-filetype=obj" "-mtriple=spirv64-unknown-unknown" "-O0" "--avoid-spirv-capabilities=Shader" "--translator-compatibility-mode" "--spirv-ext= diff --git a/sycl/test-e2e/README.md b/sycl/test-e2e/README.md index 508c35ef98afb..396aa9ef7341a 100644 --- a/sycl/test-e2e/README.md +++ b/sycl/test-e2e/README.md @@ -256,12 +256,6 @@ place. No new tests should use these features: * **gpu-intel-pvc** - Intel GPU PVC availability; * **gpu-intel-pvc-vg** - Intel GPU PVC-VG availability; -### Use the LLVM SPIR-V Backend to generate SPIR-V code - -It's possible to use the LLVM SPIR-V Backend instead of llvm-spirv tool to -convert LLVM IR to SPIR-V. This feature must be set manually by passing the -`spirv-backend` argument to `llvm-lit`. - ### llvm-lit parameters Following options can be passed to llvm-lit tool through --param option to diff --git a/sycl/test-e2e/format.py b/sycl/test-e2e/format.py index aaabb02c22f94..a69810145507d 100644 --- a/sycl/test-e2e/format.py +++ b/sycl/test-e2e/format.py @@ -175,21 +175,13 @@ def execute(self, test, litConfig): substitutions = lit.TestRunner.getDefaultSubstitutions(test, tmpDir, tmpBase) substitutions.append(("%{sycl_triple}", format(",".join(triples)))) - # -fsycl-use-spirv-backend-for-spirv-gen is needed - # to support "spirv-backend" feature - fsycl_use_spirv_backend = ( - " -fsycl-use-spirv-backend-for-spirv-gen" - if "spirv-backend" in test.config.available_features - else "" - ) # -fsycl-targets is needed for CUDA/HIP, so just use it be default so # -that new tests by default would runnable there (unless they have # -other restrictions). substitutions.append( ( "%{build}", - "%clangxx -fsycl -fsycl-targets=%{sycl_triple} %verbose_print %s" - + fsycl_use_spirv_backend, + "%clangxx -fsycl -fsycl-targets=%{sycl_triple} %verbose_print %s", ) ) if platform.system() == "Windows": diff --git a/sycl/test-e2e/lit.cfg.py b/sycl/test-e2e/lit.cfg.py index 5550722ebab19..d723d3452b7d6 100644 --- a/sycl/test-e2e/lit.cfg.py +++ b/sycl/test-e2e/lit.cfg.py @@ -225,8 +225,6 @@ def check_igc_tag_and_add_feature(): if lit_config.params.get("enable-perf-tests", False): config.available_features.add("enable-perf-tests") -if lit_config.params.get("spirv-backend", False): - config.available_features.add("spirv-backend") # Use this to make sure that any dynamic checks below are done in the build # directory and not where the sources are located. This is important for the From 45ac9546fc670f99a7c05cc560ef6e8706871528 Mon Sep 17 00:00:00 2001 From: Vyacheslav Levytskyy Date: Thu, 16 Jan 2025 17:55:20 +0100 Subject: [PATCH 16/17] Update sycl/test-e2e/InvokeSimd/lit.local.cfg Co-authored-by: Nick Sarnie --- sycl/test-e2e/InvokeSimd/lit.local.cfg | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test-e2e/InvokeSimd/lit.local.cfg b/sycl/test-e2e/InvokeSimd/lit.local.cfg index 054df935c362b..29a87ab4a0a70 100644 --- a/sycl/test-e2e/InvokeSimd/lit.local.cfg +++ b/sycl/test-e2e/InvokeSimd/lit.local.cfg @@ -7,5 +7,5 @@ config.required_features += ['gpu'] if platform.system() != "Linux": config.unsupported = True -# At the moment SPIR-V Backend has no plans to suport this feature. +# At the moment SPIR-V Backend has no plans to support this feature. config.unsupported_features += ['spirv-backend'] From d23d5eb6a32caf1b8154c1bd30484a5773628dd0 Mon Sep 17 00:00:00 2001 From: Vyacheslav Levytskyy Date: Thu, 16 Jan 2025 17:55:31 +0100 Subject: [PATCH 17/17] Update sycl/test-e2e/ESIMD/lit.local.cfg Co-authored-by: Nick Sarnie --- sycl/test-e2e/ESIMD/lit.local.cfg | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test-e2e/ESIMD/lit.local.cfg b/sycl/test-e2e/ESIMD/lit.local.cfg index f40b518bebe3a..fe0857dabc270 100644 --- a/sycl/test-e2e/ESIMD/lit.local.cfg +++ b/sycl/test-e2e/ESIMD/lit.local.cfg @@ -22,5 +22,5 @@ for substitution in config.substitutions: config.substitutions.insert(0, ("%clangxx", original_clangxx+" -Wno-error=deprecated-declarations")) -# At the moment SPIR-V Backend has no plans to suport this feature. +# At the moment SPIR-V Backend has no plans to support this feature. config.unsupported_features += ['spirv-backend']