[ESIMD] Remove -vc-codegen option for ESIMD tests

SYCL/ESIMD/BitonicSortK.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/BitonicSortKv2.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/PrefixSum.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out 20
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out 20
+// RUN: %GPU_RUN_PLACEHOLDER %t.out 20
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/Prefix_Local_sum1.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out 20
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out 20
+// RUN: %GPU_RUN_PLACEHOLDER %t.out 20
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/Prefix_Local_sum2.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out 20
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out 20
+// RUN: %GPU_RUN_PLACEHOLDER %t.out 20
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/Prefix_Local_sum3.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/Stencil.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/accessor.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl -D_CRT_SECURE_NO_WARNINGS=1 %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 // This test checks that accessor-based memory accesses work correctly in ESIMD.
 

SYCL/ESIMD/accessor_gather_scatter.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 //
 // The test checks functionality of the gather/scatter accessor-based ESIMD
 // intrinsics.

SYCL/ESIMD/accessor_load_store.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 //
 // The test checks functionality of the scalar load/store accessor-based ESIMD
 // intrinsics.

SYCL/ESIMD/dp4a.cpp

@@ -8,7 +8,7 @@
 // TODO enable on Windows
 // REQUIRES: linux && gpu
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 // TODO : Enable test for new GPU device
 // XFAIL: *
 

SYCL/ESIMD/histogram.cpp

@@ -10,7 +10,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/histogram_256_slm.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/histogram_256_slm_spec.cpp

@@ -9,7 +9,7 @@
 // REQUIRES: linux && gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out 16
+// RUN: %GPU_RUN_PLACEHOLDER %t.out 16
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/histogram_2d.cpp

@@ -10,7 +10,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/histogram_raw_send.cpp

@@ -11,7 +11,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/kmeans/kmeans.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -I%S/.. -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out %S/points.big.json
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out %S/points.big.json
+// RUN: %GPU_RUN_PLACEHOLDER %t.out %S/points.big.json
 //
 
 #include "kmeans.h"

SYCL/ESIMD/linear/linear.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -I%S/.. -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out %S/linear_in.bmp %S/linear_gold_hw.bmp
-// RUN: %ESIMD_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
 
 #include "bitmap_helpers.h"
 #include "esimd_test_utils.hpp"

SYCL/ESIMD/mandelbrot/mandelbrot.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -I%S/.. -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out %T/output.ppm %S/golden_hw.ppm
+// RUN: %GPU_RUN_PLACEHOLDER %t.out %T/output.ppm %S/golden_hw.ppm
 
 #include "esimd_test_utils.hpp"
 #include <CL/sycl.hpp>

SYCL/ESIMD/mandelbrot/mandelbrot_spec.cpp

@@ -9,7 +9,7 @@
 // REQUIRES: linux && gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -I%S/.. -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out %T/output.ppm %S/golden_hw.ppm 512 -2.09798 -1.19798 0.004 4.0
+// RUN: %GPU_RUN_PLACEHOLDER %t.out %T/output.ppm %S/golden_hw.ppm 512 -2.09798 -1.19798 0.004 4.0
 
 #include "esimd_test_utils.hpp"
 #include <CL/sycl.hpp>

SYCL/ESIMD/matrix_transpose.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/matrix_transpose_glb.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/matrix_transpose_usm.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/private_memory/pm_access_1.cpp

@@ -9,4 +9,4 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl -I%S/.. %S/Inputs/pm_common.cpp -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out 1
+// RUN: %GPU_RUN_PLACEHOLDER %t.out 1

SYCL/ESIMD/private_memory/pm_access_2.cpp

@@ -9,4 +9,4 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl -I%S/.. %S/Inputs/pm_common.cpp -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out 2
+// RUN: %GPU_RUN_PLACEHOLDER %t.out 2

SYCL/ESIMD/private_memory/pm_access_3.cpp

@@ -9,4 +9,4 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl -I%S/.. %S/Inputs/pm_common.cpp -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out 3
+// RUN: %GPU_RUN_PLACEHOLDER %t.out 3

SYCL/ESIMD/reduction.cpp

@@ -10,7 +10,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/regression/big_const_initializer.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl -I%S/.. %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 // This test checks that ESIMD program with big constant initializer list can
 // compile and run correctly.

SYCL/ESIMD/regression/unused_load.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl -I%S/.. %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 // This test checks that ESIMD JIT compilation does not crash on unused
 // block_load invocation.

SYCL/ESIMD/slm_barrier.cpp

@@ -8,7 +8,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/slm_split_barrier.cpp

@@ -7,7 +7,7 @@
 //===----------------------------------------------------------------------===//
 // REQUIRES: gpu
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 // UNSUPPORTED: cuda
 
 #include "esimd_test_utils.hpp"

SYCL/ESIMD/spec_const_redefine_esimd.cpp

@@ -1,7 +1,7 @@
 // TODO enable on Windows
 // REQUIRES: linux && gpu
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: env SYCL_PI_TRACE=2 %ESIMD_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER
+// RUN: env SYCL_PI_TRACE=2 %GPU_RUN_PLACEHOLDER %t.out 2>&1 %GPU_CHECK_PLACEHOLDER
 // UNSUPPORTED: cuda
 
 //==----------- spec_const_redefine_esimd.cpp ------------------------------==//

SYCL/ESIMD/stencil2.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/test_id_3d.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 // This test checks that multi-dimensional sycl::item can be used in ESIMD
 // kernels.

SYCL/ESIMD/vadd_2d.cpp

@@ -10,7 +10,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/vadd_2d_acc.cpp

@@ -7,7 +7,7 @@
 //===----------------------------------------------------------------------===//
 // REQUIRES: gpu
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 // UNSUPPORTED: cuda
 

SYCL/ESIMD/vadd_raw_send.cpp

@@ -9,7 +9,7 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/ESIMD/vadd_usm.cpp

@@ -9,7 +9,7 @@
 // UNSUPPORTED: cuda
 // RUN: %clangxx-esimd -fsycl %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out
-// RUN: %ESIMD_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"
 

SYCL/lit.cfg.py

@@ -147,9 +147,6 @@
                      config.sycl_be +
                      "' supported values are opencl, cuda, level_zero")
 
-esimd_run_substitute = "env SYCL_DEVICE_FILTER={SYCL_PLUGIN}:gpu SYCL_PROGRAM_COMPILE_OPTIONS=-vc-codegen".format(SYCL_PLUGIN=config.sycl_be)
-config.substitutions.append( ('%ESIMD_RUN_PLACEHOLDER',  esimd_run_substitute) )
-
 config.substitutions.append( ('%clangxx-esimd',  config.dpcpp_compiler +
                               ' ' + '-fsycl-explicit-simd' + ' ' +
                               config.cxx_flags ) )