[E2E] Introduce SPIR-V Backend support in SYCL tests and end-to-end tests (#16317)

Introduce SPIR-V Backend support in SYCL tests and end-to-end tests by
adding XFAIL & UNSUPPORTED to tests that are not passed at the moment
when Khronos Translator is substituted by SPIR-V Backend. Three
syntax-only tests are moved from e2e to tests.

---------

Co-authored-by: Nick Sarnie <nick.sarnie@intel.com>

Author: VyacheslavLevytskyy

sycl/test-e2e/AddressCast/dynamic_address_cast.cpp

@@ -12,6 +12,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 <sycl/detail/core.hpp>
 
 #include <sycl/ext/oneapi/experimental/address_cast.hpp>

sycl/test-e2e/AddressSanitizer/lit.local.cfg

@@ -13,3 +13,6 @@ if "-fsanitize=memory" in config.cxx_flags:
 
 # https://github.com/intel/llvm/issues/15953
 config.unsupported_features += ['gpu-intel-gen12']
+
+# CMPLRLLVM-64052
+config.unsupported_features += ['spirv-backend']

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 <sycl/detail/core.hpp>
 
 #include <sycl/ext/oneapi/device_global/device_global.hpp>

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 <sycl/usm.hpp>
 
 constexpr std::size_t N = 4;

sycl/test-e2e/BFloat16/bfloat16_builtins.cpp

@@ -12,6 +12,11 @@
 // RUN:  %if preview-breaking-changes-supported %{  %clangxx -fsycl -fpreview-breaking-changes %{sycl_target_opts} %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() {

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---------------------------==//
 //

sycl/test-e2e/Basic/multisource.cpp

@@ -19,6 +19,9 @@
 // RUN: %clangxx -fsycl %{sycl_target_opts} %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 <sycl/detail/core.hpp>
 
 #include <iostream>

sycl/test-e2e/Basic/multisource_spv_obj.cpp

@@ -28,6 +28,9 @@
 // RUN: %clangxx -fsycl %{sycl_target_opts} %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 <sycl/detail/core.hpp>
 
 #include <iostream>

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

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 <cassert>
 #include <sycl/detail/core.hpp>

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 && opencl && igc-dev
+// XFAIL: arch-intel_gpu_pvc && opencl && igc-dev && !spirv-backend
 // XFAIL-TRACKER: https://github.com/intel/llvm/issues/16401
 
 // RUN: %clangxx -fsycl -fsycl-targets=intel_gpu_pvc %s -o %t.out

sycl/test-e2e/DeviceCodeSplit/grf.cpp

@@ -17,6 +17,10 @@
 // XFAIL: arch-intel_gpu_pvc && opencl
 // 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

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 <iostream>
 #include <sycl/detail/core.hpp>

sycl/test-e2e/DeviceImageDependencies/NewOffloadDriver/free_function_kernels.cpp

@@ -11,6 +11,9 @@
 // XFAIL: hip
 // XFAIL-TRACKER: https://github.com/intel/llvm/issues/15742
 
+// XFAIL: spirv-backend
+// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16319
+
 #include <iostream>
 #include <sycl/detail/core.hpp>
 #include <sycl/ext/oneapi/free_function_queries.hpp>

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 <cmath>
 #include <sycl/detail/core.hpp>
 

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 <iostream>
 #include <sycl/detail/core.hpp>

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()); }

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"
 

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 <vector>
 

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 <cmath>
 #include <sycl/ext/intel/math.hpp>

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 <sycl/ext/intel/math.hpp>
 

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 <sycl/ext/intel/math.hpp>
 

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 <sycl/detail/core.hpp>
 #include <sycl/ext/intel/math.hpp>

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 support this feature.
+config.unsupported_features += ['spirv-backend']

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}

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<float> is sensitive to .get_size() vs .size() in SYCL headers

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.

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 support this feature.
+config.unsupported_features += ['spirv-backend']

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]*}}

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.

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 <sycl/detail/core.hpp>

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
 

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
 

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
 

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']

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 <sycl/detail/core.hpp>
 
 __attribute__((noinline)) long long foo(int data1, long long data2) {

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 <sycl/detail/core.hpp>
 #include <sycl/usm.hpp>
 

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 <sycl/detail/core.hpp>
 #include <sycl/usm.hpp>
 

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 <sycl/vector.hpp>
 

sycl/test-e2e/NewOffloadDriver/multisource.cpp

@@ -39,6 +39,9 @@
 // RUN: %clangxx -Wno-error=unused-command-line-argument -fsycl %{sycl_target_opts} --offload-new-driver %t.main.o %t.a -o %t4.fat
 // RUN: %{run} %t4.fat
 
+// XFAIL: spirv-backend
+// XFAIL-TRACKER: CMPLRLLVM-64059
+
 #include <sycl/detail/core.hpp>
 
 #include <iostream>