Skip to content

Commit 473fa48

Browse files
authored
[SYCL][Graph] Add E2E test for kernel bundle regression (#18173)
- Add test which was observed to be hanging after recent kernel bundle changes - Test is modified version of update test which removes update functionality so it runs on current CI Observed issue was fixed in #18157
1 parent 7d063d2 commit 473fa48

File tree

1 file changed

+58
-0
lines changed

1 file changed

+58
-0
lines changed
Lines changed: 58 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,58 @@
1+
// RUN: %{build} -o %t.out
2+
// RUN: %{run} %t.out %S/../Inputs/Kernels/update_with_indices_accessor.spv
3+
// Extra run to check for leaks in Level Zero using UR_L0_LEAKS_DEBUG
4+
// RUN: %if level_zero %{env SYCL_PI_LEVEL_ZERO_USE_IMMEDIATE_COMMANDLISTS=0 %{l0_leak_check} %{run} %t.out %S/../Inputs/Kernels/update_with_indices_accessor.spv 2>&1 | FileCheck %s --implicit-check-not=LEAK %}
5+
// Extra run to check for immediate-command-list in Level Zero
6+
// RUN: %if level_zero %{env SYCL_PI_LEVEL_ZERO_USE_IMMEDIATE_COMMANDLISTS=1 %{l0_leak_check} %{run} %t.out %S/../Inputs/Kernels/update_with_indices_accessor.spv 2>&1 | FileCheck %s --implicit-check-not=LEAK %}
7+
8+
// REQUIRES: level_zero
9+
10+
// Modified version of Update/update_with_indices_accessor_spv.cpp that does
11+
// not require the full graph aspect, test was hanging after some changes to
12+
// kernel bundles so adding this test for the CI which doesn't support update
13+
14+
#include "../graph_common.hpp"
15+
16+
int main(int, char **argv) {
17+
queue Queue{};
18+
sycl::kernel_bundle KernelBundle = loadKernelsFromFile(Queue, argv[1]);
19+
const auto getKernel =
20+
[](sycl::kernel_bundle<sycl::bundle_state::executable> &bundle,
21+
const std::string &name) {
22+
return bundle.ext_oneapi_get_kernel(name);
23+
};
24+
25+
kernel kernel = getKernel(
26+
KernelBundle, "_ZTSZZ4mainENKUlRN4sycl3_V17handlerEE_clES2_EUlvE_");
27+
28+
const size_t N = 1024;
29+
30+
std::vector<int> HostDataA(N, 0);
31+
32+
buffer BufA{HostDataA};
33+
BufA.set_write_back(false);
34+
35+
{
36+
exp_ext::command_graph Graph{
37+
Queue.get_context(),
38+
Queue.get_device(),
39+
{exp_ext::property::graph::assume_buffer_outlives_graph{}}};
40+
41+
Graph.add([&](handler &cgh) {
42+
auto Acc = BufA.get_access(cgh);
43+
44+
cgh.set_arg(0, Acc);
45+
cgh.single_task(kernel);
46+
});
47+
48+
auto ExecGraph = Graph.finalize();
49+
Queue.ext_oneapi_graph(ExecGraph).wait();
50+
}
51+
// Copy data back to host
52+
Queue.copy(BufA.get_access(), HostDataA.data()).wait();
53+
54+
for (size_t i = 0; i < N; i++) {
55+
assert(HostDataA[i] == i);
56+
}
57+
return 0;
58+
}

0 commit comments

Comments
 (0)