[SYCL][ESIMD][EMU] Removing HOST_RUN_PLACEHOLDER for ESIMD Kernels (#823)

- Host backend and esimd_emulator backend share same header file and
macro definition (~__SYCL_DEVICE_ONLY__) for memory intrinsic support
from $LLVM_SRC/sycl/include/sycl/ext/intel/experimental/esimd/detail/memory_intrin.h

- This sharing prevents the same header file from supporting both
backend types as some intrinsic support
implementations (e.g. __esimd_media_ld/st) are different and they are
chosen during kernel compilation - while backends are chosen during
kernel execution

- Therefore, tests invoking those implementations must not contain
'HOST_RUN_PLACEHOLDER'

Author: dongkyunahn-intel

SYCL/ESIMD/BitonicSortKv2.cpp

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

SYCL/ESIMD/PrefixSum.cpp

@@ -8,7 +8,6 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda || hip
 // RUN: %clangxx -fsycl %s -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out 20
 // RUN: %GPU_RUN_PLACEHOLDER %t.out 20
 
 #include "esimd_test_utils.hpp"

SYCL/ESIMD/Prefix_Local_sum1.cpp

@@ -8,7 +8,6 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda || hip
 // RUN: %clangxx -fsycl %s -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out 20
 // RUN: %GPU_RUN_PLACEHOLDER %t.out 20
 
 #include "esimd_test_utils.hpp"

SYCL/ESIMD/Prefix_Local_sum2.cpp

@@ -8,7 +8,6 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda || hip
 // RUN: %clangxx -fsycl %s -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out 20
 // RUN: %GPU_RUN_PLACEHOLDER %t.out 20
 
 #include "esimd_test_utils.hpp"

SYCL/ESIMD/Prefix_Local_sum3.cpp

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

SYCL/ESIMD/Stencil.cpp

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

SYCL/ESIMD/api/svm_gather_scatter.cpp

@@ -8,7 +8,6 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda || hip
 // RUN: %clangxx -fsycl %s -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out
 // RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 // Regression test for SVM gather/scatter API.

SYCL/ESIMD/histogram.cpp

@@ -10,7 +10,6 @@
 // TODO: esimd_emulator fails due to outdated __esimd_media_ld
 // XFAIL: esimd_emulator
 // RUN: %clangxx -fsycl %s -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out
 // RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"

SYCL/ESIMD/histogram_2d.cpp

@@ -10,7 +10,6 @@
 // TODO: esimd_emulator fails due to outdated __esimd_media_ld
 // XFAIL: esimd_emulator
 // RUN: %clangxx -fsycl %s -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out
 // RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"

SYCL/ESIMD/histogram_raw_send.cpp

@@ -12,7 +12,6 @@
 // TODO: esimd_emulator fails due to outdated __esimd_media_ld
 // XFAIL: esimd_emulator
 // RUN: %clangxx -fsycl %s -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out
 // RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 // The test checks raw send functionality with atomic write implementation

SYCL/ESIMD/kmeans/kmeans.cpp

@@ -8,7 +8,6 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda || hip
 // RUN: %clangxx -fsycl %s -I%S/.. -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out %S/points.csv
 // RUN: %GPU_RUN_PLACEHOLDER %t.out %S/points.csv
 
 #include "kmeans.h"

SYCL/ESIMD/linear/linear.cpp

@@ -10,7 +10,6 @@
 // TODO: esimd_emulator fails due to outdated __esimd_media_ld
 // 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
 
 #include "bitmap_helpers.h"

SYCL/ESIMD/matrix_transpose_usm.cpp

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

SYCL/ESIMD/reduction.cpp

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

SYCL/ESIMD/stencil2.cpp

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

SYCL/ESIMD/test_id_3d.cpp

@@ -8,7 +8,6 @@
 // REQUIRES: gpu
 // UNSUPPORTED: cuda || hip
 // RUN: %clangxx -fsycl %s -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out
 // RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 // This test checks that multi-dimensional sycl::item can be used in ESIMD

SYCL/ESIMD/vadd_2d.cpp

@@ -10,7 +10,6 @@
 // TODO: esimd_emulator fails due to outdated __esimd_media_ld
 // XFAIL: esimd_emulator
 // RUN: %clangxx -fsycl %s -o %t.out
-// RUN: %HOST_RUN_PLACEHOLDER %t.out
 // RUN: %GPU_RUN_PLACEHOLDER %t.out
 
 #include "esimd_test_utils.hpp"

SYCL/ESIMD/vadd_usm.cpp

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