Skip to content

Commit af42705

Browse files
authored
[SYCL][E2E] Workaround enum parameter free function kernel test failure (#20423)
We recently disabled dead arguments elimination optimization for free function kernels and most likely this triggered this issue. Default-constructed accessors used in the test only matter during FE stage when we instantiate necessary templates, but at LLVM IR level they are completely unused, because we do not access their fields or call any member functions. As such, with dead arguments elimination optimization disabled, SYCL RT stopped to ignore them and thus we uncovered a bug there with use-after-free. Considering that the test was aimed to test a different thing, a workaround is made to avoid sporadic failures of the test. #20225 was submitted to track the underlying issue. Most likely some kind of reference to the accessor is preserved within `handler` whose lifetime may be longer than the one of the accessor, thus causing `finalize` method to access a dead memory region.
1 parent b67580f commit af42705

File tree

1 file changed

+3
-6
lines changed

1 file changed

+3
-6
lines changed

sycl/test-e2e/FreeFunctionKernels/enum_parameter.cpp

Lines changed: 3 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -42,30 +42,27 @@ int main() {
4242

4343
bool flag1, flag2, flag3;
4444
flag1 = (flag2 = (flag3 = false));
45+
rAccType acc1;
46+
wAccType acc2;
47+
rwAccType acc3;
4548
{
4649
sycl::buffer<bool, 1> flagBuffer1(&flag1, 1);
4750
sycl::buffer<bool, 1> flagBuffer2(&flag2, 1);
4851
sycl::buffer<bool, 1> flagBuffer3(&flag3, 1);
4952
Queue.submit([&](sycl::handler &Handler) {
5053
flagType flagAcc1{flagBuffer1, Handler};
51-
rAccType acc1;
52-
5354
Handler.set_args(acc1, rMode, flagAcc1);
5455
Handler.single_task(Kernel1);
5556
});
5657

5758
Queue.submit([&](sycl::handler &Handler) {
5859
flagType flagAcc2{flagBuffer2, Handler};
59-
wAccType acc2;
60-
6160
Handler.set_args(acc2, wMode, flagAcc2);
6261
Handler.single_task(Kernel2);
6362
});
6463

6564
Queue.submit([&](sycl::handler &Handler) {
6665
flagType flagAcc3{flagBuffer3, Handler};
67-
rwAccType acc3;
68-
6966
Handler.set_args(acc3, rwMode, flagAcc3);
7067
Handler.single_task(Kernel3);
7168
});

0 commit comments

Comments
 (0)