diff --git a/SYCL/Assert/assert_in_kernels.cpp b/SYCL/Assert/assert_in_kernels.cpp index df9574badc..db0b14edda 100644 --- a/SYCL/Assert/assert_in_kernels.cpp +++ b/SYCL/Assert/assert_in_kernels.cpp @@ -1,6 +1,7 @@ // REQUIRES: linux // FIXME unsupported on CUDA and HIP until fallback libdevice becomes available // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -DSYCL_ENABLE_FALLBACK_ASSERT -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out &> %t.txt || true // RUN: %CPU_RUN_PLACEHOLDER FileCheck %s --input-file %t.txt diff --git a/SYCL/Assert/assert_in_multiple_tus.cpp b/SYCL/Assert/assert_in_multiple_tus.cpp index 6b271b12b8..a62303bb51 100644 --- a/SYCL/Assert/assert_in_multiple_tus.cpp +++ b/SYCL/Assert/assert_in_multiple_tus.cpp @@ -1,6 +1,7 @@ // REQUIRES: linux // FIXME unsupported on CUDA and HIP until fallback libdevice becomes available // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -DSYCL_ENABLE_FALLBACK_ASSERT -fsycl -fsycl-targets=%sycl_triple -I %S/Inputs %s %S/Inputs/kernels_in_file2.cpp -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out &> %t.txt || true // RUN: %CPU_RUN_PLACEHOLDER FileCheck %s --input-file %t.txt diff --git a/SYCL/Assert/assert_in_multiple_tus_one_ndebug.cpp b/SYCL/Assert/assert_in_multiple_tus_one_ndebug.cpp index dcbb1eaf9c..b124190162 100644 --- a/SYCL/Assert/assert_in_multiple_tus_one_ndebug.cpp +++ b/SYCL/Assert/assert_in_multiple_tus_one_ndebug.cpp @@ -1,6 +1,7 @@ // REQUIRES: linux // FIXME unsupported on CUDA and HIP until fallback libdevice becomes available // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -DSYCL_ENABLE_FALLBACK_ASSERT -fsycl -fsycl-targets=%sycl_triple -DDEFINE_NDEBUG_INFILE2 -I %S/Inputs %S/assert_in_multiple_tus.cpp %S/Inputs/kernels_in_file2.cpp -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out &> %t.txt || true // RUN: %CPU_RUN_PLACEHOLDER FileCheck %s --input-file %t.txt diff --git a/SYCL/Assert/assert_in_one_kernel.cpp b/SYCL/Assert/assert_in_one_kernel.cpp index 5625cad660..df41b14740 100644 --- a/SYCL/Assert/assert_in_one_kernel.cpp +++ b/SYCL/Assert/assert_in_one_kernel.cpp @@ -1,6 +1,7 @@ // REQUIRES: linux // FIXME unsupported on CUDA and HIP until fallback libdevice becomes available // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -DSYCL_ENABLE_FALLBACK_ASSERT -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out &> %t.txt || true // RUN: %CPU_RUN_PLACEHOLDER FileCheck %s --input-file %t.txt diff --git a/SYCL/Assert/assert_in_simultaneous_kernels.cpp b/SYCL/Assert/assert_in_simultaneous_kernels.cpp index 54f9f4d523..7081ed960e 100644 --- a/SYCL/Assert/assert_in_simultaneous_kernels.cpp +++ b/SYCL/Assert/assert_in_simultaneous_kernels.cpp @@ -1,6 +1,7 @@ // REQUIRES: linux // FIXME unsupported on CUDA and HIP until fallback libdevice becomes available // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -DSYCL_ENABLE_FALLBACK_ASSERT -fsycl -fsycl-targets=%sycl_triple %s -o %t.out %threads_lib // RUN: %CPU_RUN_PLACEHOLDER %t.out &> %t.txt || true // RUN: %CPU_RUN_PLACEHOLDER FileCheck %s --input-file %t.txt diff --git a/SYCL/AtomicRef/accessor.cpp b/SYCL/AtomicRef/accessor.cpp index 05c450b4b1..caea583cd1 100644 --- a/SYCL/AtomicRef/accessor.cpp +++ b/SYCL/AtomicRef/accessor.cpp @@ -2,6 +2,7 @@ // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/AtomicRef/add.cpp b/SYCL/AtomicRef/add.cpp index 1b3ab70085..c748364851 100644 --- a/SYCL/AtomicRef/add.cpp +++ b/SYCL/AtomicRef/add.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include "add.h" #include diff --git a/SYCL/AtomicRef/assignment.cpp b/SYCL/AtomicRef/assignment.cpp index 0d852b583d..dab310ad16 100644 --- a/SYCL/AtomicRef/assignment.cpp +++ b/SYCL/AtomicRef/assignment.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include "assignment.h" #include diff --git a/SYCL/AtomicRef/atomic_memory_order.cpp b/SYCL/AtomicRef/atomic_memory_order.cpp index 517ce985dc..6af2f3b14c 100644 --- a/SYCL/AtomicRef/atomic_memory_order.cpp +++ b/SYCL/AtomicRef/atomic_memory_order.cpp @@ -6,6 +6,7 @@ // L0, OpenCL, and HIP backends don't currently support // info::device::atomic_memory_order_capabilities // UNSUPPORTED: level_zero || opencl || hip +// UNSUPPORTED: esimd_emulator // NOTE: General tests for atomic memory order capabilities. diff --git a/SYCL/AtomicRef/atomic_memory_order_seq_cst.cpp b/SYCL/AtomicRef/atomic_memory_order_seq_cst.cpp index b26afb2461..d740413789 100644 --- a/SYCL/AtomicRef/atomic_memory_order_seq_cst.cpp +++ b/SYCL/AtomicRef/atomic_memory_order_seq_cst.cpp @@ -6,6 +6,7 @@ // L0, OpenCL, and HIP backends don't currently support // info::device::atomic_memory_order_capabilities // UNSUPPORTED: level_zero || opencl || hip +// UNSUPPORTED: esimd_emulator // NOTE: Tests load and store for sequentially consistent memory ordering. diff --git a/SYCL/AtomicRef/compare_exchange.cpp b/SYCL/AtomicRef/compare_exchange.cpp index 278f74e8c8..a68c9b44bb 100644 --- a/SYCL/AtomicRef/compare_exchange.cpp +++ b/SYCL/AtomicRef/compare_exchange.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include "compare_exchange.h" #include diff --git a/SYCL/AtomicRef/exchange.cpp b/SYCL/AtomicRef/exchange.cpp index 0c250bbf11..dbec0fd3c5 100644 --- a/SYCL/AtomicRef/exchange.cpp +++ b/SYCL/AtomicRef/exchange.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include "exchange.h" #include diff --git a/SYCL/AtomicRef/load.cpp b/SYCL/AtomicRef/load.cpp index 2bbb814ae1..5b5cfb279a 100644 --- a/SYCL/AtomicRef/load.cpp +++ b/SYCL/AtomicRef/load.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include "load.h" #include diff --git a/SYCL/AtomicRef/max.cpp b/SYCL/AtomicRef/max.cpp index 0e3517f922..cd525a3549 100644 --- a/SYCL/AtomicRef/max.cpp +++ b/SYCL/AtomicRef/max.cpp @@ -3,6 +3,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include "max.h" #include diff --git a/SYCL/AtomicRef/min.cpp b/SYCL/AtomicRef/min.cpp index d484911d96..251657f398 100644 --- a/SYCL/AtomicRef/min.cpp +++ b/SYCL/AtomicRef/min.cpp @@ -3,6 +3,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include "min.h" #include diff --git a/SYCL/AtomicRef/store.cpp b/SYCL/AtomicRef/store.cpp index 493dd23f2c..126be90c31 100644 --- a/SYCL/AtomicRef/store.cpp +++ b/SYCL/AtomicRef/store.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include "store.h" #include diff --git a/SYCL/AtomicRef/sub.cpp b/SYCL/AtomicRef/sub.cpp index a3ff1dd2cc..e6f10fe1c5 100644 --- a/SYCL/AtomicRef/sub.cpp +++ b/SYCL/AtomicRef/sub.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include "sub.h" #include diff --git a/SYCL/Basic/access_to_subset.cpp b/SYCL/Basic/access_to_subset.cpp index b0a80f8567..a9ee040f19 100644 --- a/SYCL/Basic/access_to_subset.cpp +++ b/SYCL/Basic/access_to_subset.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==---------- access_to_subset.cpp --- access to subset of buffer test ----==// // diff --git a/SYCL/Basic/accessor/accessor.cpp b/SYCL/Basic/accessor/accessor.cpp index b30cbcbe2e..991af750cf 100644 --- a/SYCL/Basic/accessor/accessor.cpp +++ b/SYCL/Basic/accessor/accessor.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==----------------accessor.cpp - SYCL accessor basic test ----------------==// // diff --git a/SYCL/Basic/accessor/device_accessor_deduction.cpp b/SYCL/Basic/accessor/device_accessor_deduction.cpp index c63102cb00..ba9d2ae197 100644 --- a/SYCL/Basic/accessor/device_accessor_deduction.cpp +++ b/SYCL/Basic/accessor/device_accessor_deduction.cpp @@ -3,3 +3,4 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator diff --git a/SYCL/Basic/accessor/get_device_access_deduction.cpp b/SYCL/Basic/accessor/get_device_access_deduction.cpp index bf80d9a6c9..61aed92d44 100644 --- a/SYCL/Basic/accessor/get_device_access_deduction.cpp +++ b/SYCL/Basic/accessor/get_device_access_deduction.cpp @@ -3,3 +3,4 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator diff --git a/SYCL/Basic/barrier_order.cpp b/SYCL/Basic/barrier_order.cpp index 60217603d8..3272939c78 100644 --- a/SYCL/Basic/barrier_order.cpp +++ b/SYCL/Basic/barrier_order.cpp @@ -2,6 +2,7 @@ // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Basic/bit_cast/bit_cast.cpp b/SYCL/Basic/bit_cast/bit_cast.cpp index fa8245a54c..26b5506d2b 100644 --- a/SYCL/Basic/bit_cast/bit_cast.cpp +++ b/SYCL/Basic/bit_cast/bit_cast.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/boolean.cpp b/SYCL/Basic/boolean.cpp index 22aa2eeb9f..18a17af581 100644 --- a/SYCL/Basic/boolean.cpp +++ b/SYCL/Basic/boolean.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Basic/buffer/buffer.cpp b/SYCL/Basic/buffer/buffer.cpp index 60aab62c02..401f71db9d 100644 --- a/SYCL/Basic/buffer/buffer.cpp +++ b/SYCL/Basic/buffer/buffer.cpp @@ -5,6 +5,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t2.out // RUN: %GPU_RUN_PLACEHOLDER %t2.out // RUN: %ACC_RUN_PLACEHOLDER %t2.out +// UNSUPPORTED: esimd_emulator //==------------------- buffer.cpp - SYCL buffer basic test ----------------==// // diff --git a/SYCL/Basic/buffer/buffer_container.cpp b/SYCL/Basic/buffer/buffer_container.cpp index 585f31c5b9..93dcba6717 100644 --- a/SYCL/Basic/buffer/buffer_container.cpp +++ b/SYCL/Basic/buffer/buffer_container.cpp @@ -5,6 +5,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t2.out // RUN: %GPU_RUN_PLACEHOLDER %t2.out // RUN: %ACC_RUN_PLACEHOLDER %t2.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/buffer/buffer_dev_to_dev.cpp b/SYCL/Basic/buffer/buffer_dev_to_dev.cpp index d44302226b..085fbc9ca5 100644 --- a/SYCL/Basic/buffer/buffer_dev_to_dev.cpp +++ b/SYCL/Basic/buffer/buffer_dev_to_dev.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==---------- buffer_dev_to_dev.cpp - SYCL buffer basic test --------------==// // diff --git a/SYCL/Basic/buffer/buffer_full_copy.cpp b/SYCL/Basic/buffer/buffer_full_copy.cpp index a1d77fa19e..33b39d568f 100644 --- a/SYCL/Basic/buffer/buffer_full_copy.cpp +++ b/SYCL/Basic/buffer/buffer_full_copy.cpp @@ -5,6 +5,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t2.out // RUN: %GPU_RUN_PLACEHOLDER %t2.out // RUN: %ACC_RUN_PLACEHOLDER %t2.out +// UNSUPPORTED: esimd_emulator //==------------- buffer_full_copy.cpp - SYCL buffer basic test ------------==// // diff --git a/SYCL/Basic/buffer/reinterpret.cpp b/SYCL/Basic/buffer/reinterpret.cpp index 3890dc7598..d96c4b13a0 100644 --- a/SYCL/Basic/buffer/reinterpret.cpp +++ b/SYCL/Basic/buffer/reinterpret.cpp @@ -4,6 +4,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // // XFAIL: level_zero&&gpu +// UNSUPPORTED: esimd_emulator //==---------- reinterpret.cpp --- SYCL buffer reinterpret basic test ------==// // diff --git a/SYCL/Basic/buffer/subbuffer.cpp b/SYCL/Basic/buffer/subbuffer.cpp index 90c8a5bf72..28725ba828 100644 --- a/SYCL/Basic/buffer/subbuffer.cpp +++ b/SYCL/Basic/buffer/subbuffer.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // //==---------- subbuffer.cpp --- sub-buffer basic test ---------------------==// // diff --git a/SYCL/Basic/compare_exchange_strong.cpp b/SYCL/Basic/compare_exchange_strong.cpp index 48e9a52097..d14eabf828 100644 --- a/SYCL/Basic/compare_exchange_strong.cpp +++ b/SYCL/Basic/compare_exchange_strong.cpp @@ -2,6 +2,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include using namespace cl::sycl; diff --git a/SYCL/Basic/device_equality.cpp b/SYCL/Basic/device_equality.cpp index 24b27c63db..7a9465ca8b 100644 --- a/SYCL/Basic/device_equality.cpp +++ b/SYCL/Basic/device_equality.cpp @@ -2,6 +2,7 @@ // RUN: %HOST_RUN_PLACEHOLDER %t1.out // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==------- device_equality.cpp - SYCL device equality test ----------------==// // diff --git a/SYCL/Basic/device_event.cpp b/SYCL/Basic/device_event.cpp index 3dee230031..fdc9d46712 100644 --- a/SYCL/Basic/device_event.cpp +++ b/SYCL/Basic/device_event.cpp @@ -8,6 +8,7 @@ // Crashes on AMD, returns error "Barrier is not supported on the host device // yet." with Nvidia. // XFAIL: hip_amd || hip_nvidia +// UNSUPPORTED: esimd_emulator //==--------device_event.cpp - SYCL class device_event test ----------------==// // diff --git a/SYCL/Basic/enqueue_barrier.cpp b/SYCL/Basic/enqueue_barrier.cpp index 1f90286f64..fb3c3b6ade 100644 --- a/SYCL/Basic/enqueue_barrier.cpp +++ b/SYCL/Basic/enqueue_barrier.cpp @@ -6,6 +6,7 @@ // The test is failing sporadically on Windows OpenCL RTs // Disabling on windows until fixed // UNSUPPORTED: windows +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Basic/event.cpp b/SYCL/Basic/event.cpp index e1365377bd..1bae60ea28 100644 --- a/SYCL/Basic/event.cpp +++ b/SYCL/Basic/event.cpp @@ -2,6 +2,7 @@ // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==--------------- event.cpp - SYCL event test ----------------------------==// // diff --git a/SYCL/Basic/event_profiling_info.cpp b/SYCL/Basic/event_profiling_info.cpp index 2a7f7a5a10..23c7949b38 100644 --- a/SYCL/Basic/event_profiling_info.cpp +++ b/SYCL/Basic/event_profiling_info.cpp @@ -4,6 +4,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==------------------- event_profiling_info.cpp ---------------------------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. diff --git a/SYCL/Basic/free_function_queries/free_function_queries.cpp b/SYCL/Basic/free_function_queries/free_function_queries.cpp index 7dc453146b..589a2d4eb1 100644 --- a/SYCL/Basic/free_function_queries/free_function_queries.cpp +++ b/SYCL/Basic/free_function_queries/free_function_queries.cpp @@ -3,6 +3,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==- free_function_queries.cpp - SYCL free function queries test -=// // diff --git a/SYCL/Basic/get_backend.cpp b/SYCL/Basic/get_backend.cpp index 94d8aabeb6..98a0add85f 100644 --- a/SYCL/Basic/get_backend.cpp +++ b/SYCL/Basic/get_backend.cpp @@ -3,6 +3,7 @@ // // Failing on HIP AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator // //==----------------- get_backend.cpp ------------------------==// // This is a test of get_backend(). diff --git a/SYCL/Basic/group_async_copy.cpp b/SYCL/Basic/group_async_copy.cpp index e6f4efab46..18059e95fa 100644 --- a/SYCL/Basic/group_async_copy.cpp +++ b/SYCL/Basic/group_async_copy.cpp @@ -6,6 +6,7 @@ // // Crashes on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Basic/handler/handler_copy_with_offset.cpp b/SYCL/Basic/handler/handler_copy_with_offset.cpp index 6659b70b3f..27f415c280 100644 --- a/SYCL/Basic/handler/handler_copy_with_offset.cpp +++ b/SYCL/Basic/handler/handler_copy_with_offset.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==--- handler_copy_with_offset.cpp - SYCL handler copy with offset test --==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. diff --git a/SYCL/Basic/handler/handler_generic_integral_lambda.cpp b/SYCL/Basic/handler/handler_generic_integral_lambda.cpp index 7c7205e91c..c9b8369e81 100644 --- a/SYCL/Basic/handler/handler_generic_integral_lambda.cpp +++ b/SYCL/Basic/handler/handler_generic_integral_lambda.cpp @@ -2,6 +2,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==-------------- handler_generic_integral_lambda.cpp ---------------------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. diff --git a/SYCL/Basic/handler/handler_mem_op.cpp b/SYCL/Basic/handler/handler_mem_op.cpp index 7fc5439a0d..44d22d4ec2 100644 --- a/SYCL/Basic/handler/handler_mem_op.cpp +++ b/SYCL/Basic/handler/handler_mem_op.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda +// UNSUPPORTED: esimd_emulator //==- handler.cpp - SYCL handler explicit memory operations test -*- C++-*--==// // diff --git a/SYCL/Basic/host-task-dependency.cpp b/SYCL/Basic/host-task-dependency.cpp index 2484e80f5a..594c8e922b 100644 --- a/SYCL/Basic/host-task-dependency.cpp +++ b/SYCL/Basic/host-task-dependency.cpp @@ -6,6 +6,7 @@ // TODO: Behaviour is unstable for level zero on Windows. Enable when fixed. // TODO: The test is sporadically fails on CUDA. Enable when fixed. // UNSUPPORTED: (windows && level_zero) || cuda || hip_nvidia +// UNSUPPORTED: esimd_emulator #define SYCL2020_DISABLE_DEPRECATION_WARNINGS diff --git a/SYCL/Basic/image/image.cpp b/SYCL/Basic/image/image.cpp index e083b53949..87c44a2d21 100644 --- a/SYCL/Basic/image/image.cpp +++ b/SYCL/Basic/image/image.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Basic/image/image_accessor_range.cpp b/SYCL/Basic/image/image_accessor_range.cpp index 1b0a61b3bf..7074367033 100755 --- a/SYCL/Basic/image/image_accessor_range.cpp +++ b/SYCL/Basic/image/image_accessor_range.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // CUDA does not support SYCL 1.2.1 images. // // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out diff --git a/SYCL/Basic/image/image_accessor_readsampler.cpp b/SYCL/Basic/image/image_accessor_readsampler.cpp index 379c1552dc..841f0e90f7 100644 --- a/SYCL/Basic/image/image_accessor_readsampler.cpp +++ b/SYCL/Basic/image/image_accessor_readsampler.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // CUDA cannot support SYCL 1.2.1 images. // // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out diff --git a/SYCL/Basic/image/image_accessor_readwrite.cpp b/SYCL/Basic/image/image_accessor_readwrite.cpp index e656275f7a..ec3c72a138 100644 --- a/SYCL/Basic/image/image_accessor_readwrite.cpp +++ b/SYCL/Basic/image/image_accessor_readwrite.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // CUDA cannot support SYCL 1.2.1 images. // // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out diff --git a/SYCL/Basic/image/image_max_size.cpp b/SYCL/Basic/image/image_max_size.cpp index 9535f82477..0dcf7b4cc7 100644 --- a/SYCL/Basic/image/image_max_size.cpp +++ b/SYCL/Basic/image/image_max_size.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip || (windows && opencl && gpu) +// UNSUPPORTED: esimd_emulator // CUDA does not support info::device::image3d_max_width query. // TODO: Irregular runtime fails on Windows/opencl:gpu require analysis. diff --git a/SYCL/Basic/image/image_read.cpp b/SYCL/Basic/image/image_read.cpp index f76b891297..5d57277b86 100644 --- a/SYCL/Basic/image/image_read.cpp +++ b/SYCL/Basic/image/image_read.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Basic/image/image_sample.cpp b/SYCL/Basic/image/image_sample.cpp index e4bc20b184..d3e834f683 100644 --- a/SYCL/Basic/image/image_sample.cpp +++ b/SYCL/Basic/image/image_sample.cpp @@ -2,6 +2,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/image/image_write.cpp b/SYCL/Basic/image/image_write.cpp index 3f4aa66d7f..3cb90be2ca 100644 --- a/SYCL/Basic/image/image_write.cpp +++ b/SYCL/Basic/image/image_write.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // TODO: re-enable on cuda device. // See https://github.com/intel/llvm/issues/1542#issuecomment-707877817 for more // details. diff --git a/SYCL/Basic/image/srgba-read.cpp b/SYCL/Basic/image/srgba-read.cpp index c1b6dbd959..bf53ce3b14 100644 --- a/SYCL/Basic/image/srgba-read.cpp +++ b/SYCL/Basic/image/srgba-read.cpp @@ -5,6 +5,7 @@ // XFAIL: level_zero // UNSUPPORTED: cuda // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/kernel_info.cpp b/SYCL/Basic/kernel_info.cpp index 60ef3539cb..25aa1ce133 100644 --- a/SYCL/Basic/kernel_info.cpp +++ b/SYCL/Basic/kernel_info.cpp @@ -5,6 +5,7 @@ // // Fail is flaky for level_zero, enable when fixed. // UNSUPPORTED: level_zero +// UNSUPPORTED: esimd_emulator //==--- kernel_info.cpp - SYCL kernel info test ----------------------------==// // diff --git a/SYCL/Basic/memory-consumption.cpp b/SYCL/Basic/memory-consumption.cpp index 861096212a..78d28b52b2 100644 --- a/SYCL/Basic/memory-consumption.cpp +++ b/SYCL/Basic/memory-consumption.cpp @@ -6,6 +6,7 @@ // processing OCL_ICD_FILENAMES debug environment variable which causes // extra memory allocation on device creation. // UNSUPPORTED: windows, opencl +// UNSUPPORTED: esimd_emulator // //==-----memory-consumption.cpp - SYCL memory consumption basic test ------==// // diff --git a/SYCL/Basic/multi_ptr.cpp b/SYCL/Basic/multi_ptr.cpp index e771a62763..13cc977ab9 100644 --- a/SYCL/Basic/multi_ptr.cpp +++ b/SYCL/Basic/multi_ptr.cpp @@ -8,6 +8,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==--------------- multi_ptr.cpp - SYCL multi_ptr test --------------------==// // diff --git a/SYCL/Basic/multisource.cpp b/SYCL/Basic/multisource.cpp index 11679f8c2e..70c33a083f 100644 --- a/SYCL/Basic/multisource.cpp +++ b/SYCL/Basic/multisource.cpp @@ -24,6 +24,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.fat // RUN: %GPU_RUN_PLACEHOLDER %t.fat // RUN: %ACC_RUN_PLACEHOLDER %t.fat +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/parallel_for_disable_range_roundup.cpp b/SYCL/Basic/parallel_for_disable_range_roundup.cpp index cc27950a31..74a3b50795 100755 --- a/SYCL/Basic/parallel_for_disable_range_roundup.cpp +++ b/SYCL/Basic/parallel_for_disable_range_roundup.cpp @@ -7,6 +7,7 @@ // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -sycl-std=2020 %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER SYCL_PARALLEL_FOR_RANGE_ROUNDING_TRACE=1 %t.out %GPU_CHECK_PLACEHOLDER --check-prefix CHECK-ENABLED +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/parallel_for_indexers.cpp b/SYCL/Basic/parallel_for_indexers.cpp index 578090a0d3..68b3cd391f 100644 --- a/SYCL/Basic/parallel_for_indexers.cpp +++ b/SYCL/Basic/parallel_for_indexers.cpp @@ -8,6 +8,7 @@ // // Incorrect results with hip on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/parallel_for_range_roundup.cpp b/SYCL/Basic/parallel_for_range_roundup.cpp index 7ae3280d42..e23374bbdc 100644 --- a/SYCL/Basic/parallel_for_range_roundup.cpp +++ b/SYCL/Basic/parallel_for_range_roundup.cpp @@ -1,5 +1,6 @@ // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: env SYCL_PARALLEL_FOR_RANGE_ROUNDING_TRACE=1 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/queue/release.cpp b/SYCL/Basic/queue/release.cpp index c690671a6a..be721b4c6b 100644 --- a/SYCL/Basic/queue/release.cpp +++ b/SYCL/Basic/queue/release.cpp @@ -4,6 +4,7 @@ // RUN: env SYCL_PI_TRACE=2 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator #include int main() { diff --git a/SYCL/Basic/reqd_work_group_size.cpp b/SYCL/Basic/reqd_work_group_size.cpp index 5fb9de35d8..5be425b987 100644 --- a/SYCL/Basic/reqd_work_group_size.cpp +++ b/SYCL/Basic/reqd_work_group_size.cpp @@ -5,6 +5,7 @@ // // Failing negative test with HIP // XFAIL: hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/scalar_vec_access.cpp b/SYCL/Basic/scalar_vec_access.cpp index 245c0e9c03..a0be692929 100644 --- a/SYCL/Basic/scalar_vec_access.cpp +++ b/SYCL/Basic/scalar_vec_access.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER // RUN: %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER // RUN: %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER +// UNSUPPORTED: esimd_emulator //==------- scalar_vec_access.cpp - SYCL scalar access to vec test ---------==// // diff --git a/SYCL/Basic/span.cpp b/SYCL/Basic/span.cpp index db45ce35c2..5c56da85aa 100644 --- a/SYCL/Basic/span.cpp +++ b/SYCL/Basic/span.cpp @@ -5,6 +5,7 @@ // // Fails to release USM pointer on HIP for NVIDIA // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Basic/stream/auto_flush.cpp b/SYCL/Basic/stream/auto_flush.cpp index 75a9192040..eeb6a25858 100644 --- a/SYCL/Basic/stream/auto_flush.cpp +++ b/SYCL/Basic/stream/auto_flush.cpp @@ -5,6 +5,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator //==-------------- copy.cpp - SYCL stream obect auto flushing test ---------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. diff --git a/SYCL/Basic/stream/release_resources_test.cpp b/SYCL/Basic/stream/release_resources_test.cpp index 8b7d2ee73d..c1261efc06 100644 --- a/SYCL/Basic/stream/release_resources_test.cpp +++ b/SYCL/Basic/stream/release_resources_test.cpp @@ -2,6 +2,7 @@ // RUN: env SYCL_PI_TRACE=2 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER // RUN: env SYCL_PI_TRACE=2 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER // RUN: env SYCL_PI_TRACE=2 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER +// UNSUPPORTED: esimd_emulator // Check that buffer used by a stream object is released. diff --git a/SYCL/Basic/stream/stream.cpp b/SYCL/Basic/stream/stream.cpp index 30a12c8f20..8ca4f5315a 100644 --- a/SYCL/Basic/stream/stream.cpp +++ b/SYCL/Basic/stream/stream.cpp @@ -7,6 +7,7 @@ // // Missing built-ins on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator //==------------------ stream.cpp - SYCL stream basic test -----------------==// // diff --git a/SYCL/Basic/stream/stream_copies_buffer_sync.cpp b/SYCL/Basic/stream/stream_copies_buffer_sync.cpp index 2275127e1a..81ceacd179 100644 --- a/SYCL/Basic/stream/stream_copies_buffer_sync.cpp +++ b/SYCL/Basic/stream/stream_copies_buffer_sync.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER // RUN: %GPU_RUN_ON_LINUX_PLACEHOLDER %t.out %GPU_CHECK_ON_LINUX_PLACEHOLDER // RUN: %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/submit_barrier.cpp b/SYCL/Basic/submit_barrier.cpp index f7a64f93e9..291b02ead1 100755 --- a/SYCL/Basic/submit_barrier.cpp +++ b/SYCL/Basic/submit_barrier.cpp @@ -2,6 +2,7 @@ // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Basic/swizzle_op.cpp b/SYCL/Basic/swizzle_op.cpp index abfcee0303..120de1c936 100644 --- a/SYCL/Basic/swizzle_op.cpp +++ b/SYCL/Basic/swizzle_op.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==------------ swizzle_op.cpp - SYCL SwizzleOp basic test ----------------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. diff --git a/SYCL/Basic/sycl-namespace.cpp b/SYCL/Basic/sycl-namespace.cpp index 3304bb4464..134211408e 100644 --- a/SYCL/Basic/sycl-namespace.cpp +++ b/SYCL/Basic/sycl-namespace.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Basic/unused_pointer.cpp b/SYCL/Basic/unused_pointer.cpp index 9c8f572b7e..cb1f38a095 100644 --- a/SYCL/Basic/unused_pointer.cpp +++ b/SYCL/Basic/unused_pointer.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==---------- unused_pointer.cpp - test pointers in struct --------------==// // diff --git a/SYCL/Basic/vector_byte.cpp b/SYCL/Basic/vector_byte.cpp index b559487149..ef06c16fa9 100644 --- a/SYCL/Basic/vector_byte.cpp +++ b/SYCL/Basic/vector_byte.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==---------- vector_byte.cpp - SYCL vec<> for std::byte test -------------==// // diff --git a/SYCL/Basic/vector_operators.cpp b/SYCL/Basic/vector_operators.cpp index 4298ed76cf..4ba9e17059 100644 --- a/SYCL/Basic/vector_operators.cpp +++ b/SYCL/Basic/vector_operators.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==---------- vector_operators.cpp - SYCL vec<> operators test ------------==// // diff --git a/SYCL/Config/kernel_from_file.cpp b/SYCL/Config/kernel_from_file.cpp index f1fd61ffc3..0bf2413062 100644 --- a/SYCL/Config/kernel_from_file.cpp +++ b/SYCL/Config/kernel_from_file.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // CUDA and HIP don't support SPIR-V. // FIXME Disabled fallback assert as it'll require either online linking or diff --git a/SYCL/Config/select_device.cpp b/SYCL/Config/select_device.cpp index 9c9dfed2c7..020e7ca8d1 100644 --- a/SYCL/Config/select_device.cpp +++ b/SYCL/Config/select_device.cpp @@ -39,6 +39,7 @@ // REQUIRES: gpu // // XFAIL: cuda || hip +// UNSUPPORTED: esimd_emulator // // TODO: Update this test when SYCL_DEVICE_FILTER support in enabled. diff --git a/SYCL/DeprecatedFeatures/ESIMD/histogram_256_slm_spec.cpp b/SYCL/DeprecatedFeatures/ESIMD/histogram_256_slm_spec.cpp index dfa77e1a17..29b601c294 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/histogram_256_slm_spec.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/histogram_256_slm_spec.cpp @@ -7,6 +7,7 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl %s -I%S/Inputs -D__SYCL_INTERNAL_API -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out 16 diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_bool.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_bool.cpp index c7cd3f4ae3..26cee3f6c9 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_bool.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_bool.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_char.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_char.cpp index 18d0f3ea43..c979a97ded 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_char.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_char.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_double.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_double.cpp index 4a8715c582..c2cb014189 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_double.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_double.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_float.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_float.cpp index f270c58580..243f521c79 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_float.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_float.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_int.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_int.cpp index f9185f5c76..a425b38856 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_int.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_int.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_int64.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_int64.cpp index 2e4c6294ad..46e3118810 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_int64.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_int64.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_redefine_esimd.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_redefine_esimd.cpp index 178edec960..d1f242249f 100755 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_redefine_esimd.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_redefine_esimd.cpp @@ -4,6 +4,7 @@ // RUN: %clangxx -DSYCL_DISABLE_FALLBACK_ASSERT -D__SYCL_INTERNAL_API -fsycl -I%S/Inputs %s -o %t.out // RUN: env SYCL_PI_TRACE=2 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator //==----------- spec_const_redefine_esimd.cpp ------------------------------==// // diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_short.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_short.cpp index f07697a874..d084c631b8 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_short.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_short.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_uchar.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_uchar.cpp index 905ee0ba64..920901cfe1 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_uchar.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_uchar.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_uint.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_uint.cpp index 2f1b179e24..80304f56f6 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_uint.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_uint.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_uint64.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_uint64.cpp index 4216079e86..24654cc1a2 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_uint64.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_uint64.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/ESIMD/spec_const_ushort.cpp b/SYCL/DeprecatedFeatures/ESIMD/spec_const_ushort.cpp index d9d72e94dc..eb29338783 100644 --- a/SYCL/DeprecatedFeatures/ESIMD/spec_const_ushort.cpp +++ b/SYCL/DeprecatedFeatures/ESIMD/spec_const_ushort.cpp @@ -9,6 +9,7 @@ // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-in-functor.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-in-functor.cpp index ce8fca7ee7..d47de24a86 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-in-functor.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-in-functor.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // // RUN: %clangxx -fsycl %s -D__SYCL_INTERNAL_API -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out %HOST_CHECK_PLACEHOLDER diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-type.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-type.cpp index cae3a46de0..f841adcc01 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-type.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/composite-type.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/multiple-usages-of-composite.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/multiple-usages-of-composite.cpp index 460b1f9407..a3db29cea4 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/multiple-usages-of-composite.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/multiple-usages-of-composite.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out -v // RUN: %HOST_RUN_PLACEHOLDER %t.out %HOST_CHECK_PLACEHOLDER diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_hw.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_hw.cpp index 3245c3aec3..a4fd1450e0 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_hw.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_hw.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -D__SYCL_INTERNAL_API %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_neg.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_neg.cpp index f2930c58c4..13a1710e89 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_neg.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_neg.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // Specialization constants are not supported on FPGA h/w and emulator. // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // //==----------- spec_const_hw.cpp ------------------------------------------==// // diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_redefine.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_redefine.cpp index 57be8ae484..ded1591057 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_redefine.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/spec_const_redefine.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // // FIXME Disable fallback assert so that it doesn't interferes with number of // program builds at run-time diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants.cpp index ea52b2fa79..5cd926e47b 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // Specialization constants are not supported on FPGA h/w and emulator. // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // //==----------- specialization_constants.cpp -------------------------------==// // diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_negative.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_negative.cpp index 963ec6362e..b2cb2e496e 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_negative.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_negative.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // Specialization constants are not supported on FPGA h/w and emulator. // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // //==----------- specialization_constants_negative.cpp ----------------------==// // diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_override.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_override.cpp index 4e538d55c5..6d27bb5b9a 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_override.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/specialization_constants_override.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // Specialization constants are not supported on FPGA h/w and emulator. // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // //==----------- specialization_constants_override.cpp ----------------------==// // diff --git a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/unpacked-composite-type.cpp b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/unpacked-composite-type.cpp index 122b88c7c4..84bf680cb9 100644 --- a/SYCL/DeprecatedFeatures/SpecConsts1.2.1/unpacked-composite-type.cpp +++ b/SYCL/DeprecatedFeatures/SpecConsts1.2.1/unpacked-composite-type.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // // RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out %HOST_CHECK_PLACEHOLDER diff --git a/SYCL/DeprecatedFeatures/basic-program.cpp b/SYCL/DeprecatedFeatures/basic-program.cpp index 37c46ad477..dee17f3a96 100644 --- a/SYCL/DeprecatedFeatures/basic-program.cpp +++ b/SYCL/DeprecatedFeatures/basic-program.cpp @@ -1,4 +1,5 @@ // XFAIL: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/DeprecatedFeatures/get-options.cpp b/SYCL/DeprecatedFeatures/get-options.cpp index 134f087147..170fcc007c 100644 --- a/SYCL/DeprecatedFeatures/get-options.cpp +++ b/SYCL/DeprecatedFeatures/get-options.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out // XFAIL: cuda || hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeprecatedFeatures/get_backend.cpp b/SYCL/DeprecatedFeatures/get_backend.cpp index acaf8ad196..9a29589c45 100644 --- a/SYCL/DeprecatedFeatures/get_backend.cpp +++ b/SYCL/DeprecatedFeatures/get_backend.cpp @@ -1,5 +1,6 @@ // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -D__SYCL_INTERNAL_API %s -o %t.out // RUN: env SYCL_DEVICE_FILTER=%sycl_be %t.out +// UNSUPPORTED: esimd_emulator // //==----------------- get_backend.cpp ------------------------==// // This is a test of get_backend(). diff --git a/SYCL/DeprecatedFeatures/kernel-and-program.cpp b/SYCL/DeprecatedFeatures/kernel-and-program.cpp index 4384772f3e..b0917de8a2 100644 --- a/SYCL/DeprecatedFeatures/kernel-and-program.cpp +++ b/SYCL/DeprecatedFeatures/kernel-and-program.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out // XFAIL: hip +// UNSUPPORTED: esimd_emulator //==--- kernel-and-program.cpp - SYCL kernel/program test ------------------==// // diff --git a/SYCL/DeprecatedFeatures/kernel_info.cpp b/SYCL/DeprecatedFeatures/kernel_info.cpp index 428ece522b..525e41b158 100644 --- a/SYCL/DeprecatedFeatures/kernel_info.cpp +++ b/SYCL/DeprecatedFeatures/kernel_info.cpp @@ -5,6 +5,7 @@ // // Fail is flaky for level_zero, enable when fixed. // UNSUPPORTED: level_zero +// UNSUPPORTED: esimd_emulator // // XFAIL: hip_nvidia diff --git a/SYCL/DeprecatedFeatures/parallel_for_range.cpp b/SYCL/DeprecatedFeatures/parallel_for_range.cpp index 056563bbde..f9c24d73ae 100644 --- a/SYCL/DeprecatedFeatures/parallel_for_range.cpp +++ b/SYCL/DeprecatedFeatures/parallel_for_range.cpp @@ -4,6 +4,7 @@ // // Failing on HIP AMD and HIP NVIDIA // UNSUPPORTED: hip_amd || hip_nvidia +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -D__SYCL_INTERNAL_API -fsycl -fno-sycl-id-queries-fit-in-int -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/DeprecatedFeatures/program-merge-options-env.cpp b/SYCL/DeprecatedFeatures/program-merge-options-env.cpp index 80c1d3bd5e..7ea1bfd917 100644 --- a/SYCL/DeprecatedFeatures/program-merge-options-env.cpp +++ b/SYCL/DeprecatedFeatures/program-merge-options-env.cpp @@ -2,6 +2,7 @@ // RUN: env SYCL_PI_TRACE=-1 SYCL_PROGRAM_COMPILE_OPTIONS=-DENV_COMPILE_OPTS SYCL_PROGRAM_LINK_OPTIONS=-DENV_LINK_OPTS SYCL_DEVICE_FILTER=%sycl_be %t.out | FileCheck %s // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include "program-merge-options.hpp" // CHECK: piProgramBuild diff --git a/SYCL/DeprecatedFeatures/program-merge-options.cpp b/SYCL/DeprecatedFeatures/program-merge-options.cpp index 6436e2f693..44d2a89136 100644 --- a/SYCL/DeprecatedFeatures/program-merge-options.cpp +++ b/SYCL/DeprecatedFeatures/program-merge-options.cpp @@ -2,6 +2,7 @@ // RUN: env SYCL_PI_TRACE=-1 SYCL_DEVICE_FILTER=%sycl_be %t.out | FileCheck %s // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // Debug option -g is not passed to device code compiler when CL-style driver // is used and /DEBUG options is passed. diff --git a/SYCL/DeprecatedFeatures/program_link.cpp b/SYCL/DeprecatedFeatures/program_link.cpp index dd79fa9dfd..177c79640e 100644 --- a/SYCL/DeprecatedFeatures/program_link.cpp +++ b/SYCL/DeprecatedFeatures/program_link.cpp @@ -9,6 +9,7 @@ // // Hits an assertion on AMD with multiple GPUs available, fails trace on Nvidia. // XFAIL: hip_amd || hip_nvidia +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/DeviceCodeSplit/split-per-kernel.cpp b/SYCL/DeviceCodeSplit/split-per-kernel.cpp index ea602ba758..eb06c7e193 100644 --- a/SYCL/DeviceCodeSplit/split-per-kernel.cpp +++ b/SYCL/DeviceCodeSplit/split-per-kernel.cpp @@ -4,6 +4,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceCodeSplit/split-per-source-main.cpp b/SYCL/DeviceCodeSplit/split-per-source-main.cpp index 7e93958441..cf57f850e0 100644 --- a/SYCL/DeviceCodeSplit/split-per-source-main.cpp +++ b/SYCL/DeviceCodeSplit/split-per-source-main.cpp @@ -4,6 +4,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator #include "Inputs/split-per-source.h" diff --git a/SYCL/DeviceLib/ITTAnnotations/atomic.cpp b/SYCL/DeviceLib/ITTAnnotations/atomic.cpp index 290c710e73..8d05d57ae8 100644 --- a/SYCL/DeviceLib/ITTAnnotations/atomic.cpp +++ b/SYCL/DeviceLib/ITTAnnotations/atomic.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-instrument-device-code %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out diff --git a/SYCL/DeviceLib/ITTAnnotations/barrier.cpp b/SYCL/DeviceLib/ITTAnnotations/barrier.cpp index a671224e01..fe9a405e68 100644 --- a/SYCL/DeviceLib/ITTAnnotations/barrier.cpp +++ b/SYCL/DeviceLib/ITTAnnotations/barrier.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-instrument-device-code %s -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/DeviceLib/built-ins/nan.cpp b/SYCL/DeviceLib/built-ins/nan.cpp index eaa0246bf8..36b105edb8 100644 --- a/SYCL/DeviceLib/built-ins/nan.cpp +++ b/SYCL/DeviceLib/built-ins/nan.cpp @@ -6,6 +6,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // XFAIL: cuda +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/printf.cpp b/SYCL/DeviceLib/built-ins/printf.cpp index 5b6d7a2cf3..2e9a44972a 100644 --- a/SYCL/DeviceLib/built-ins/printf.cpp +++ b/SYCL/DeviceLib/built-ins/printf.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // CUDA and HIP don't support printf. // // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out diff --git a/SYCL/DeviceLib/built-ins/scalar_common.cpp b/SYCL/DeviceLib/built-ins/scalar_common.cpp index 8dedbaaad9..f7d6784187 100644 --- a/SYCL/DeviceLib/built-ins/scalar_common.cpp +++ b/SYCL/DeviceLib/built-ins/scalar_common.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/scalar_geometric.cpp b/SYCL/DeviceLib/built-ins/scalar_geometric.cpp index 56c5a4ee3c..a9fc1579f3 100644 --- a/SYCL/DeviceLib/built-ins/scalar_geometric.cpp +++ b/SYCL/DeviceLib/built-ins/scalar_geometric.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/scalar_integer.cpp b/SYCL/DeviceLib/built-ins/scalar_integer.cpp index eb5e862f92..26ff6b5df8 100644 --- a/SYCL/DeviceLib/built-ins/scalar_integer.cpp +++ b/SYCL/DeviceLib/built-ins/scalar_integer.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/scalar_math.cpp b/SYCL/DeviceLib/built-ins/scalar_math.cpp index 1ce9fd09a2..5314768bbf 100644 --- a/SYCL/DeviceLib/built-ins/scalar_math.cpp +++ b/SYCL/DeviceLib/built-ins/scalar_math.cpp @@ -6,6 +6,7 @@ // // Incorrect results with hip on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/scalar_math_2.cpp b/SYCL/DeviceLib/built-ins/scalar_math_2.cpp index 266bf8c258..fbf1783bc7 100644 --- a/SYCL/DeviceLib/built-ins/scalar_math_2.cpp +++ b/SYCL/DeviceLib/built-ins/scalar_math_2.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/scalar_relational.cpp b/SYCL/DeviceLib/built-ins/scalar_relational.cpp index 1e459accc8..9666a34d29 100644 --- a/SYCL/DeviceLib/built-ins/scalar_relational.cpp +++ b/SYCL/DeviceLib/built-ins/scalar_relational.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/vector_common.cpp b/SYCL/DeviceLib/built-ins/vector_common.cpp index 623a05ef5e..14e6d8550f 100644 --- a/SYCL/DeviceLib/built-ins/vector_common.cpp +++ b/SYCL/DeviceLib/built-ins/vector_common.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/vector_geometric.cpp b/SYCL/DeviceLib/built-ins/vector_geometric.cpp index 0acb1bd195..7470f427df 100644 --- a/SYCL/DeviceLib/built-ins/vector_geometric.cpp +++ b/SYCL/DeviceLib/built-ins/vector_geometric.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/vector_integer.cpp b/SYCL/DeviceLib/built-ins/vector_integer.cpp index e3505121cd..0e48b9558b 100644 --- a/SYCL/DeviceLib/built-ins/vector_integer.cpp +++ b/SYCL/DeviceLib/built-ins/vector_integer.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/vector_math.cpp b/SYCL/DeviceLib/built-ins/vector_math.cpp index 2056e16688..86adcd83c7 100644 --- a/SYCL/DeviceLib/built-ins/vector_math.cpp +++ b/SYCL/DeviceLib/built-ins/vector_math.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DeviceLib/built-ins/vector_relational.cpp b/SYCL/DeviceLib/built-ins/vector_relational.cpp index 93e28e739f..fd85224544 100644 --- a/SYCL/DeviceLib/built-ins/vector_relational.cpp +++ b/SYCL/DeviceLib/built-ins/vector_relational.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/DotProduct/dot_product_int_test.cpp b/SYCL/DotProduct/dot_product_int_test.cpp index 6f51200c2d..47051c9dda 100644 --- a/SYCL/DotProduct/dot_product_int_test.cpp +++ b/SYCL/DotProduct/dot_product_int_test.cpp @@ -6,6 +6,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/DotProduct/dot_product_vec_test.cpp b/SYCL/DotProduct/dot_product_vec_test.cpp index ce9c87bab8..e27c950624 100644 --- a/SYCL/DotProduct/dot_product_vec_test.cpp +++ b/SYCL/DotProduct/dot_product_vec_test.cpp @@ -6,6 +6,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/ESIMD/BitonicSortK.cpp b/SYCL/ESIMD/BitonicSortK.cpp index 0bbfe53416..0c8f4593c3 100644 --- a/SYCL/ESIMD/BitonicSortK.cpp +++ b/SYCL/ESIMD/BitonicSortK.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/accessor.cpp b/SYCL/ESIMD/accessor.cpp index 0ee005b834..a4e1db842f 100644 --- a/SYCL/ESIMD/accessor.cpp +++ b/SYCL/ESIMD/accessor.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl -D_CRT_SECURE_NO_WARNINGS=1 %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/accessor_gather_scatter.cpp b/SYCL/ESIMD/accessor_gather_scatter.cpp index 8bd66b41e7..4c1cf1c8f4 100644 --- a/SYCL/ESIMD/accessor_gather_scatter.cpp +++ b/SYCL/ESIMD/accessor_gather_scatter.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '()' operator for 'Kernel' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/accessor_load_store.cpp b/SYCL/ESIMD/accessor_load_store.cpp index 4fba632496..1eb831cd76 100644 --- a/SYCL/ESIMD/accessor_load_store.cpp +++ b/SYCL/ESIMD/accessor_load_store.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '()' operator for 'Kernel' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/ballot.cpp b/SYCL/ESIMD/api/ballot.cpp index 470fa36880..a6c039e109 100755 --- a/SYCL/ESIMD/api/ballot.cpp +++ b/SYCL/ESIMD/api/ballot.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '()' operator for 'test' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/api/esimd_bit_ops.cpp b/SYCL/ESIMD/api/esimd_bit_ops.cpp index 4d36e7bb33..235c83e417 100644 --- a/SYCL/ESIMD/api/esimd_bit_ops.cpp +++ b/SYCL/ESIMD/api/esimd_bit_ops.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable 'single_task()' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -fsycl-device-code-split=per_kernel -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/simd_binop_integer_promotion.cpp b/SYCL/ESIMD/api/simd_binop_integer_promotion.cpp index f94917d0af..6c66877e03 100644 --- a/SYCL/ESIMD/api/simd_binop_integer_promotion.cpp +++ b/SYCL/ESIMD/api/simd_binop_integer_promotion.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '()' operator for 'test' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/simd_copy_to_from.cpp b/SYCL/ESIMD/api/simd_copy_to_from.cpp index 92e3e7a631..39eff524fb 100644 --- a/SYCL/ESIMD/api/simd_copy_to_from.cpp +++ b/SYCL/ESIMD/api/simd_copy_to_from.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '()' operator for 'testAcc' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/api/simd_memory_access.cpp b/SYCL/ESIMD/api/simd_memory_access.cpp index c715aa91da..67ac027e7e 100644 --- a/SYCL/ESIMD/api/simd_memory_access.cpp +++ b/SYCL/ESIMD/api/simd_memory_access.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '()' operator for 'Kernel' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/simd_negation_operator.cpp b/SYCL/ESIMD/api/simd_negation_operator.cpp index 72c1438b12..630765cae9 100644 --- a/SYCL/ESIMD/api/simd_negation_operator.cpp +++ b/SYCL/ESIMD/api/simd_negation_operator.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/simd_subscript_operator.cpp b/SYCL/ESIMD/api/simd_subscript_operator.cpp index b9c14cf1b3..1a512932e9 100644 --- a/SYCL/ESIMD/api/simd_subscript_operator.cpp +++ b/SYCL/ESIMD/api/simd_subscript_operator.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/simd_view_copy_move_assign.cpp b/SYCL/ESIMD/api/simd_view_copy_move_assign.cpp index 34b5a31fcd..ffd1f15eee 100644 --- a/SYCL/ESIMD/api/simd_view_copy_move_assign.cpp +++ b/SYCL/ESIMD/api/simd_view_copy_move_assign.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/simd_view_negation_operator.cpp b/SYCL/ESIMD/api/simd_view_negation_operator.cpp index a8405192e1..bb9376021a 100644 --- a/SYCL/ESIMD/api/simd_view_negation_operator.cpp +++ b/SYCL/ESIMD/api/simd_view_negation_operator.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/simd_view_subscript_operator.cpp b/SYCL/ESIMD/api/simd_view_subscript_operator.cpp index d26859a99c..d4376ada69 100644 --- a/SYCL/ESIMD/api/simd_view_subscript_operator.cpp +++ b/SYCL/ESIMD/api/simd_view_subscript_operator.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/slm_gather_scatter.cpp b/SYCL/ESIMD/api/slm_gather_scatter.cpp index aa480d420e..63619c192c 100644 --- a/SYCL/ESIMD/api/slm_gather_scatter.cpp +++ b/SYCL/ESIMD/api/slm_gather_scatter.cpp @@ -1,5 +1,7 @@ // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_scatter_scaled +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/slm_gather_scatter_heavy.cpp b/SYCL/ESIMD/api/slm_gather_scatter_heavy.cpp index 8c19da0905..d669825968 100644 --- a/SYCL/ESIMD/api/slm_gather_scatter_heavy.cpp +++ b/SYCL/ESIMD/api/slm_gather_scatter_heavy.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '()' operator for 'GatherKernel' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/api/slm_gather_scatter_rgba.cpp b/SYCL/ESIMD/api/slm_gather_scatter_rgba.cpp index 6ec3de3de9..f82946136b 100644 --- a/SYCL/ESIMD/api/slm_gather_scatter_rgba.cpp +++ b/SYCL/ESIMD/api/slm_gather_scatter_rgba.cpp @@ -1,5 +1,7 @@ // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_scatter_scaled +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // diff --git a/SYCL/ESIMD/esimd_check_vc_codegen.cpp b/SYCL/ESIMD/esimd_check_vc_codegen.cpp new file mode 100644 index 0000000000..1c677f3366 --- /dev/null +++ b/SYCL/ESIMD/esimd_check_vc_codegen.cpp @@ -0,0 +1,104 @@ +//==-------- esimd_check_vc_codegen.cpp - DPC++ ESIMD on-device test -------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// +// REQUIRES: gpu +// UNSUPPORTED: cuda || hip +// esimd_emulator does not support online-compiler that invokes 'piProgramBuild' +// UNSUPPORTED: esimd_emulator +// RUN: %clangxx -fsycl %s -o %t.out +// RUN: env SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER + +#include "esimd_test_utils.hpp" + +#include +#include +#include + +using namespace cl::sycl; + +int main(void) { + constexpr unsigned Size = 1024 * 128; + constexpr unsigned VL = 16; + + float *A = new float[Size]; + float *B = new float[Size]; + float *C = new float[Size]; + + for (unsigned i = 0; i < Size; ++i) { + A[i] = B[i] = i; + C[i] = 0.0f; + } + + try { + buffer bufa(A, range<1>(Size)); + buffer bufb(B, range<1>(Size)); + buffer bufc(C, range<1>(Size)); + + // We need that many workgroups + range<1> GlobalRange{Size / VL}; + + // We need that many threads in each group + range<1> LocalRange{1}; + + queue q(esimd_test::ESIMDSelector{}, esimd_test::createExceptionHandler()); + + auto dev = q.get_device(); + std::cout << "Running on " << dev.get_info() << "\n"; + + auto e = q.submit([&](handler &cgh) { + auto PA = bufa.get_access(cgh); + auto PB = bufb.get_access(cgh); + auto PC = bufc.get_access(cgh); + cgh.parallel_for( + GlobalRange * LocalRange, [=](id<1> i) SYCL_ESIMD_KERNEL { + using namespace sycl::ext::intel::experimental::esimd; + unsigned int offset = i * VL * sizeof(float); + simd va; + va.copy_from(PA, offset); + simd vb; + vb.copy_from(PB, offset); + simd vc = va + vb; + vc.copy_to(PC, offset); + }); + }); + e.wait(); + } catch (sycl::exception const &e) { + std::cout << "SYCL exception caught: " << e.what() << '\n'; + + delete[] A; + delete[] B; + delete[] C; + return 1; + } + + int err_cnt = 0; + + for (unsigned i = 0; i < Size; ++i) { + if (A[i] + B[i] != C[i]) { + if (++err_cnt < 10) { + std::cout << "failed at index " << i << ", " << C[i] << " != " << A[i] + << " + " << B[i] << "\n"; + } + } + } + if (err_cnt > 0) { + std::cout << " pass rate: " + << ((float)(Size - err_cnt) / (float)Size) * 100.0f << "% (" + << (Size - err_cnt) << "/" << Size << ")\n"; + } + + delete[] A; + delete[] B; + delete[] C; + + std::cout << (err_cnt > 0 ? "FAILED\n" : "Passed\n"); + return err_cnt > 0 ? 1 : 0; +} + +// CHECK: ---> piProgramBuild( +// CHECK: : {{.*}}-vc-codegen +// CHECK: ) ---> pi_result : PI_SUCCESS diff --git a/SYCL/ESIMD/fp_in_phi.cpp b/SYCL/ESIMD/fp_in_phi.cpp index 8e73d272ab..1fface3c09 100644 --- a/SYCL/ESIMD/fp_in_phi.cpp +++ b/SYCL/ESIMD/fp_in_phi.cpp @@ -13,6 +13,8 @@ // RUN: %clangxx -Xclang -fsycl-allow-func-ptr -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_scatter_scaled +// XFAIL: esimd_emulator // // The test checks that ESIMD kernels correctly handle function pointers as // arguments of LLVM's PHI function. diff --git a/SYCL/ESIMD/histogram.cpp b/SYCL/ESIMD/histogram.cpp index 9b062fbe80..9245f57841 100644 --- a/SYCL/ESIMD/histogram.cpp +++ b/SYCL/ESIMD/histogram.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '__esimd_media_ld/st' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/histogram_256_slm.cpp b/SYCL/ESIMD/histogram_256_slm.cpp index 5a072c2f4f..72ad60919f 100644 --- a/SYCL/ESIMD/histogram_256_slm.cpp +++ b/SYCL/ESIMD/histogram_256_slm.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_scatter_scaled +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/histogram_256_slm_spec_2020.cpp b/SYCL/ESIMD/histogram_256_slm_spec_2020.cpp index 3b33141d69..b57e196c9b 100644 --- a/SYCL/ESIMD/histogram_256_slm_spec_2020.cpp +++ b/SYCL/ESIMD/histogram_256_slm_spec_2020.cpp @@ -1,6 +1,8 @@ // TODO enable on Windows // REQUIRES: linux && gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out 16 diff --git a/SYCL/ESIMD/histogram_2d.cpp b/SYCL/ESIMD/histogram_2d.cpp index 24904148e1..2678aecc0e 100644 --- a/SYCL/ESIMD/histogram_2d.cpp +++ b/SYCL/ESIMD/histogram_2d.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '__esimd_media_ld/st' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/histogram_raw_send.cpp b/SYCL/ESIMD/histogram_raw_send.cpp index a63897acfb..c9b5e7a7c6 100644 --- a/SYCL/ESIMD/histogram_raw_send.cpp +++ b/SYCL/ESIMD/histogram_raw_send.cpp @@ -8,6 +8,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '__esimd_media_ld/st' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/linear/linear.cpp b/SYCL/ESIMD/linear/linear.cpp index 6b96a9970f..142a9088a5 100644 --- a/SYCL/ESIMD/linear/linear.cpp +++ b/SYCL/ESIMD/linear/linear.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '__esimd_media_ld/st' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -I%S/.. -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out %S/linear_in.bmp %S/linear_gold_hw.bmp // RUN: %GPU_RUN_PLACEHOLDER %t.out %S/linear_in.bmp %S/linear_gold_hw.bmp diff --git a/SYCL/ESIMD/mandelbrot/mandelbrot.cpp b/SYCL/ESIMD/mandelbrot/mandelbrot.cpp index 57c8e74b72..2509cf8c4a 100644 --- a/SYCL/ESIMD/mandelbrot/mandelbrot.cpp +++ b/SYCL/ESIMD/mandelbrot/mandelbrot.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO/FIXME: esimd_emulator support. Comparison failure. +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -I%S/.. -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out %T/output.ppm %S/golden_hw.ppm diff --git a/SYCL/ESIMD/mandelbrot/mandelbrot_spec.cpp b/SYCL/ESIMD/mandelbrot/mandelbrot_spec.cpp index c31a90f9f3..78685556a5 100644 --- a/SYCL/ESIMD/mandelbrot/mandelbrot_spec.cpp +++ b/SYCL/ESIMD/mandelbrot/mandelbrot_spec.cpp @@ -8,6 +8,8 @@ // TODO enable on Windows // REQUIRES: linux && gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -I%S/.. -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out %T/output_spec.ppm %S/golden_hw.ppm 512 -2.09798 -1.19798 0.004 4.0 diff --git a/SYCL/ESIMD/matrix_transpose.cpp b/SYCL/ESIMD/matrix_transpose.cpp index dbe33d616b..ddde4b8f3e 100644 --- a/SYCL/ESIMD/matrix_transpose.cpp +++ b/SYCL/ESIMD/matrix_transpose.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/matrix_transpose2.cpp b/SYCL/ESIMD/matrix_transpose2.cpp index b587769092..2d2857c5e4 100644 --- a/SYCL/ESIMD/matrix_transpose2.cpp +++ b/SYCL/ESIMD/matrix_transpose2.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '__esimd_media_ld/st' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/matrix_transpose_glb.cpp b/SYCL/ESIMD/matrix_transpose_glb.cpp index ec13559499..aae61928b9 100644 --- a/SYCL/ESIMD/matrix_transpose_glb.cpp +++ b/SYCL/ESIMD/matrix_transpose_glb.cpp @@ -20,6 +20,11 @@ using namespace cl::sycl; using namespace std; using namespace sycl::ext::intel::experimental::esimd; +// TODO: For esimd_emulator, completion time depends on CPU +// performance. MAX_SIZE may need to be adjusted in order to prevent +// timeout failure because of slow CPU. +#define ESIMD_EMULATOR_MATRIX_MAX_SIZE (1U << 10) + void initMatrix(int *M, unsigned N) { assert(N >= 8 && (((N - 1) & N) == 0) && "only power of 2 (>= 16) is supported"); @@ -239,6 +244,15 @@ ESIMD_INLINE void transpose16(int *buf, int MZ, int block_col, int block_row) { bool runTest(unsigned MZ, unsigned block_size) { queue q(esimd_test::ESIMDSelector{}, esimd_test::createExceptionHandler(), property::queue::enable_profiling{}); + if ((MZ > ESIMD_EMULATOR_MATRIX_MAX_SIZE) && + (q.get_context().get_platform().get_backend() == + cl::sycl::backend::ext_intel_esimd_emulator)) { + cerr << "\nMatrix larger than limit size (" + << ESIMD_EMULATOR_MATRIX_MAX_SIZE + << ") is ignored for esimd_emulator due to timeout.\n"; + cerr << "Requested matrix size = " << MZ << "\n"; + return true; + } int *M = malloc_shared(MZ * MZ, q); initMatrix(M, MZ); diff --git a/SYCL/ESIMD/matrix_transpose_usm.cpp b/SYCL/ESIMD/matrix_transpose_usm.cpp index 6268295b5b..bec8ffcfee 100644 --- a/SYCL/ESIMD/matrix_transpose_usm.cpp +++ b/SYCL/ESIMD/matrix_transpose_usm.cpp @@ -17,6 +17,11 @@ #include #include +// TODO: For esimd_emulator, completion time depends on CPU +// performance. MAX_SIZE may need to be adjusted in order to prevent +// timeout failure because of slow CPU. +#define ESIMD_EMULATOR_MATRIX_MAX_SIZE (1U << 8) + using namespace cl::sycl; using namespace std; using namespace sycl::ext::intel::experimental::esimd; @@ -242,6 +247,15 @@ ESIMD_INLINE void transpose16(int *buf, int MZ, int block_col, int block_row) { } bool runTest(queue &q, unsigned MZ, unsigned block_size) { + if ((MZ > ESIMD_EMULATOR_MATRIX_MAX_SIZE) && + (q.get_context().get_platform().get_backend() == + cl::sycl::backend::ext_intel_esimd_emulator)) { + cerr << "\nMatrix larger than limit size (" + << ESIMD_EMULATOR_MATRIX_MAX_SIZE + << ") is ignored for esimd_emulator due to timeout.\n"; + cerr << "Requested matrix size = " << MZ << "\n"; + return true; + } int *M = malloc_shared(MZ * MZ, q); initMatrix(M, MZ); diff --git a/SYCL/ESIMD/printf.cpp b/SYCL/ESIMD/printf.cpp index cb7d6e1cf0..224803b9b5 100644 --- a/SYCL/ESIMD/printf.cpp +++ b/SYCL/ESIMD/printf.cpp @@ -9,6 +9,8 @@ // REQUIRES: gpu // UNSUPPORTED: cuda || hip // CUDA and HIP don't support printf. +// TODO: esimd_emulator support - enable 'single_task()' +// XFAIL: esimd_emulator // // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER diff --git a/SYCL/ESIMD/regression/big_const_initializer.cpp b/SYCL/ESIMD/regression/big_const_initializer.cpp index 9941f6b8d5..b312c309c6 100644 --- a/SYCL/ESIMD/regression/big_const_initializer.cpp +++ b/SYCL/ESIMD/regression/big_const_initializer.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_scatter_scaled +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/regression/globals.cpp b/SYCL/ESIMD/regression/globals.cpp index 2345027f8f..731655a5ae 100644 --- a/SYCL/ESIMD/regression/globals.cpp +++ b/SYCL/ESIMD/regression/globals.cpp @@ -2,6 +2,8 @@ // UNSUPPORTED: cuda || hip // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// TODO: esimd_emulator support - enable 'this_sub_group' +// XFAIL: esimd_emulator // This test checks loads from SPIRV builtin globals work correctly in ESIMD. diff --git a/SYCL/ESIMD/regression/operator_decrement.cpp b/SYCL/ESIMD/regression/operator_decrement.cpp index 379af94a43..355a6e7b1b 100644 --- a/SYCL/ESIMD/regression/operator_decrement.cpp +++ b/SYCL/ESIMD/regression/operator_decrement.cpp @@ -9,6 +9,8 @@ // // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/regression/unused_load.cpp b/SYCL/ESIMD/regression/unused_load.cpp index 3d9ae61e46..f1ab7372be 100644 --- a/SYCL/ESIMD/regression/unused_load.cpp +++ b/SYCL/ESIMD/regression/unused_load.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/slm_barrier.cpp b/SYCL/ESIMD/slm_barrier.cpp index 2e58afc433..f8c5f2b83a 100644 --- a/SYCL/ESIMD/slm_barrier.cpp +++ b/SYCL/ESIMD/slm_barrier.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_scatter4_scaled +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/slm_split_barrier.cpp b/SYCL/ESIMD/slm_split_barrier.cpp index 9b180bca68..86b6734f03 100644 --- a/SYCL/ESIMD/slm_split_barrier.cpp +++ b/SYCL/ESIMD/slm_split_barrier.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_scatter4_scaled +// XFAIL: esimd_emulator #include "esimd_test_utils.hpp" diff --git a/SYCL/ESIMD/spec_const/spec_const_bool.cpp b/SYCL/ESIMD/spec_const/spec_const_bool.cpp index 7476d496aa..921c118584 100644 --- a/SYCL/ESIMD/spec_const/spec_const_bool.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_bool.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_char.cpp b/SYCL/ESIMD/spec_const/spec_const_char.cpp index 3d7b6137c7..8b77593edf 100644 --- a/SYCL/ESIMD/spec_const/spec_const_char.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_char.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_double.cpp b/SYCL/ESIMD/spec_const/spec_const_double.cpp index 146a170f37..523f8a288b 100644 --- a/SYCL/ESIMD/spec_const/spec_const_double.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_double.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_float.cpp b/SYCL/ESIMD/spec_const/spec_const_float.cpp index 04e0209862..3d02a1ccc7 100644 --- a/SYCL/ESIMD/spec_const/spec_const_float.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_float.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_int.cpp b/SYCL/ESIMD/spec_const/spec_const_int.cpp index 13b3b4227f..ced19c7968 100644 --- a/SYCL/ESIMD/spec_const/spec_const_int.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_int.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_int64.cpp b/SYCL/ESIMD/spec_const/spec_const_int64.cpp index f67085eaf7..f5b5bfdbb4 100644 --- a/SYCL/ESIMD/spec_const/spec_const_int64.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_int64.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_short.cpp b/SYCL/ESIMD/spec_const/spec_const_short.cpp index 3c296dab27..cd8d3fc618 100644 --- a/SYCL/ESIMD/spec_const/spec_const_short.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_short.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_uchar.cpp b/SYCL/ESIMD/spec_const/spec_const_uchar.cpp index cd934e2ba0..648affdf45 100644 --- a/SYCL/ESIMD/spec_const/spec_const_uchar.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_uchar.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_uint.cpp b/SYCL/ESIMD/spec_const/spec_const_uint.cpp index a0f93df45b..bda4af2e55 100644 --- a/SYCL/ESIMD/spec_const/spec_const_uint.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_uint.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_uint64.cpp b/SYCL/ESIMD/spec_const/spec_const_uint64.cpp index aa6a2814bf..df1e9a1d69 100644 --- a/SYCL/ESIMD/spec_const/spec_const_uint64.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_uint64.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/spec_const/spec_const_ushort.cpp b/SYCL/ESIMD/spec_const/spec_const_ushort.cpp index f0558d7b74..283f847361 100644 --- a/SYCL/ESIMD/spec_const/spec_const_ushort.cpp +++ b/SYCL/ESIMD/spec_const/spec_const_ushort.cpp @@ -9,6 +9,8 @@ // RUN: %clangxx -fsycl -I%S/.. %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable online compiler +// XFAIL: esimd_emulator #include diff --git a/SYCL/ESIMD/sycl_esimd_mix.cpp b/SYCL/ESIMD/sycl_esimd_mix.cpp index b3052f374f..9a7b561b08 100644 --- a/SYCL/ESIMD/sycl_esimd_mix.cpp +++ b/SYCL/ESIMD/sycl_esimd_mix.cpp @@ -10,9 +10,11 @@ // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// esimd_emulator supports only ESIMD kernels +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out -// RUN: env SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK,CHECK-NO-VAR -// RUN: env SYCL_PROGRAM_COMPILE_OPTIONS="-g" SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK,CHECK-WITH-VAR +// RUN: %GPU_RUN_PLACEHOLDER %t.out +// RUN: env SYCL_PROGRAM_COMPILE_OPTIONS="-g" %GPU_RUN_PLACEHOLDER %t.out #include "esimd_test_utils.hpp" @@ -123,24 +125,5 @@ int main(void) { return 0; } -// Regular SYCL kernel is compiled without -vc-codegen option - -// CHECK-LABEL: ---> piProgramBuild( -// CHECK-NOT: -vc-codegen -// CHECK-WITH-VAR: -g -// CHECK-NOT: -vc-codegen -// CHECK: ) ---> pi_result : PI_SUCCESS -// CHECK-LABEL: ---> piKernelCreate( -// CHECK: : {{.*}}SyclKernel -// CHECK: ) ---> pi_result : PI_SUCCESS - -// For ESIMD kernels, -vc-codegen option is always preserved, -// regardless of SYCL_PROGRAM_COMPILE_OPTIONS value. - -// CHECK-LABEL: ---> piProgramBuild( -// CHECK-NO-VAR: -vc-codegen -// CHECK-WITH-VAR: -g -vc-codegen -// CHECK: ) ---> pi_result : PI_SUCCESS -// CHECK-LABEL: ---> piKernelCreate( -// CHECK: : {{.*}}EsimdKernel -// CHECK: ) ---> pi_result : PI_SUCCESS +// 'CHECK' commands for checking 'vc-codegen' are moved to +// 'sycl_esimd_mix_check_build_opts.cpp' diff --git a/SYCL/ESIMD/sycl_esimd_mix_check_build_opts.cpp b/SYCL/ESIMD/sycl_esimd_mix_check_build_opts.cpp new file mode 100644 index 0000000000..c391b199bc --- /dev/null +++ b/SYCL/ESIMD/sycl_esimd_mix_check_build_opts.cpp @@ -0,0 +1,148 @@ +//==--- sycl_esimd_mix_check_build_opts.cpp - DPC++ ESIMD on-device test ---==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// +// This is basic test for mixing SYCL and ESIMD kernels in the same source and +// in the same program . + +// REQUIRES: gpu +// UNSUPPORTED: cuda || hip +// esimd_emulator does not support online-compiler that invokes 'piProgramBuild' +// UNSUPPORTED: esimd_emulator +// RUN: %clangxx -fsycl %s -o %t.out +// RUN: env SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK,CHECK-NO-VAR +// RUN: env SYCL_PROGRAM_COMPILE_OPTIONS="-g" SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER --check-prefixes=CHECK,CHECK-WITH-VAR + +#include "esimd_test_utils.hpp" + +#include +#include +#include + +using namespace cl::sycl; + +bool checkResult(const std::vector &A, int Inc) { + int err_cnt = 0; + unsigned Size = A.size(); + + for (unsigned i = 0; i < Size; ++i) { + if (A[i] != i + Inc) + if (++err_cnt < 10) + std::cerr << "failed at A[" << i << "]: " << A[i] << " != " << i + Inc + << "\n"; + } + + if (err_cnt > 0) { + std::cout << " pass rate: " + << ((float)(Size - err_cnt) / (float)Size) * 100.0f << "% (" + << (Size - err_cnt) << "/" << Size << ")\n"; + return false; + } + return true; +} + +int main(void) { + constexpr unsigned Size = 32; + constexpr unsigned VL = 16; + + std::vector A(Size); + + for (unsigned i = 0; i < Size; ++i) { + A[i] = i; + } + + try { + buffer bufa(A.data(), range<1>(Size)); + + // We need that many workgroups + cl::sycl::range<1> GlobalRange{Size}; + // We need that many threads in each group + cl::sycl::range<1> LocalRange{1}; + + queue q(gpu_selector{}, esimd_test::createExceptionHandler()); + + auto dev = q.get_device(); + std::cout << "Running on " << dev.get_info() << "\n"; + + auto e = q.submit([&](handler &cgh) { + auto PA = bufa.get_access(cgh); + cgh.parallel_for(GlobalRange * LocalRange, + [=](id<1> i) { PA[i] = PA[i] + 1; }); + }); + e.wait(); + } catch (cl::sycl::exception const &e) { + std::cout << "SYCL exception caught: " << e.what() << '\n'; + return 2; + } + + if (checkResult(A, 1)) { + std::cout << "SYCL kernel passed\n"; + } else { + std::cout << "SYCL kernel failed\n"; + return 1; + } + + try { + buffer bufa(A.data(), range<1>(Size)); + + // We need that many workgroups + cl::sycl::range<1> GlobalRange{Size / VL}; + // We need that many threads in each group + cl::sycl::range<1> LocalRange{1}; + + queue q(esimd_test::ESIMDSelector{}, esimd_test::createExceptionHandler()); + + auto dev = q.get_device(); + std::cout << "Running on " << dev.get_info() << "\n"; + + auto e = q.submit([&](handler &cgh) { + auto PA = bufa.get_access(cgh); + cgh.parallel_for( + GlobalRange * LocalRange, [=](id<1> i) SYCL_ESIMD_KERNEL { + using namespace sycl::ext::intel::experimental::esimd; + unsigned int offset = i * VL * sizeof(float); + simd va; + va.copy_from(PA, offset); + simd vc = va + 1; + vc.copy_to(PA, offset); + }); + }); + e.wait(); + } catch (cl::sycl::exception const &e) { + std::cout << "SYCL exception caught: " << e.what() << '\n'; + return 2; + } + + if (checkResult(A, 2)) { + std::cout << "ESIMD kernel passed\n"; + } else { + std::cout << "ESIMD kernel failed\n"; + return 1; + } + return 0; +} + +// Regular SYCL kernel is compiled without -vc-codegen option + +// CHECK-LABEL: ---> piProgramBuild( +// CHECK-NOT: -vc-codegen +// CHECK-WITH-VAR: -g +// CHECK-NOT: -vc-codegen +// CHECK: ) ---> pi_result : PI_SUCCESS +// CHECK-LABEL: ---> piKernelCreate( +// CHECK: : {{.*}}SyclKernel +// CHECK: ) ---> pi_result : PI_SUCCESS + +// For ESIMD kernels, -vc-codegen option is always preserved, +// regardless of SYCL_PROGRAM_COMPILE_OPTIONS value. + +// CHECK-LABEL: ---> piProgramBuild( +// CHECK-NO-VAR: -vc-codegen +// CHECK-WITH-VAR: -g -vc-codegen +// CHECK: ) ---> pi_result : PI_SUCCESS +// CHECK-LABEL: ---> piKernelCreate( +// CHECK: : {{.*}}EsimdKernel +// CHECK: ) ---> pi_result : PI_SUCCESS diff --git a/SYCL/ESIMD/vadd_1d.cpp b/SYCL/ESIMD/vadd_1d.cpp index 78bf4e04a5..48bc4dac52 100644 --- a/SYCL/ESIMD/vadd_1d.cpp +++ b/SYCL/ESIMD/vadd_1d.cpp @@ -7,8 +7,10 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out -// RUN: env SYCL_PI_TRACE=-1 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER +// RUN: %GPU_RUN_PLACEHOLDER %t.out #include "esimd_test_utils.hpp" @@ -97,6 +99,5 @@ int main(void) { return err_cnt > 0 ? 1 : 0; } -// CHECK: ---> piProgramBuild( -// CHECK: : {{.*}}-vc-codegen -// CHECK: ) ---> pi_result : PI_SUCCESS +// 'CHECK' commands for checking 'vc-codegen' are moved to +// 'sycl_esimd_mix_check_build_opts.cpp' diff --git a/SYCL/ESIMD/vadd_2d.cpp b/SYCL/ESIMD/vadd_2d.cpp index 3037281833..89a7874a0f 100644 --- a/SYCL/ESIMD/vadd_2d.cpp +++ b/SYCL/ESIMD/vadd_2d.cpp @@ -7,6 +7,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '__esimd_media_ld/st' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/ESIMD/vadd_2d_acc.cpp b/SYCL/ESIMD/vadd_2d_acc.cpp index 7533573e7f..f3020e0c28 100644 --- a/SYCL/ESIMD/vadd_2d_acc.cpp +++ b/SYCL/ESIMD/vadd_2d_acc.cpp @@ -10,6 +10,8 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable __esimd_oword_ld_unaligned +// XFAIL: esimd_emulator // The test checks that 2D workitem addressing works correctly with SIMD // kernels. diff --git a/SYCL/ESIMD/vadd_raw_send.cpp b/SYCL/ESIMD/vadd_raw_send.cpp index 649d05aaf7..1e1096630d 100644 --- a/SYCL/ESIMD/vadd_raw_send.cpp +++ b/SYCL/ESIMD/vadd_raw_send.cpp @@ -8,6 +8,8 @@ //===----------------------------------------------------------------------===// // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// TODO: esimd_emulator support - enable '()' operator for 'main::lambda' +// XFAIL: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/FilterSelector/select.cpp b/SYCL/FilterSelector/select.cpp index 710037ae73..e5cdb08bde 100644 --- a/SYCL/FilterSelector/select.cpp +++ b/SYCL/FilterSelector/select.cpp @@ -6,6 +6,7 @@ // // Failing on HIP AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator //==------------------- select.cpp - filter_selector test ------------------==// // @@ -43,6 +44,7 @@ int main() { bool HasHIPDevices = false; bool HasOpenCLGPU = false; bool HasLevelZeroGPU = false; + bool HasESIMDEmuDevices = false; for (auto &Dev : Devs) { auto Backend = Dev.get_platform().get_backend(); @@ -54,6 +56,8 @@ int main() { HasCUDADevices = true; } else if (Backend == backend::ext_oneapi_hip) { HasHIPDevices = true; + } else if (Backend == backend::ext_intel_esimd_emulator) { + HasESIMDEmuDevices = true; } } @@ -72,6 +76,7 @@ int main() { std::cout << "HasHIPDevices = " << HasHIPDevices << std::endl; std::cout << "HasOpenCLGPU = " << HasOpenCLGPU << std::endl; std::cout << "HasLevelZeroGPU = " << HasLevelZeroGPU << std::endl; + std::cout << "HasESIMDEmuDevices = " << HasESIMDEmuDevices << std::endl; if (!CPUs.empty()) { std::cout << "Test 'cpu'"; @@ -217,5 +222,18 @@ int main() { std::cout << "...PASS" << std::endl; } + if (HasESIMDEmuDevices) { + std::cout << "Test 'esimd_emulator'"; + device d21(ext::oneapi::filter_selector("esimd_emulator")); + assert(d21.get_platform().get_backend() == + backend::ext_intel_esimd_emulator); + std::cout << "...PASS" << std::endl; + + std::cout << "test 'esimd_emulator:gpu'"; + device d22(ext::oneapi::filter_selector("esimd_emulator:gpu")); + assert(d22.is_gpu() && d22.get_platform().get_backend() == + backend::ext_intel_esimd_emulator); + std::cout << "...PASS" << std::endl; + } return 0; } diff --git a/SYCL/FilterSelector/select_device_esimd_emulator.cpp b/SYCL/FilterSelector/select_device_esimd_emulator.cpp new file mode 100644 index 0000000000..41e7cab8a4 --- /dev/null +++ b/SYCL/FilterSelector/select_device_esimd_emulator.cpp @@ -0,0 +1,68 @@ +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out +// RUN: env SYCL_DEVICE_FILTER=esimd_emulator:gpu,host %t.out +// +// Checks if only specified device types can be acquired from select_device +// when SYCL_DEVICE_FILTER is set +// Checks that no device is selected when no device of desired type is +// available. +// +// REQUIRES: esimd_emulator,gpu,host + +#include +#include + +using namespace cl::sycl; +using namespace std; + +int main() { + const char *envVal = getenv("SYCL_DEVICE_FILTER"); + string forcedPIs; + if (envVal) { + cout << "SYCL_DEVICE_FILTER=" << envVal << std::endl; + forcedPIs = envVal; + } + + { + default_selector ds; + device d = ds.select_device(); + string name = d.get_platform().get_info(); + assert(name.find("ESIMD_EMULATOR") != string::npos); + cout << "ESIMD_EMULATOR GPU Device is found: " << boolalpha << d.is_gpu() + << std::endl; + } + { + gpu_selector gs; + device d = gs.select_device(); + string name = d.get_platform().get_info(); + assert(name.find("ESIMD_EMULATOR") != string::npos); + cout << name << " is found: " << boolalpha << d.is_gpu() << std::endl; + } + { + cpu_selector cs; + try { + device d = cs.select_device(); + cerr << "CPU device is found in error: " << d.is_cpu() << std::endl; + return -1; + } catch (...) { + cout << "Expectedly, cpu device is not found." << std::endl; + } + } + // HOST device is always available regardless of SYCL_DEVICE_FILTER + { + host_selector hs; + device d = hs.select_device(); + cout << "HOST device is found: " << d.is_host() << std::endl; + } + { + accelerator_selector as; + try { + device d = as.select_device(); + cerr << "ACC device is found in error: " << d.is_accelerator() + << std::endl; + } catch (...) { + cout << "Expectedly, ACC device is not found." << std::endl; + } + } + + return 0; +} diff --git a/SYCL/Functor/functor_inheritance.cpp b/SYCL/Functor/functor_inheritance.cpp index da477f5e65..b0bbd542ae 100644 --- a/SYCL/Functor/functor_inheritance.cpp +++ b/SYCL/Functor/functor_inheritance.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Functor/kernel_functor.cpp b/SYCL/Functor/kernel_functor.cpp index a43556d773..dfd5a6e4a0 100644 --- a/SYCL/Functor/kernel_functor.cpp +++ b/SYCL/Functor/kernel_functor.cpp @@ -3,6 +3,7 @@ // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==--- kernel_functor.cpp - // This test illustrates defining kernels as named function objects (functors) diff --git a/SYCL/GroupAlgorithm/SYCL2020/permute_select.cpp b/SYCL/GroupAlgorithm/SYCL2020/permute_select.cpp index b0fbb35982..7c50e2f68e 100644 --- a/SYCL/GroupAlgorithm/SYCL2020/permute_select.cpp +++ b/SYCL/GroupAlgorithm/SYCL2020/permute_select.cpp @@ -2,6 +2,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // //==------------ permute_select.cpp -*- C++ -*-----------------------------===// // diff --git a/SYCL/GroupAlgorithm/SYCL2020/shift_left_right.cpp b/SYCL/GroupAlgorithm/SYCL2020/shift_left_right.cpp index e3c97cac0e..19e6ca4c3c 100644 --- a/SYCL/GroupAlgorithm/SYCL2020/shift_left_right.cpp +++ b/SYCL/GroupAlgorithm/SYCL2020/shift_left_right.cpp @@ -2,6 +2,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // //==------------ shift_left_right.cpp -*- C++ -*----------------------------==// // diff --git a/SYCL/GroupAlgorithm/back_to_back_collectives.cpp b/SYCL/GroupAlgorithm/back_to_back_collectives.cpp index 5bdd126138..807e9deeaa 100644 --- a/SYCL/GroupAlgorithm/back_to_back_collectives.cpp +++ b/SYCL/GroupAlgorithm/back_to_back_collectives.cpp @@ -6,6 +6,7 @@ // // Missing __spirv_GroupIAdd on AMD: // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/GroupLocalMemory/group_local_memory.cpp b/SYCL/GroupLocalMemory/group_local_memory.cpp index a2473d7aa9..3942292ba8 100644 --- a/SYCL/GroupLocalMemory/group_local_memory.cpp +++ b/SYCL/GroupLocalMemory/group_local_memory.cpp @@ -4,6 +4,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/GroupLocalMemory/no_early_opt.cpp b/SYCL/GroupLocalMemory/no_early_opt.cpp index 3d68ec0e77..8f161d1186 100644 --- a/SYCL/GroupLocalMemory/no_early_opt.cpp +++ b/SYCL/GroupLocalMemory/no_early_opt.cpp @@ -4,6 +4,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator // The test checks that multiple calls to the same template instantiation of a // group local memory function result in separate allocations, even with device diff --git a/SYCL/HierPar/hier_par_basic.cpp b/SYCL/HierPar/hier_par_basic.cpp index d21e710b83..7871977314 100644 --- a/SYCL/HierPar/hier_par_basic.cpp +++ b/SYCL/HierPar/hier_par_basic.cpp @@ -14,6 +14,7 @@ // // Linking issues on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator // This test checks hierarchical parallelism invocation APIs, but without any // data or code with side-effects between the work group and work item scopes. diff --git a/SYCL/HierPar/hier_par_wgscope.cpp b/SYCL/HierPar/hier_par_wgscope.cpp index 190732fd84..47316825d1 100644 --- a/SYCL/HierPar/hier_par_wgscope.cpp +++ b/SYCL/HierPar/hier_par_wgscope.cpp @@ -7,6 +7,7 @@ // // Linking issues on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator //==- hier_par_wgscope.cpp --- hierarchical parallelism test for WG scope---==// // diff --git a/SYCL/HierPar/hier_par_wgscope_O0.cpp b/SYCL/HierPar/hier_par_wgscope_O0.cpp index e87fdeb9d2..e5088222d7 100644 --- a/SYCL/HierPar/hier_par_wgscope_O0.cpp +++ b/SYCL/HierPar/hier_par_wgscope_O0.cpp @@ -15,6 +15,7 @@ // // Linking issues on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator // 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 diff --git a/SYCL/HostInteropTask/host-task-dependency2.cpp b/SYCL/HostInteropTask/host-task-dependency2.cpp index 72ad4cf229..a16bed6ff8 100644 --- a/SYCL/HostInteropTask/host-task-dependency2.cpp +++ b/SYCL/HostInteropTask/host-task-dependency2.cpp @@ -8,6 +8,8 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out 10 // RUN: %ACC_RUN_PLACEHOLDER %t.out 10 +// UNSUPPORTED: esimd_emulator + #include #include diff --git a/SYCL/HostInteropTask/host-task-dependency3.cpp b/SYCL/HostInteropTask/host-task-dependency3.cpp index 8e8486f606..0d7aaf153d 100644 --- a/SYCL/HostInteropTask/host-task-dependency3.cpp +++ b/SYCL/HostInteropTask/host-task-dependency3.cpp @@ -7,6 +7,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out 10 // RUN: %GPU_RUN_PLACEHOLDER %t.out 10 // RUN: %ACC_RUN_PLACEHOLDER %t.out 10 +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/HostInteropTask/host-task-failure.cpp b/SYCL/HostInteropTask/host-task-failure.cpp index 0ba45cee11..6ab7e1ecca 100644 --- a/SYCL/HostInteropTask/host-task-failure.cpp +++ b/SYCL/HostInteropTask/host-task-failure.cpp @@ -2,6 +2,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/HostInteropTask/host-task-two-queues.cpp b/SYCL/HostInteropTask/host-task-two-queues.cpp index 9da120f575..eaaaa111b9 100644 --- a/SYCL/HostInteropTask/host-task-two-queues.cpp +++ b/SYCL/HostInteropTask/host-task-two-queues.cpp @@ -5,6 +5,7 @@ // // TODO: Flaky fail on Level Zero that is why mark as unsupported temporarily. // UNSUPPORTED: level_zero +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/HostInteropTask/host-task.cpp b/SYCL/HostInteropTask/host-task.cpp index d9823f7e9e..e198186ab2 100644 --- a/SYCL/HostInteropTask/host-task.cpp +++ b/SYCL/HostInteropTask/host-task.cpp @@ -10,6 +10,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out 3 // RUN: %GPU_RUN_PLACEHOLDER %t.out 3 // RUN: %ACC_RUN_PLACEHOLDER %t.out 3 +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/InorderQueue/in_order_buffs.cpp b/SYCL/InorderQueue/in_order_buffs.cpp index be96e3b849..0f3047455a 100644 --- a/SYCL/InorderQueue/in_order_buffs.cpp +++ b/SYCL/InorderQueue/in_order_buffs.cpp @@ -2,6 +2,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==-------- ordered_buffs.cpp - SYCL buffers in ordered queues test--------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. diff --git a/SYCL/InorderQueue/in_order_dmemll.cpp b/SYCL/InorderQueue/in_order_dmemll.cpp index 32518c3e8e..6df6c4aba6 100644 --- a/SYCL/InorderQueue/in_order_dmemll.cpp +++ b/SYCL/InorderQueue/in_order_dmemll.cpp @@ -2,6 +2,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator // //==----------- ordered_dmemll.cpp - Device Memory Linked List test --------==// // It uses an ordered queue where explicit waiting is not necessary between diff --git a/SYCL/InorderQueue/in_order_kernels.cpp b/SYCL/InorderQueue/in_order_kernels.cpp index 10147e2b55..98d7be9b2b 100644 --- a/SYCL/InorderQueue/in_order_kernels.cpp +++ b/SYCL/InorderQueue/in_order_kernels.cpp @@ -7,6 +7,7 @@ // // Linking issues on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator // SYCL ordered queue kernel shortcut test // diff --git a/SYCL/InorderQueue/in_order_usm_explicit.cpp b/SYCL/InorderQueue/in_order_usm_explicit.cpp index 1fddaeaf04..1eb032ec1c 100644 --- a/SYCL/InorderQueue/in_order_usm_explicit.cpp +++ b/SYCL/InorderQueue/in_order_usm_explicit.cpp @@ -2,6 +2,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // SYCL in ordered queues explicit USM test. // Simple test checking explicit USM functionality using a Queue with the // in_order property. diff --git a/SYCL/KernelAndProgram/build-log.cpp b/SYCL/KernelAndProgram/build-log.cpp index cbb8560c04..83270e3a6f 100644 --- a/SYCL/KernelAndProgram/build-log.cpp +++ b/SYCL/KernelAndProgram/build-log.cpp @@ -1,4 +1,5 @@ // XFAIL: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -DGPU %s -o %t_gpu.out // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/KernelAndProgram/cache-build-result.cpp b/SYCL/KernelAndProgram/cache-build-result.cpp index d9458ec16b..cb1767a96e 100644 --- a/SYCL/KernelAndProgram/cache-build-result.cpp +++ b/SYCL/KernelAndProgram/cache-build-result.cpp @@ -4,6 +4,7 @@ // RUN: env SYCL_CACHE_PERSISTENT=1 %GPU_RUN_PLACEHOLDER %t_gpu.out // RUN: env SYCL_CACHE_PERSISTENT=1 %ACC_RUN_PLACEHOLDER %t.out // XFAIL: cuda || hip +// UNSUPPORTED: esimd_emulator #include SYCL_EXTERNAL diff --git a/SYCL/KernelAndProgram/kernel-bundle-merge-options-env.cpp b/SYCL/KernelAndProgram/kernel-bundle-merge-options-env.cpp index 21c3aefc6a..0d39cf22ab 100644 --- a/SYCL/KernelAndProgram/kernel-bundle-merge-options-env.cpp +++ b/SYCL/KernelAndProgram/kernel-bundle-merge-options-env.cpp @@ -3,6 +3,7 @@ // RUN: %GPU_RUN_PLACEHOLDER SYCL_PI_TRACE=-1 SYCL_PROGRAM_COMPILE_OPTIONS=-DENV_COMPILE_OPTS SYCL_PROGRAM_LINK_OPTIONS=-DENV_LINK_OPTS %t.out %GPU_CHECK_PLACEHOLDER // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include "kernel-bundle-merge-options.hpp" // CHECK: piProgramBuild diff --git a/SYCL/KernelAndProgram/kernel-bundle-merge-options.cpp b/SYCL/KernelAndProgram/kernel-bundle-merge-options.cpp index c2abf27bcf..4064472a09 100644 --- a/SYCL/KernelAndProgram/kernel-bundle-merge-options.cpp +++ b/SYCL/KernelAndProgram/kernel-bundle-merge-options.cpp @@ -2,6 +2,7 @@ // RUN: %GPU_RUN_PLACEHOLDER SYCL_PI_TRACE=-1 %t.out %GPU_CHECK_PLACEHOLDER // REQUIRES: gpu // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // Debug option -g is not passed to device code compiler when CL-style driver // is used and /DEBUG options is passed. diff --git a/SYCL/KernelAndProgram/multiple-kernel-linking.cpp b/SYCL/KernelAndProgram/multiple-kernel-linking.cpp index 0294b4bf20..26e84dc72c 100644 --- a/SYCL/KernelAndProgram/multiple-kernel-linking.cpp +++ b/SYCL/KernelAndProgram/multiple-kernel-linking.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fno-sycl-early-optimizations -fsycl-device-code-split=per_kernel %s -o %t_per_kernel.out // RUN: %clangxx -fsycl -fno-sycl-early-optimizations -fsycl-device-code-split=per_source %s -o %t_per_source.out diff --git a/SYCL/KernelParams/non-standard-layout.cpp b/SYCL/KernelParams/non-standard-layout.cpp index 8d7a4e55aa..9304a71800 100644 --- a/SYCL/KernelParams/non-standard-layout.cpp +++ b/SYCL/KernelParams/non-standard-layout.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/KernelParams/struct_kernel_param.cpp b/SYCL/KernelParams/struct_kernel_param.cpp index 9ffe4724ce..d282f150e1 100644 --- a/SYCL/KernelParams/struct_kernel_param.cpp +++ b/SYCL/KernelParams/struct_kernel_param.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==-struct_kernel_param.cpp-Checks passing structs as kernel params--------==// // diff --git a/SYCL/KernelParams/union_kernel_param.cpp b/SYCL/KernelParams/union_kernel_param.cpp index 2a8c02661a..24b624ac7b 100644 --- a/SYCL/KernelParams/union_kernel_param.cpp +++ b/SYCL/KernelParams/union_kernel_param.cpp @@ -5,6 +5,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Plugin/enqueue-arg-order-buffer.cpp b/SYCL/Plugin/enqueue-arg-order-buffer.cpp index a517896e6e..803a175fd1 100644 --- a/SYCL/Plugin/enqueue-arg-order-buffer.cpp +++ b/SYCL/Plugin/enqueue-arg-order-buffer.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip_nvidia +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: env SYCL_PI_TRACE=2 %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER // RUN: env SYCL_PI_TRACE=2 %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER diff --git a/SYCL/Plugin/enqueue-arg-order-image.cpp b/SYCL/Plugin/enqueue-arg-order-image.cpp index 22cec83165..558312effb 100644 --- a/SYCL/Plugin/enqueue-arg-order-image.cpp +++ b/SYCL/Plugin/enqueue-arg-order-image.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // Native images are created with host pointers only with host unified memory // support, enforce it for this test. diff --git a/SYCL/Plugin/retain_events.cpp b/SYCL/Plugin/retain_events.cpp index 7611a54ad9..abc6a37059 100644 --- a/SYCL/Plugin/retain_events.cpp +++ b/SYCL/Plugin/retain_events.cpp @@ -1,6 +1,7 @@ // REQUIRES: gpu // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Plugin/sycl-ls-gpu-default-any.cpp b/SYCL/Plugin/sycl-ls-gpu-default-any.cpp index 909fcb861d..9132ce8472 100644 --- a/SYCL/Plugin/sycl-ls-gpu-default-any.cpp +++ b/SYCL/Plugin/sycl-ls-gpu-default-any.cpp @@ -21,3 +21,4 @@ // The test crashed on CUDA CI machines with the latest OpenCL GPU RT // (21.19.19792). // UNSUPPORTED: cuda || hip +// XFAIL: esimd_emulator diff --git a/SYCL/Plugin/sycl-ls-gpu-esimd-emulator.cpp b/SYCL/Plugin/sycl-ls-gpu-esimd-emulator.cpp new file mode 100644 index 0000000000..a8e7182631 --- /dev/null +++ b/SYCL/Plugin/sycl-ls-gpu-esimd-emulator.cpp @@ -0,0 +1,15 @@ +// REQUIRES: gpu, esimd_emulator + +// RUN: sycl-ls --verbose >%t.default.out +// RUN: FileCheck %s --check-prefixes=CHECK-GPU-BUILTIN,CHECK-GPU-CUSTOM --input-file %t.default.out + +// CHECK-GPU-BUILTIN: gpu_selector(){{.*}}gpu, {{.*}}ESIMD_EMULATOR +// CHECK-GPU-CUSTOM: custom_selector(gpu){{.*}}gpu, {{.*}}ESIMD_EMULATOR + +//= sycl-ls-gpu-esimd-emulator.cpp - Test ESIMD_EMULATOR selected gpu device =// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// diff --git a/SYCL/Plugin/sycl-ls-gpu-sycl-be.cpp b/SYCL/Plugin/sycl-ls-gpu-sycl-be.cpp index 2a38fa6b63..b751b452ce 100644 --- a/SYCL/Plugin/sycl-ls-gpu-sycl-be.cpp +++ b/SYCL/Plugin/sycl-ls-gpu-sycl-be.cpp @@ -1,4 +1,4 @@ -// REQUIRES: gpu, cuda, hip, opencl, sycl-ls +// REQUIRES: gpu, cuda, hip, opencl, esimd_emulator, sycl-ls // RUN: sycl-ls --verbose >%t.default.out // RUN: FileCheck %s --check-prefixes=CHECK-BUILTIN-GPU-OPENCL,CHECK-CUSTOM-GPU-OPENCL --input-file %t.default.out @@ -21,8 +21,13 @@ // CHECK-BUILTIN-GPU-HIP: gpu_selector(){{.*}}gpu, {{.*}}HIP // CHECK-CUSTOM-GPU-HIP: custom_selector(gpu){{.*}}gpu, {{.*}}HIP -//==---- sycl-ls-gpu-sycl-be.cpp - SYCL test for discovered/selected devices -//--==// +// RUN: env SYCL_DEVICE_FILTER=esimd_emulator sycl-ls --verbose >%t.esimd_emulator.out +// RUN: FileCheck %s --check-prefixes=CHECK-BUILTIN-GPU-ESIMD-EMULATOR,CHECK-CUSTOM-GPU-ESIMD-EMULATOR --input-file %t.esimdemulator.out + +// CHECK-BUILTIN-GPU-ESIMD-EMULATOR: gpu_selector(){{.*}}gpu, {{.*}}ESIMD_EMULATOR +// CHECK-CUSTOM-GPU-ESIMD-EMULATOR: custom_selector(gpu){{.*}}gpu, {{.*}}ESIMD_EMULATOR + +//==-- sycl-ls-gpu-sycl-be.cpp - SYCL test for discovered/selected devices -==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. // See https://llvm.org/LICENSE.txt for license information. diff --git a/SYCL/Plugin/sycl-ls.cpp b/SYCL/Plugin/sycl-ls.cpp index eed3f820b4..ae33e83eb9 100644 --- a/SYCL/Plugin/sycl-ls.cpp +++ b/SYCL/Plugin/sycl-ls.cpp @@ -1,5 +1,5 @@ // RUN: sycl-ls --verbose | grep "Device \[" | wc -l >%t.verbose.out -// RUN: sycl-ls | wc -l >%t.concise.out +// RUN: sycl-ls | grep "^\[[a-z0-9_]*:[a-z]*:[0-9]*\]\ " | wc -l >%t.concise.out // RUN: diff %t.verbose.out %t.concise.out //==---- sycl-ls.cpp - SYCL test for consistency of sycl-ls output ---------==// diff --git a/SYCL/README.md b/SYCL/README.md index 8bab44e134..056705de93 100644 --- a/SYCL/README.md +++ b/SYCL/README.md @@ -81,7 +81,7 @@ list of configurations. Each configuration includes backend separated from comma-separated list of target devices with colon. Example: ``` --DCHECK_SYCL_ALL="opencl:cpu,host;level_zero:gpu,host;cuda:gpu;hip:gpu" +-DCHECK_SYCL_ALL="opencl:cpu,host;level_zero:gpu,host;cuda:gpu;hip:gpu,esimd_emulator:gpu" ``` ***SYCL_BE*** SYCL backend to be used for testing. Supported values are: @@ -89,11 +89,12 @@ from comma-separated list of target devices with colon. Example: - **cuda** - for CUDA backend; - **hip** - for HIP backend; - **level_zero** - Level Zero backend. + - **esimd_emulator** - ESIMD Emulator backend. ***SYCL_TARGET_DEVICES*** comma separated list of target devices for testing. Default value is cpu,gpu,acc,host. Supported values are: - **cpu** - CPU device available in OpenCL backend only; - - **gpu** - GPU device available in OpenCL, Level Zero, CUDA, and HIP backends; + - **gpu** - GPU device available in OpenCL, Level Zero, CUDA, HIP, and ESIMD_EMULATOR backends; - **acc** - FPGA emulator device available in OpenCL backend only; - **host** - SYCL Host device available with all backends. @@ -134,7 +135,7 @@ unavailable. * **windows**, **linux** - host OS; * **cpu**, **gpu**, **host**, **accelerator** - target device; - * **cuda**, **hip**, **opencl**, **level_zero** - target backend; + * **cuda**, **hip**, **opencl**, **level_zero**, **esimd_emulator**- target backend; * **sycl-ls** - sycl-ls tool availability; * **cl_options** - CL command line options recognized (or not) by compiler; * **opencl_icd** - OpenCL ICD loader availability; @@ -152,7 +153,7 @@ configure specific single test execution in the command line: * **dpcpp_compiler** - full path to dpcpp compiler; * **target_device** - comma-separated list of target devices (cpu, gpu, acc, host); - * **sycl_be** - SYCL backend to be used (opencl, level_zero, cuda, hip); + * **sycl_be** - SYCL backend to be used (opencl, level_zero, cuda, hip, esimd_emulator); * **dump_ir** - if IR dumping is supported for compiler (True, False); * **gpu-intel-dg1** - tells LIT infra that Intel GPU DG1 is present in the system. It is developer / CI infra responsibility to make sure that the diff --git a/SYCL/Reduction/reduction_big_data.cpp b/SYCL/Reduction/reduction_big_data.cpp index c490d08da2..b4782baf3e 100644 --- a/SYCL/Reduction/reduction_big_data.cpp +++ b/SYCL/Reduction/reduction_big_data.cpp @@ -6,6 +6,7 @@ // Missing __spirv_GroupFMax on AMD, error message `Group algorithms are not // supported on host device` on Nvidia. // XFAIL: hip_amd || hip_nvidia +// UNSUPPORTED: esimd_emulator // RUNx: %HOST_RUN_PLACEHOLDER %t.out // TODO: Enable the test for HOST when it supports ext::oneapi::reduce() and diff --git a/SYCL/Reduction/reduction_nd_conditional.cpp b/SYCL/Reduction/reduction_nd_conditional.cpp index a330c40831..ecc40f254f 100644 --- a/SYCL/Reduction/reduction_nd_conditional.cpp +++ b/SYCL/Reduction/reduction_nd_conditional.cpp @@ -8,6 +8,7 @@ // parallel_for with reduction requires work group size not bigger than 1` on // Nvidia. // XFAIL: hip_amd || hip_nvidia +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(nd_range, reduction, func) // with reduction and conditional increment of the reduction variable. diff --git a/SYCL/Reduction/reduction_nd_lambda.cpp b/SYCL/Reduction/reduction_nd_lambda.cpp index 3e14952233..43908fc58d 100644 --- a/SYCL/Reduction/reduction_nd_lambda.cpp +++ b/SYCL/Reduction/reduction_nd_lambda.cpp @@ -7,6 +7,7 @@ // Inconsistently fails on HIP AMD, error message `Barrier is not supported on // the host device yet.` on HIP Nvidia. // UNSUPPORTED: hip_amd || hip_nvidia +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(nd_range, reduction, lambda) diff --git a/SYCL/Reduction/reduction_nd_s0_dw.cpp b/SYCL/Reduction/reduction_nd_s0_dw.cpp index a53009fc17..9fd39f5b8f 100644 --- a/SYCL/Reduction/reduction_nd_s0_dw.cpp +++ b/SYCL/Reduction/reduction_nd_s0_dw.cpp @@ -9,6 +9,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(nd_range, reduction, func) // with reductions initialized with 0-dimensional discard_write accessor. diff --git a/SYCL/Reduction/reduction_nd_s0_rw.cpp b/SYCL/Reduction/reduction_nd_s0_rw.cpp index 5c13fcb1cc..0665e58a00 100644 --- a/SYCL/Reduction/reduction_nd_s0_rw.cpp +++ b/SYCL/Reduction/reduction_nd_s0_rw.cpp @@ -9,6 +9,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(nd_range, reduction, func) // with reductions initialized with 0-dimensional read_write accessor. diff --git a/SYCL/Reduction/reduction_nd_s1_dw.cpp b/SYCL/Reduction/reduction_nd_s1_dw.cpp index 0cb0ac04fb..7bb2bc3c6a 100644 --- a/SYCL/Reduction/reduction_nd_s1_dw.cpp +++ b/SYCL/Reduction/reduction_nd_s1_dw.cpp @@ -10,6 +10,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(nd_range, reduction, func) // with reductions initialized with 1-dimensional discard_write accessor diff --git a/SYCL/Reduction/reduction_nd_s1_rw.cpp b/SYCL/Reduction/reduction_nd_s1_rw.cpp index 4d476c0027..9ed508f670 100644 --- a/SYCL/Reduction/reduction_nd_s1_rw.cpp +++ b/SYCL/Reduction/reduction_nd_s1_rw.cpp @@ -9,6 +9,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(nd_range, reduction, func) // with reductions initialized with 1-dimensional read_write accessor diff --git a/SYCL/Reduction/reduction_placeholder.cpp b/SYCL/Reduction/reduction_placeholder.cpp index 2324a4df03..ce57a0dab3 100644 --- a/SYCL/Reduction/reduction_placeholder.cpp +++ b/SYCL/Reduction/reduction_placeholder.cpp @@ -6,6 +6,7 @@ // Missing __spirv_GroupIAdd, __spirv_GroupFMin on AMD, error message `Group // algorithms are not supported on host device.` on Nvidia. // XFAIL: hip_amd || hip_nvidia +// UNSUPPORTED: esimd_emulator // RUNx: %HOST_RUN_PLACEHOLDER %t.out // TODO: Enable the test for HOST when it supports ext::oneapi::reduce() and diff --git a/SYCL/Reduction/reduction_queue_parallel_for.cpp b/SYCL/Reduction/reduction_queue_parallel_for.cpp index 6fa5484ee7..42999c2d90 100644 --- a/SYCL/Reduction/reduction_queue_parallel_for.cpp +++ b/SYCL/Reduction/reduction_queue_parallel_for.cpp @@ -9,6 +9,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // RUNx: %HOST_RUN_PLACEHOLDER %t.out // TODO: Enable the test for HOST when it supports ext::oneapi::reduce() and diff --git a/SYCL/Reduction/reduction_range_1d_s0_dw.cpp b/SYCL/Reduction/reduction_range_1d_s0_dw.cpp index d62c30e8b3..6332ed6da8 100644 --- a/SYCL/Reduction/reduction_range_1d_s0_dw.cpp +++ b/SYCL/Reduction/reduction_range_1d_s0_dw.cpp @@ -8,6 +8,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(range<1>, reduction, func) // with reductions initialized with 0-dimensional discard_write accessor diff --git a/SYCL/Reduction/reduction_range_1d_s0_rw.cpp b/SYCL/Reduction/reduction_range_1d_s0_rw.cpp index 74317e6d41..de2b5f2518 100644 --- a/SYCL/Reduction/reduction_range_1d_s0_rw.cpp +++ b/SYCL/Reduction/reduction_range_1d_s0_rw.cpp @@ -8,6 +8,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(range<1>, reduction, func) // with reductions initialized with 0-dimensional read_write accessor diff --git a/SYCL/Reduction/reduction_range_1d_s1_dw.cpp b/SYCL/Reduction/reduction_range_1d_s1_dw.cpp index 1ef65eef16..1d69fc5a45 100644 --- a/SYCL/Reduction/reduction_range_1d_s1_dw.cpp +++ b/SYCL/Reduction/reduction_range_1d_s1_dw.cpp @@ -5,6 +5,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator #include "reduction_range_scalar.hpp" diff --git a/SYCL/Reduction/reduction_range_1d_s1_rw.cpp b/SYCL/Reduction/reduction_range_1d_s1_rw.cpp index 359aa2f0fe..54e39a2c75 100644 --- a/SYCL/Reduction/reduction_range_1d_s1_rw.cpp +++ b/SYCL/Reduction/reduction_range_1d_s1_rw.cpp @@ -5,6 +5,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(range<1>, reduction, func) // with reductions initialized with 1-dimensional read_write accessor diff --git a/SYCL/Reduction/reduction_range_2d_s1_dw.cpp b/SYCL/Reduction/reduction_range_2d_s1_dw.cpp index 5a04e5773b..21caf8c571 100644 --- a/SYCL/Reduction/reduction_range_2d_s1_dw.cpp +++ b/SYCL/Reduction/reduction_range_2d_s1_dw.cpp @@ -8,6 +8,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(range<2>, reduction, func) // with reductions initialized with 1-dimensional discard_write accessor diff --git a/SYCL/Reduction/reduction_range_2d_s1_rw.cpp b/SYCL/Reduction/reduction_range_2d_s1_rw.cpp index b2fb2ba14d..c26e72e968 100644 --- a/SYCL/Reduction/reduction_range_2d_s1_rw.cpp +++ b/SYCL/Reduction/reduction_range_2d_s1_rw.cpp @@ -8,6 +8,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(range<2>, reduction, func) // with reductions initialized with 1-dimensional discard_write accessor diff --git a/SYCL/Reduction/reduction_range_3d_s1_dw.cpp b/SYCL/Reduction/reduction_range_3d_s1_dw.cpp index c4b7a4ab6e..9cef9ed89c 100644 --- a/SYCL/Reduction/reduction_range_3d_s1_dw.cpp +++ b/SYCL/Reduction/reduction_range_3d_s1_dw.cpp @@ -8,6 +8,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(range<3>, reduction, func) // with reductions initialized with 1-dimensional discard_write accessor diff --git a/SYCL/Reduction/reduction_range_3d_s1_rw.cpp b/SYCL/Reduction/reduction_range_3d_s1_rw.cpp index 79bc4eed55..aa7594dfb8 100644 --- a/SYCL/Reduction/reduction_range_3d_s1_rw.cpp +++ b/SYCL/Reduction/reduction_range_3d_s1_rw.cpp @@ -8,6 +8,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(range<3>, reduction, func) // with reductions initialized with 1-dimensional discard_write accessor diff --git a/SYCL/Reduction/reduction_range_lambda.cpp b/SYCL/Reduction/reduction_range_lambda.cpp index 67a8218a70..436acd931c 100644 --- a/SYCL/Reduction/reduction_range_lambda.cpp +++ b/SYCL/Reduction/reduction_range_lambda.cpp @@ -2,6 +2,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(range, reduction, lambda) // with reductions initialized with 1-dimensional accessor accessing diff --git a/SYCL/Reduction/reduction_range_usm_dw.cpp b/SYCL/Reduction/reduction_range_usm_dw.cpp index e50626464e..7ff4cc6bfd 100644 --- a/SYCL/Reduction/reduction_range_usm_dw.cpp +++ b/SYCL/Reduction/reduction_range_usm_dw.cpp @@ -9,6 +9,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator #include "reduction_range_scalar.hpp" diff --git a/SYCL/Reduction/reduction_reducer_op_eq.cpp b/SYCL/Reduction/reduction_reducer_op_eq.cpp index db04c2bcd0..52bcb4719b 100644 --- a/SYCL/Reduction/reduction_reducer_op_eq.cpp +++ b/SYCL/Reduction/reduction_reducer_op_eq.cpp @@ -5,6 +5,7 @@ // // On nvidia a reduction appears to be unexpectedly executed via the host. // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator // This test checks that operators ++, +=, *=, |=, &=, ^= are supported // whent the corresponding std::plus<>, std::multiplies, etc are defined. diff --git a/SYCL/Reduction/reduction_usm.cpp b/SYCL/Reduction/reduction_usm.cpp index a381140b9b..8ede462877 100644 --- a/SYCL/Reduction/reduction_usm.cpp +++ b/SYCL/Reduction/reduction_usm.cpp @@ -9,6 +9,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // This test performs basic checks of parallel_for(nd_range, reduction, func) // with reductions initialized with USM var. It tests both diff --git a/SYCL/Reduction/reduction_usm_dw.cpp b/SYCL/Reduction/reduction_usm_dw.cpp index 3a7ebaf309..545436ebf3 100644 --- a/SYCL/Reduction/reduction_usm_dw.cpp +++ b/SYCL/Reduction/reduction_usm_dw.cpp @@ -5,6 +5,7 @@ // TODO: test disabled due to sporadic fails in level_zero:gpu RT. // UNSUPPORTED: linux && level_zero +// UNSUPPORTED: esimd_emulator // // Missing __spirv_GroupIAdd, __spirv_GroupFMin, __spirv_GroupFMax on AMD, error // message `Group algorithms are not supported on host device` on Nvidia. diff --git a/SYCL/Regression/atomic_load.cpp b/SYCL/Regression/atomic_load.cpp index a8afef2c86..8d9467c121 100644 --- a/SYCL/Regression/atomic_load.cpp +++ b/SYCL/Regression/atomic_load.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include using namespace cl::sycl; diff --git a/SYCL/Regression/commandlist/gpu.cpp b/SYCL/Regression/commandlist/gpu.cpp index c9d093c66d..1d04ba27b6 100644 --- a/SYCL/Regression/commandlist/gpu.cpp +++ b/SYCL/Regression/commandlist/gpu.cpp @@ -1,5 +1,6 @@ // REQUIRES: gpu, linux // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // // RUN: %clangxx -fsycl %S/Inputs/FindPrimesSYCL.cpp %S/Inputs/main.cpp -o %t.out -lpthread // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Regression/device_num.cpp b/SYCL/Regression/device_num.cpp index 3ff8f1afd8..7cf1d8a495 100644 --- a/SYCL/Regression/device_num.cpp +++ b/SYCL/Regression/device_num.cpp @@ -8,6 +8,7 @@ // The test is using all available BEs but CUDA machine in CI does not have // functional OpenCL RT // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Regression/group.cpp b/SYCL/Regression/group.cpp index 3ada0d637e..d7f925d3ba 100644 --- a/SYCL/Regression/group.cpp +++ b/SYCL/Regression/group.cpp @@ -6,6 +6,7 @@ // // Crashes on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator //==-- group.cpp - Regression tests for cl::sycl::group API bug fixes. -----==// // diff --git a/SYCL/Regression/host_unified_memory.cpp b/SYCL/Regression/host_unified_memory.cpp index 586149b9c2..4ec23e185c 100644 --- a/SYCL/Regression/host_unified_memory.cpp +++ b/SYCL/Regression/host_unified_memory.cpp @@ -1,6 +1,7 @@ // REQUIRES: gpu && linux // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %GPU_RUN_PLACEHOLDER SYCL_HOST_UNIFIED_MEMORY=1 %t.out +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Regression/image_access.cpp b/SYCL/Regression/image_access.cpp index 90fbde3cb1..4ebe2d47d5 100644 --- a/SYCL/Regression/image_access.cpp +++ b/SYCL/Regression/image_access.cpp @@ -5,6 +5,7 @@ // No execution of FPGA because it does not support images // // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // CUDA cannot support OpenCL spec conform images. //==-------------- image_access.cpp - SYCL image accessors test -----------==// diff --git a/SYCL/Regression/implicit_atomic_conversion.cpp b/SYCL/Regression/implicit_atomic_conversion.cpp index 75ff63807f..38624c1271 100644 --- a/SYCL/Regression/implicit_atomic_conversion.cpp +++ b/SYCL/Regression/implicit_atomic_conversion.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Regression/kernel_bundle_ignore_sycl_external.cpp b/SYCL/Regression/kernel_bundle_ignore_sycl_external.cpp index 672f30e5f2..ef3c9d1a51 100644 --- a/SYCL/Regression/kernel_bundle_ignore_sycl_external.cpp +++ b/SYCL/Regression/kernel_bundle_ignore_sycl_external.cpp @@ -4,6 +4,7 @@ // RUN: %ACC_RUN_PLACEHOLDER %t.out // // XFAIL: cuda || hip_nvidia +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Regression/kernel_name_class.cpp b/SYCL/Regression/kernel_name_class.cpp index bfb5065dd6..da992ff47b 100644 --- a/SYCL/Regression/kernel_name_class.cpp +++ b/SYCL/Regression/kernel_name_class.cpp @@ -8,6 +8,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.ext.out // RUN: %GPU_RUN_PLACEHOLDER %t.ext.out // RUN: %ACC_RUN_PLACEHOLDER %t.ext.out +// UNSUPPORTED: esimd_emulator //==-- kernel_name_class.cpp - SYCL kernel naming variants test ------------==// // diff --git a/SYCL/Regression/kernel_name_inside_sycl_namespace.cpp b/SYCL/Regression/kernel_name_inside_sycl_namespace.cpp index c47acf6792..6748b1c8dc 100644 --- a/SYCL/Regression/kernel_name_inside_sycl_namespace.cpp +++ b/SYCL/Regression/kernel_name_inside_sycl_namespace.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==------- kernel_name_inside_sycl_namespace.cpp - Regression test --------==// // diff --git a/SYCL/Regression/kernel_unnamed.cpp b/SYCL/Regression/kernel_unnamed.cpp index 76e6a120fd..364c42be9e 100644 --- a/SYCL/Regression/kernel_unnamed.cpp +++ b/SYCL/Regression/kernel_unnamed.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==-- kernel_unnamed.cpp - SYCL kernel naming variants test ------------==// // diff --git a/SYCL/Regression/nontrivial_device_copyable_value.cpp b/SYCL/Regression/nontrivial_device_copyable_value.cpp index 73d9ac49f3..67f27add1f 100644 --- a/SYCL/Regression/nontrivial_device_copyable_value.cpp +++ b/SYCL/Regression/nontrivial_device_copyable_value.cpp @@ -3,6 +3,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // // Note: Tests that non-trivially copyable types marked as device-copyable are // copied and used correctly on the device. diff --git a/SYCL/Regression/private_array_init_test.cpp b/SYCL/Regression/private_array_init_test.cpp index 58ad352069..9676ef4330 100644 --- a/SYCL/Regression/private_array_init_test.cpp +++ b/SYCL/Regression/private_array_init_test.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==- private_array_init_test.cpp - Regression test for private array init -==// // diff --git a/SYCL/Regression/static-buffer-dtor.cpp b/SYCL/Regression/static-buffer-dtor.cpp index 0b38b33895..cfb6445856 100644 --- a/SYCL/Regression/static-buffer-dtor.cpp +++ b/SYCL/Regression/static-buffer-dtor.cpp @@ -17,6 +17,7 @@ // Failing on HIP AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/Regression/unoptimized_stream.cpp b/SYCL/Regression/unoptimized_stream.cpp index fae5af2bcb..88f2ef7399 100644 --- a/SYCL/Regression/unoptimized_stream.cpp +++ b/SYCL/Regression/unoptimized_stream.cpp @@ -4,6 +4,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: cuda +// UNSUPPORTED: esimd_emulator // Disable test due to flaky failure on CUDA(issue #387) // NOTE: The libclc target used by the CUDA backend used to generate atomic load diff --git a/SYCL/Sampler/basic-rw-float.cpp b/SYCL/Sampler/basic-rw-float.cpp index a4ee73de08..fdca69f5ad 100644 --- a/SYCL/Sampler/basic-rw-float.cpp +++ b/SYCL/Sampler/basic-rw-float.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/basic-rw.cpp b/SYCL/Sampler/basic-rw.cpp index 29fa88dee0..9a474467e0 100644 --- a/SYCL/Sampler/basic-rw.cpp +++ b/SYCL/Sampler/basic-rw.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/normalized-clamp-linear-float.cpp b/SYCL/Sampler/normalized-clamp-linear-float.cpp index 506793aaf5..a78f1ca3f2 100644 --- a/SYCL/Sampler/normalized-clamp-linear-float.cpp +++ b/SYCL/Sampler/normalized-clamp-linear-float.cpp @@ -5,6 +5,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // UNSUPPORTED: level_zero && windows +// UNSUPPORTED: esimd_emulator // XFAIL: cuda // LevelZero on Windows hangs with normalized coordinates. Waiting on fix. diff --git a/SYCL/Sampler/normalized-clamp-nearest.cpp b/SYCL/Sampler/normalized-clamp-nearest.cpp index 36e5fa5b60..7d2a43b2f1 100644 --- a/SYCL/Sampler/normalized-clamp-nearest.cpp +++ b/SYCL/Sampler/normalized-clamp-nearest.cpp @@ -7,6 +7,7 @@ // TODO: enable this test after flaky bug is gone on Windows // UNSUPPORTED: windows +// UNSUPPORTED: esimd_emulator // LevelZero has a bug wherein it always returns the first pixel value. // Will re-enable once fixed. diff --git a/SYCL/Sampler/normalized-clampedge-linear-float.cpp b/SYCL/Sampler/normalized-clampedge-linear-float.cpp index f3fde942c7..1a2f516c9d 100644 --- a/SYCL/Sampler/normalized-clampedge-linear-float.cpp +++ b/SYCL/Sampler/normalized-clampedge-linear-float.cpp @@ -3,6 +3,7 @@ // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // XFAIL: cuda || hip +// UNSUPPORTED: esimd_emulator // CUDA works with image_channel_type::fp32, but not with any 8-bit per channel // type (such as unorm_int8) diff --git a/SYCL/Sampler/normalized-clampedge-nearest.cpp b/SYCL/Sampler/normalized-clampedge-nearest.cpp index f0a7c516e4..3a05efe3e1 100644 --- a/SYCL/Sampler/normalized-clampedge-nearest.cpp +++ b/SYCL/Sampler/normalized-clampedge-nearest.cpp @@ -6,6 +6,7 @@ // Missing __spirv_ImageWrite, __spirv_SampledImage, // __spirv_ImageSampleExplicitLod on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator /* This file sets up an image, initializes it with data, diff --git a/SYCL/Sampler/normalized-mirror-linear-float.cpp b/SYCL/Sampler/normalized-mirror-linear-float.cpp index ef168082e9..3fb8b22369 100644 --- a/SYCL/Sampler/normalized-mirror-linear-float.cpp +++ b/SYCL/Sampler/normalized-mirror-linear-float.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/normalized-mirror-nearest.cpp b/SYCL/Sampler/normalized-mirror-nearest.cpp index 84c9ad9c96..8a9bc26399 100644 --- a/SYCL/Sampler/normalized-mirror-nearest.cpp +++ b/SYCL/Sampler/normalized-mirror-nearest.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/normalized-none-linear-float.cpp b/SYCL/Sampler/normalized-none-linear-float.cpp index b7b8e5b04a..2aae8d2768 100644 --- a/SYCL/Sampler/normalized-none-linear-float.cpp +++ b/SYCL/Sampler/normalized-none-linear-float.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/normalized-none-nearest.cpp b/SYCL/Sampler/normalized-none-nearest.cpp index 6f17e6cd25..52559f4503 100644 --- a/SYCL/Sampler/normalized-none-nearest.cpp +++ b/SYCL/Sampler/normalized-none-nearest.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/normalized-repeat-linear-float.cpp b/SYCL/Sampler/normalized-repeat-linear-float.cpp index f66cf6e862..0e98cbb9dd 100644 --- a/SYCL/Sampler/normalized-repeat-linear-float.cpp +++ b/SYCL/Sampler/normalized-repeat-linear-float.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/normalized-repeat-nearest.cpp b/SYCL/Sampler/normalized-repeat-nearest.cpp index efa7395dd4..884f618ec9 100644 --- a/SYCL/Sampler/normalized-repeat-nearest.cpp +++ b/SYCL/Sampler/normalized-repeat-nearest.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/unnormalized-clamp-linear-float.cpp b/SYCL/Sampler/unnormalized-clamp-linear-float.cpp index f5f4bdd1b1..2e36b9036c 100644 --- a/SYCL/Sampler/unnormalized-clamp-linear-float.cpp +++ b/SYCL/Sampler/unnormalized-clamp-linear-float.cpp @@ -6,6 +6,7 @@ // XFAIL: cuda // Temporary disabled (#204) // UNSUPPORTED: level_zero && windows +// UNSUPPORTED: esimd_emulator // CUDA works with image_channel_type::fp32, but not with any 8-bit per channel // type (such as unorm_int8) diff --git a/SYCL/Sampler/unnormalized-clamp-nearest.cpp b/SYCL/Sampler/unnormalized-clamp-nearest.cpp index 4006fa6d34..b192f0f7f0 100644 --- a/SYCL/Sampler/unnormalized-clamp-nearest.cpp +++ b/SYCL/Sampler/unnormalized-clamp-nearest.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/unnormalized-clampedge-linear-float.cpp b/SYCL/Sampler/unnormalized-clampedge-linear-float.cpp index 296302e1d7..35a3afdf33 100644 --- a/SYCL/Sampler/unnormalized-clampedge-linear-float.cpp +++ b/SYCL/Sampler/unnormalized-clampedge-linear-float.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/unnormalized-clampedge-nearest.cpp b/SYCL/Sampler/unnormalized-clampedge-nearest.cpp index dc1fb2a06c..61953084c0 100644 --- a/SYCL/Sampler/unnormalized-clampedge-nearest.cpp +++ b/SYCL/Sampler/unnormalized-clampedge-nearest.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/unnormalized-none-linear-float.cpp b/SYCL/Sampler/unnormalized-none-linear-float.cpp index 5a7087b320..1892b767d6 100644 --- a/SYCL/Sampler/unnormalized-none-linear-float.cpp +++ b/SYCL/Sampler/unnormalized-none-linear-float.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Sampler/unnormalized-none-nearest.cpp b/SYCL/Sampler/unnormalized-none-nearest.cpp index fadefde39e..8bd5045289 100644 --- a/SYCL/Sampler/unnormalized-none-nearest.cpp +++ b/SYCL/Sampler/unnormalized-none-nearest.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Scheduler/BasicSchedulerTests.cpp b/SYCL/Scheduler/BasicSchedulerTests.cpp index d0fb040629..e7b517b79f 100644 --- a/SYCL/Scheduler/BasicSchedulerTests.cpp +++ b/SYCL/Scheduler/BasicSchedulerTests.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==------------------- BasicSchedulerTests.cpp ----------------------------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. diff --git a/SYCL/Scheduler/CommandCleanupThreadSafety.cpp b/SYCL/Scheduler/CommandCleanupThreadSafety.cpp index d801c071df..0af2d3e603 100644 --- a/SYCL/Scheduler/CommandCleanupThreadSafety.cpp +++ b/SYCL/Scheduler/CommandCleanupThreadSafety.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: windows +// UNSUPPORTED: esimd_emulator // RUN: %clangxx -fsycl %s -o %t.out -lpthread // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Scheduler/DataMovement.cpp b/SYCL/Scheduler/DataMovement.cpp index 07acfbb059..a94aa4bb02 100644 --- a/SYCL/Scheduler/DataMovement.cpp +++ b/SYCL/Scheduler/DataMovement.cpp @@ -2,6 +2,7 @@ // RUN: %HOST_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // //==-------------------------- DataMovement.cpp ----------------------------==// // diff --git a/SYCL/Scheduler/InOrderQueueDeps.cpp b/SYCL/Scheduler/InOrderQueueDeps.cpp index 104b1a13d9..882c934418 100644 --- a/SYCL/Scheduler/InOrderQueueDeps.cpp +++ b/SYCL/Scheduler/InOrderQueueDeps.cpp @@ -9,6 +9,7 @@ // The tested functionality is disabled with Level Zero until it is supported by // the plugin. // UNSUPPORTED: level_zero +// UNSUPPORTED: esimd_emulator //==----------------------- InOrderQueueDeps.cpp ---------------------------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. diff --git a/SYCL/Scheduler/MemObjRemapping.cpp b/SYCL/Scheduler/MemObjRemapping.cpp index 92f675d53b..a15e5252ee 100644 --- a/SYCL/Scheduler/MemObjRemapping.cpp +++ b/SYCL/Scheduler/MemObjRemapping.cpp @@ -3,6 +3,7 @@ // RUN: env SYCL_HOST_UNIFIED_MEMORY=1 SYCL_PI_TRACE=2 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator #include #include #include diff --git a/SYCL/Scheduler/MultipleDevices.cpp b/SYCL/Scheduler/MultipleDevices.cpp index 000d30676c..e0ea58b384 100644 --- a/SYCL/Scheduler/MultipleDevices.cpp +++ b/SYCL/Scheduler/MultipleDevices.cpp @@ -2,6 +2,7 @@ // RUN: env SYCL_DEVICE_FILTER=%sycl_be %t.out // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator //===- MultipleDevices.cpp - Test checking multi-device execution --------===// // diff --git a/SYCL/Scheduler/ReleaseResourcesTest.cpp b/SYCL/Scheduler/ReleaseResourcesTest.cpp index 8a51878677..f23260742c 100644 --- a/SYCL/Scheduler/ReleaseResourcesTest.cpp +++ b/SYCL/Scheduler/ReleaseResourcesTest.cpp @@ -5,6 +5,7 @@ // RUN: env SYCL_PI_TRACE=2 %ACC_RUN_PLACEHOLDER %t.out 2>&1 %ACC_CHECK_PLACEHOLDER // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator //==------------------- ReleaseResourcesTests.cpp --------------------------==// // diff --git a/SYCL/SeparateCompile/same-kernel.cpp b/SYCL/SeparateCompile/same-kernel.cpp index 42407276a3..f363a54ee0 100644 --- a/SYCL/SeparateCompile/same-kernel.cpp +++ b/SYCL/SeparateCompile/same-kernel.cpp @@ -16,6 +16,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t-same-kernel.exe // RUN: %GPU_RUN_PLACEHOLDER %t-same-kernel.exe // RUN: %ACC_RUN_PLACEHOLDER %t-same-kernel.exe +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/SeparateCompile/sycl-external.cpp b/SYCL/SeparateCompile/sycl-external.cpp index b988f1af71..df70d7319d 100644 --- a/SYCL/SeparateCompile/sycl-external.cpp +++ b/SYCL/SeparateCompile/sycl-external.cpp @@ -18,6 +18,7 @@ // // Linking issues with HIP AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/SeparateCompile/test.cpp b/SYCL/SeparateCompile/test.cpp index a7ec38194a..276c8f3f51 100644 --- a/SYCL/SeparateCompile/test.cpp +++ b/SYCL/SeparateCompile/test.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator // CUDA and HIP don't support SPIR-V. // // FIXME Disabled fallback assert as it'll require either online linking or diff --git a/SYCL/SpecConstants/2020/handler-api.cpp b/SYCL/SpecConstants/2020/handler-api.cpp index 4ba012b29e..2bb398bc0e 100644 --- a/SYCL/SpecConstants/2020/handler-api.cpp +++ b/SYCL/SpecConstants/2020/handler-api.cpp @@ -14,6 +14,7 @@ // FIXME: ACC devices use emulation path, which is not yet supported // FIXME: CUDA uses emulation path, which is not yet supported // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/SpecConstants/2020/host_apis.cpp b/SYCL/SpecConstants/2020/host_apis.cpp index c84276b489..c4b3ea887b 100644 --- a/SYCL/SpecConstants/2020/host_apis.cpp +++ b/SYCL/SpecConstants/2020/host_apis.cpp @@ -3,6 +3,7 @@ // UNSUPPORTED: cuda // UNSUPPORTED: hip +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/SpecConstants/2020/kernel-bundle-api.cpp b/SYCL/SpecConstants/2020/kernel-bundle-api.cpp index 82022bfcbb..7c1127d1a4 100644 --- a/SYCL/SpecConstants/2020/kernel-bundle-api.cpp +++ b/SYCL/SpecConstants/2020/kernel-bundle-api.cpp @@ -14,6 +14,7 @@ // FIXME: ACC devices use emulation path, which is not yet supported // FIXME: CUDA uses emulation path, which is not yet supported // UNSUPPORTED: cuda || hip +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/SubGroup/barrier.cpp b/SYCL/SubGroup/barrier.cpp index d973e587dd..87fc71c912 100644 --- a/SYCL/SubGroup/barrier.cpp +++ b/SYCL/SubGroup/barrier.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator //==---------- barrier.cpp - SYCL sub_group barrier test -------*- C++ -*---==// // diff --git a/SYCL/SubGroup/broadcast.cpp b/SYCL/SubGroup/broadcast.cpp index 9f1d27964e..52331e37d7 100644 --- a/SYCL/SubGroup/broadcast.cpp +++ b/SYCL/SubGroup/broadcast.cpp @@ -6,6 +6,7 @@ // // Missing __spirv_GroupBroadcast on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator //==--------- broadcast.cpp - SYCL sub_group broadcast test ----*- C++ -*---==// // diff --git a/SYCL/SubGroup/broadcast_fp16.cpp b/SYCL/SubGroup/broadcast_fp16.cpp index c9a136a600..93373489bd 100644 --- a/SYCL/SubGroup/broadcast_fp16.cpp +++ b/SYCL/SubGroup/broadcast_fp16.cpp @@ -3,6 +3,7 @@ // // Missing __spirv_GroupBroadcast on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator //==---- broadcast_fp16.cpp - SYCL sub_group broadcast test ----*- C++ -*---==// // diff --git a/SYCL/SubGroup/broadcast_fp64.cpp b/SYCL/SubGroup/broadcast_fp64.cpp index 3be5266acb..ddee3e5af5 100644 --- a/SYCL/SubGroup/broadcast_fp64.cpp +++ b/SYCL/SubGroup/broadcast_fp64.cpp @@ -6,6 +6,7 @@ // // Missing __spirv_GroupBroadcast on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator //==----- broadcast_fp64.cpp - SYCL sub_group broadcast test ----*- C++ -*--==// // diff --git a/SYCL/SubGroup/common.cpp b/SYCL/SubGroup/common.cpp index 2a8730e97c..ddc3f117e5 100644 --- a/SYCL/SubGroup/common.cpp +++ b/SYCL/SubGroup/common.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // //==-------------- common.cpp - SYCL sub_group common test -----*- C++ -*---==// // diff --git a/SYCL/SubGroup/generic-shuffle.cpp b/SYCL/SubGroup/generic-shuffle.cpp index 3ec9ef31bd..cc84d0bc22 100644 --- a/SYCL/SubGroup/generic-shuffle.cpp +++ b/SYCL/SubGroup/generic-shuffle.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // // Missing __spirv_SubgroupShuffleINTEL, __spirv_SubgroupShuffleUpINTEL, // __spirv_SubgroupShuffleDownINTEL, __spirv_SubgroupShuffleXorINTEL on AMD diff --git a/SYCL/SubGroup/shuffle.cpp b/SYCL/SubGroup/shuffle.cpp index b351f4635f..b17da3ab4d 100644 --- a/SYCL/SubGroup/shuffle.cpp +++ b/SYCL/SubGroup/shuffle.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // // Missing __spirv_SubgroupShuffleINTEL, __spirv_SubgroupShuffleUpINTEL, // __spirv_SubgroupShuffleDownINTEL, __spirv_SubgroupShuffleXorINTEL on AMD diff --git a/SYCL/SubGroup/shuffle_fp16.cpp b/SYCL/SubGroup/shuffle_fp16.cpp index 93da06fa69..6925116b52 100644 --- a/SYCL/SubGroup/shuffle_fp16.cpp +++ b/SYCL/SubGroup/shuffle_fp16.cpp @@ -4,6 +4,7 @@ // Missing __spirv_SubgroupShuffleINTEL, __spirv_SubgroupShuffleUpINTEL, // __spirv_SubgroupShuffleDownINTEL, __spirv_SubgroupShuffleXorINTEL on AMD // XFAIL: hip_amd +// UNSUPPORTED: esimd_emulator // Even though `gfx908` and `gfx906` support halfs, libspirv is currently // built with `tahiti` as the target CPU, which means that clang rejects // AMD built-ins using halfs, for that reason half support has to stay diff --git a/SYCL/SubGroup/shuffle_fp64.cpp b/SYCL/SubGroup/shuffle_fp64.cpp index cb863cf993..4e465491dd 100644 --- a/SYCL/SubGroup/shuffle_fp64.cpp +++ b/SYCL/SubGroup/shuffle_fp64.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator // // Missing __spirv_SubgroupShuffleINTEL, __spirv_SubgroupShuffleUpINTEL, // __spirv_SubgroupShuffleDownINTEL, __spirv_SubgroupShuffleXorINTEL on AMD diff --git a/SYCL/SubGroup/sub_group_as.cpp b/SYCL/SubGroup/sub_group_as.cpp index 2e26abdce5..e4af5fb322 100644 --- a/SYCL/SubGroup/sub_group_as.cpp +++ b/SYCL/SubGroup/sub_group_as.cpp @@ -10,6 +10,7 @@ // __spirv_SubgroupBlockWriteINTEL on AMD // error message `Barrier is not supported on the host device yet.` on Nvidia. // XFAIL: hip_amd || hip_nvidia +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/SubGroup/sub_group_as_vec.cpp b/SYCL/SubGroup/sub_group_as_vec.cpp index 592c14d138..65c050bc34 100644 --- a/SYCL/SubGroup/sub_group_as_vec.cpp +++ b/SYCL/SubGroup/sub_group_as_vec.cpp @@ -10,6 +10,7 @@ // __spirv_SubgroupBlockWriteINTEL on AMD // error message `Barrier is not supported on the host device yet.` on Nvidia. // XFAIL: hip_amd || hip_nvidia +// UNSUPPORTED: esimd_emulator #include "helper.hpp" #include diff --git a/SYCL/SubGroup/sub_groups_sycl2020.cpp b/SYCL/SubGroup/sub_groups_sycl2020.cpp index 4047af0fa6..40b1668c72 100644 --- a/SYCL/SubGroup/sub_groups_sycl2020.cpp +++ b/SYCL/SubGroup/sub_groups_sycl2020.cpp @@ -5,6 +5,7 @@ // Assertion `!MHostPlatform && "Plugin is not available for Host."' failed on // Nvidia. // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/SubGroupMask/Basic.cpp b/SYCL/SubGroupMask/Basic.cpp index cffdd7ee25..660be9c1fd 100644 --- a/SYCL/SubGroupMask/Basic.cpp +++ b/SYCL/SubGroupMask/Basic.cpp @@ -1,6 +1,7 @@ // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // REQUIRES: gpu // UNSUPPORTED: cuda, hip +// UNSUPPORTED: esimd_emulator // GroupNonUniformBallot capability is supported on Intel GPU only // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/SubGroupMask/GroupSize.cpp b/SYCL/SubGroupMask/GroupSize.cpp index 3f70fe58cb..bb93ab04a4 100644 --- a/SYCL/SubGroupMask/GroupSize.cpp +++ b/SYCL/SubGroupMask/GroupSize.cpp @@ -1,6 +1,7 @@ // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // REQUIRES: gpu // UNSUPPORTED: cuda, hip +// UNSUPPORTED: esimd_emulator // GroupNonUniformBallot capability is supported on Intel GPU only // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/SYCL/Tracing/buffer_printers.cpp b/SYCL/Tracing/buffer_printers.cpp index 7335fc395c..c2cbb2125c 100644 --- a/SYCL/Tracing/buffer_printers.cpp +++ b/SYCL/Tracing/buffer_printers.cpp @@ -4,6 +4,7 @@ // RUN: env SYCL_PI_TRACE=2 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator #include #include diff --git a/SYCL/Tracing/image_printers.cpp b/SYCL/Tracing/image_printers.cpp index a923b42c67..7d49010ee9 100644 --- a/SYCL/Tracing/image_printers.cpp +++ b/SYCL/Tracing/image_printers.cpp @@ -4,6 +4,7 @@ // // Unsupported hip call on AMD // UNSUPPORTED: hip_amd +// UNSUPPORTED: esimd_emulator // Test image-specific printers of the Plugin Interace // diff --git a/SYCL/Tracing/pi_tracing_test.cpp b/SYCL/Tracing/pi_tracing_test.cpp index 1ddd4933b2..f51e33af5a 100644 --- a/SYCL/Tracing/pi_tracing_test.cpp +++ b/SYCL/Tracing/pi_tracing_test.cpp @@ -4,6 +4,7 @@ // RUN: env SYCL_PI_TRACE=-1 %ACC_RUN_PLACEHOLDER %t.out %ACC_CHECK_PLACEHOLDER // // XFAIL: hip_nvidia +// UNSUPPORTED: esimd_emulator // Test tracing of the Plugin Interface diff --git a/SYCL/USM/allocator_container.cpp b/SYCL/USM/allocator_container.cpp index d8055a96a2..d24a6d656d 100644 --- a/SYCL/USM/allocator_container.cpp +++ b/SYCL/USM/allocator_container.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==------ allocator_container.cpp - USM allocator in containers test ------==// // diff --git a/SYCL/USM/allocator_shared.cpp b/SYCL/USM/allocator_shared.cpp index ade1ff39ba..271450ed51 100644 --- a/SYCL/USM/allocator_shared.cpp +++ b/SYCL/USM/allocator_shared.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==-------- allocator_shared.cpp - Allocate Shared test -------------------==// // diff --git a/SYCL/USM/allocator_vector.cpp b/SYCL/USM/allocator_vector.cpp index 118b9db07b..c0ef962754 100644 --- a/SYCL/USM/allocator_vector.cpp +++ b/SYCL/USM/allocator_vector.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator // //==---- allocator_vector.cpp - Allocator Container test -------------------==// // diff --git a/SYCL/USM/badmalloc.cpp b/SYCL/USM/badmalloc.cpp index 95934477c1..bf86ad5071 100644 --- a/SYCL/USM/badmalloc.cpp +++ b/SYCL/USM/badmalloc.cpp @@ -1,4 +1,5 @@ // UNSUPPORTED: windows +// UNSUPPORTED: esimd_emulator // // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out // RUN: %HOST_RUN_PLACEHOLDER %t1.out diff --git a/SYCL/USM/copy.cpp b/SYCL/USM/copy.cpp index a21b40a64c..52a8ab0a66 100644 --- a/SYCL/USM/copy.cpp +++ b/SYCL/USM/copy.cpp @@ -11,6 +11,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/USM/dep_events.cpp b/SYCL/USM/dep_events.cpp index 71bab5d7ef..feb6a39e07 100644 --- a/SYCL/USM/dep_events.cpp +++ b/SYCL/USM/dep_events.cpp @@ -9,6 +9,7 @@ // cuMemPrefetchAsync returns cudaErrorInvalidDevice for this OS // Test is temporarily disabled until this is resolved // UNSUPPORTED: cuda && windows +// UNSUPPORTED: esimd_emulator // // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out // RUN: %HOST_RUN_PLACEHOLDER %t1.out diff --git a/SYCL/USM/depends_on.cpp b/SYCL/USM/depends_on.cpp index f52408c640..eb8a8587a1 100644 --- a/SYCL/USM/depends_on.cpp +++ b/SYCL/USM/depends_on.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==----------------- depends_on.cpp - depends_on test ---------------------==// // diff --git a/SYCL/USM/dmemll.cpp b/SYCL/USM/dmemll.cpp index 74b19cf9f7..19cfa47afb 100644 --- a/SYCL/USM/dmemll.cpp +++ b/SYCL/USM/dmemll.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==------------------- dmemll.cpp - Device Memory Linked List test --------==// // diff --git a/SYCL/USM/dmemllaligned.cpp b/SYCL/USM/dmemllaligned.cpp index cfcc1ea767..93ac0eda57 100644 --- a/SYCL/USM/dmemllaligned.cpp +++ b/SYCL/USM/dmemllaligned.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==---- dmemllaligned.cpp - Aligned Device Memory Linked List test --------==// // diff --git a/SYCL/USM/fill.cpp b/SYCL/USM/fill.cpp index 5885c901d7..86d1c86f17 100644 --- a/SYCL/USM/fill.cpp +++ b/SYCL/USM/fill.cpp @@ -11,6 +11,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/USM/hmemll.cpp b/SYCL/USM/hmemll.cpp index f5f9a5378e..8f6a39e130 100644 --- a/SYCL/USM/hmemll.cpp +++ b/SYCL/USM/hmemll.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==------------------- hmemll.cpp - Host Memory Linked List test ----------==// // diff --git a/SYCL/USM/hmemllaligned.cpp b/SYCL/USM/hmemllaligned.cpp index b3ca33647a..bfc30eae36 100644 --- a/SYCL/USM/hmemllaligned.cpp +++ b/SYCL/USM/hmemllaligned.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==---- hmemllaligned.cpp - Aligned Host Memory Linked List test ----------==// // diff --git a/SYCL/USM/memadvise.cpp b/SYCL/USM/memadvise.cpp index d8b1a6a12d..ecbb8a9a29 100644 --- a/SYCL/USM/memadvise.cpp +++ b/SYCL/USM/memadvise.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==---------------- memadvise.cpp - Shared Memory Linked List test --------==// // diff --git a/SYCL/USM/memcpy.cpp b/SYCL/USM/memcpy.cpp index f0e1c54364..12a926b0a2 100644 --- a/SYCL/USM/memcpy.cpp +++ b/SYCL/USM/memcpy.cpp @@ -11,6 +11,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/USM/memset.cpp b/SYCL/USM/memset.cpp index f85478851c..6948905f37 100644 --- a/SYCL/USM/memset.cpp +++ b/SYCL/USM/memset.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==---- memset.cpp - USM memset test --------------------------------------==// // diff --git a/SYCL/USM/mixed.cpp b/SYCL/USM/mixed.cpp index a37e71a264..e44eb4f735 100644 --- a/SYCL/USM/mixed.cpp +++ b/SYCL/USM/mixed.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==------------------- mixed.cpp - Mixed Memory test ---------------------==// // diff --git a/SYCL/USM/mixed2.cpp b/SYCL/USM/mixed2.cpp index 42efa86822..f513c7b2e3 100644 --- a/SYCL/USM/mixed2.cpp +++ b/SYCL/USM/mixed2.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==------------------- mixed2.cpp - Mixed Memory test ---------------------==// // diff --git a/SYCL/USM/mixed2template.cpp b/SYCL/USM/mixed2template.cpp index 9feadabf5a..f4ddc2f1f7 100644 --- a/SYCL/USM/mixed2template.cpp +++ b/SYCL/USM/mixed2template.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==---------- mixed2template.cpp - Mixed Memory with Templatestest --------==// // diff --git a/SYCL/USM/mixed_queue.cpp b/SYCL/USM/mixed_queue.cpp index bfcb044cf9..05be97d37d 100644 --- a/SYCL/USM/mixed_queue.cpp +++ b/SYCL/USM/mixed_queue.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==-------------- mixed_queue.cpp - Mixed Memory test ---------------------==// // diff --git a/SYCL/USM/pfor_flatten.cpp b/SYCL/USM/pfor_flatten.cpp index 109965fdfe..4c401fe803 100644 --- a/SYCL/USM/pfor_flatten.cpp +++ b/SYCL/USM/pfor_flatten.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==--------------- pfor_flatten.cpp - Kernel Launch Flattening test -------==// // diff --git a/SYCL/USM/pfor_flatten_range_shortcut.cpp b/SYCL/USM/pfor_flatten_range_shortcut.cpp index ee3fcabe43..86c46c9f11 100644 --- a/SYCL/USM/pfor_flatten_range_shortcut.cpp +++ b/SYCL/USM/pfor_flatten_range_shortcut.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==-- pfor_flatten_range_shortcut.cpp - Kernel Launch Flattening test -----==// // diff --git a/SYCL/USM/pointer_query.cpp b/SYCL/USM/pointer_query.cpp index 330e08a1ab..eb1ba53fe6 100644 --- a/SYCL/USM/pointer_query.cpp +++ b/SYCL/USM/pointer_query.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==-------------- pointer_query.cpp - Pointer Query test ------------------==// // diff --git a/SYCL/USM/prefetch.cpp b/SYCL/USM/prefetch.cpp index 21a5b59885..6a695ce8fa 100644 --- a/SYCL/USM/prefetch.cpp +++ b/SYCL/USM/prefetch.cpp @@ -9,6 +9,7 @@ // cuMemPrefetchAsync returns cudaErrorInvalidDevice for this OS // Test is temporarily disabled until this is resolved // UNSUPPORTED: cuda && windows +// UNSUPPORTED: esimd_emulator // // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out // RUN: %HOST_RUN_PLACEHOLDER %t1.out diff --git a/SYCL/USM/queue_wait.cpp b/SYCL/USM/queue_wait.cpp index 4d5d69a62b..9f1b6617c3 100644 --- a/SYCL/USM/queue_wait.cpp +++ b/SYCL/USM/queue_wait.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out +// UNSUPPORTED: esimd_emulator #include diff --git a/SYCL/USM/smemll.cpp b/SYCL/USM/smemll.cpp index 8c54109912..b912748e7c 100644 --- a/SYCL/USM/smemll.cpp +++ b/SYCL/USM/smemll.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==------------------- smemll.cpp - Shared Memory Linked List test --------==// // diff --git a/SYCL/USM/smemllaligned.cpp b/SYCL/USM/smemllaligned.cpp index b3c2db8c2c..8c39bb8669 100644 --- a/SYCL/USM/smemllaligned.cpp +++ b/SYCL/USM/smemllaligned.cpp @@ -3,6 +3,7 @@ // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out // RUN: %ACC_RUN_PLACEHOLDER %t1.out +// UNSUPPORTED: esimd_emulator //==---- smemllaligned.cpp - Aligned Shared Memory Linked List test --------==// // diff --git a/SYCL/lit.cfg.py b/SYCL/lit.cfg.py index 4e28715879..f73e98ecbd 100644 --- a/SYCL/lit.cfg.py +++ b/SYCL/lit.cfg.py @@ -171,7 +171,8 @@ 'opencl', 'cuda', 'hip', - 'level_zero'] + 'level_zero', + 'esimd_emulator'] if config.sycl_be not in supported_sycl_be: lit_config.error("Unknown SYCL BE specified '" +