ESIMD/unified_memory_api/prefetch_acc_stateful_pvc.cpp failed on unrelated changes

[Xewar313]

Describe the bug

https://github.com/intel/llvm/actions/runs/13763836130/job/38491973677?pr=17297

To reproduce

failed on CI

Environment

failed on CI

Additional context

[AlexeySachkov]

Fail log for reference:

FAIL: SYCL :: ESIMD/unified_memory_api/prefetch_acc_stateful_pvc.cpp (99 of 2281)
******************** TEST 'SYCL :: ESIMD/unified_memory_api/prefetch_acc_stateful_pvc.cpp' FAILED ********************
Exit Code: 1

Command Output (stdout):
--
Running on Intel(R) Data Center GPU Max 1100, driver=[1.6.32567.170000]
Running case: T=int8_t, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=int8_t, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=int8_t, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=int8_t, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int8_t, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int8_t, N=32, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int8_t, N=64, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int8_t, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int8_t, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int8_t, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int8_t, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int8_t, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int8_t, N=32, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int8_t, N=64, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int8_t, N=6, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int16_t, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=int16_t, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=int16_t, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=int16_t, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int16_t, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int16_t, N=32, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int16_t, N=64, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int16_t, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int16_t, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int16_t, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int16_t, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int16_t, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int16_t, N=32, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int16_t, N=64, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int16_t, N=6, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=half, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=half, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=half, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=half, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=half, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=half, N=32, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=half, N=64, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=half, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=half, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=half, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=half, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=half, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=half, N=32, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=half, N=64, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=half, N=6, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=uint32_t, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=uint32_t, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=uint32_t, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=uint32_t, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=uint32_t, N=32, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=uint32_t, N=64, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=uint32_t, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=32, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=64, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=6, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=16, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=32, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=uint32_t, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=uint32_t, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=uint32_t, N=6, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=float, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=float, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=float, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=float, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=float, N=32, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=float, N=64, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=float, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=32, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=64, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=6, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=16, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=32, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=float, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=float, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=float, N=6, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=32, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=64, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=32, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=64, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=6, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=16, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=32, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=sycl::ext::intel::experimental::esimd::tfloat32, N=6, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=int64_t, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=int64_t, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=int64_t, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int64_t, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int64_t, N=32, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int64_t, N=64, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int64_t, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=32, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=64, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=6, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=16, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=32, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int64_t, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=int64_t, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=int64_t, N=6, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=double, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=double, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=0
Running case: T=double, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=double, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=double, N=32, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=double, N=64, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=double, N=1, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=2, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=4, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=8, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=16, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=32, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=64, VS=1, MaskStride=2, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=6, VS=1, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=16, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=32, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=double, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=0
Running case: T=double, N=64, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
Running case: T=double, N=6, VS=2, MaskStride=3, Groups=8, Threads=16, use_mask=1
ACC case: T=int8_t,N=4 passed
ACC case: T=int8_t,N=8 passed
ACC case: T=int8_t,N=12 passed
ACC case: T=int8_t,N=16 passed
ACC case: T=int8_t,N=32 passed
ACC case: T=int8_t,N=64 passed
ACC case: T=int8_t,N=128 passed
ACC case: T=int8_t,N=256 passed
ACC case: T=int8_t,N=512 passed
ACC case: T=int16_t,N=2 passed
ACC case: T=int16_t,N=4 passed
ACC case: T=int16_t,N=6 passed
ACC case: T=int16_t,N=8 passed
ACC case: T=int16_t,N=16 passed
ACC case: T=int16_t,N=32 passed
ACC case: T=int16_t,N=64 passed
ACC case: T=int16_t,N=128 passed
ACC case: T=int16_t,N=256 passed
ACC case: T=half,N=2 passed
ACC case: T=half,N=4 passed
ACC case: T=half,N=6 passed
ACC case: T=half,N=8 passed
ACC case: T=half,N=16 passed
ACC case: T=half,N=32 passed
ACC case: T=half,N=64 passed
ACC case: T=half,N=128 passed
ACC case: T=half,N=256 passed
ACC case: T=uint32_t,N=1 passed
ACC case: T=uint32_t,N=2 passed
ACC case: T=uint32_t,N=3 passed
ACC case: T=uint32_t,N=4 passed
ACC case: T=uint32_t,N=8 passed
ACC case: T=uint32_t,N=16 passed
ACC case: T=uint32_t,N=32 passed
ACC case: T=uint32_t,N=64 passed
ACC case: T=uint32_t,N=128 passed
ACC case: T=float,N=1 passed
ACC case: T=float,N=2 passed
ACC case: T=float,N=3 passed
ACC case: T=float,N=4 passed
ACC case: T=float,N=8 passed
ACC case: T=float,N=16 passed
ACC case: T=float,N=32 passed
ACC case: T=float,N=64 passed
ACC case: T=float,N=128 passed
ACC case: T=sycl::ext::intel::experimental::esimd::tfloat32,N=1 passed
ACC case: T=sycl::ext::intel::experimental::esimd::tfloat32,N=2 passed
ACC case: T=sycl::ext::intel::experimental::esimd::tfloat32,N=3 passed
ACC case: T=sycl::ext::intel::experimental::esimd::tfloat32,N=4 passed
ACC case: T=sycl::ext::intel::experimental::esimd::tfloat32,N=8 passed
ACC case: T=sycl::ext::intel::experimental::esimd::tfloat32,N=16 passed
ACC case: T=sycl::ext::intel::experimental::esimd::tfloat32,N=32 passed
ACC case: T=sycl::ext::intel::experimental::esimd::tfloat32,N=64 passed
ACC case: T=sycl::ext::intel::experimental::esimd::tfloat32,N=128 passed
ACC case: T=int64_t,N=1 passed
ACC case: T=int64_t,N=2 passed
ACC case: T=int64_t,N=3 passed
ACC case: T=int64_t,N=4 passed
ACC case: T=int64_t,N=8 passed
ACC case: T=int64_t,N=16 passed
ACC case: T=int64_t,N=32 passed
ACC case: T=int64_t,N=64 passed
ACC case: T=double,N=1 passed
ACC case: T=double,N=2 passed
ACC case: T=double,N=3 passed
ACC case: T=double,N=4 passed
ACC case: T=double,N=8 passed
ACC case: T=double,N=16 passed
ACC case: T=double,N=32 passed
ACC case: T=double,N=64out[192] = 0x0 vs etalon = 0x3b436e301d0c79ba
out[193] = 0x0 vs etalon = 0x6d22e2823f316b47
out[194] = 0x0 vs etalon = 0x7272cdad42c407cc
out[195] = 0x0 vs etalon = 0x5b490ef04a7c7211
out[196] = 0x0 vs etalon = 0x24dd44b17cd8dfb7
out[197] = 0x0 vs etalon = 0x6f7b3b2c67d5a126
out[198] = 0x0 vs etalon = 0x6c60f7730541b391
out[199] = 0x0 vs etalon = 0x302aaab969cca878
out[200] = 0x0 vs etalon = 0x5154df72078734b1
out[201] = 0x0 vs etalon = 0xa69d43e06a2ac1a
out[202] = 0x0 vs etalon = 0x5df2d9a411a3596
out[203] = 0x0 vs etalon = 0x1d8a6cf40ff652ca
out[204] = 0x0 vs etalon = 0x3ec6c6514c6c5381
out[205] = 0x0 vs etalon = 0x1576ff79688ce314
out[206] = 0x0 vs etalon = 0x548d8dca1b46b326
out[207] = 0x0 vs etalon = 0x1bcce2250fd0fbfb
out[208] = 0x0 vs etalon = 0x38532ce108efc4a7
out[209] = 0x0 vs etalon = 0x4f0267422ac5fa8e
out[210] = 0x0 vs etalon = 0x4bb3cc732a4b7633
out[211] = 0x0 vs etalon = 0x75426c9f70911125
out[212] = 0x0 vs etalon = 0x27[2455](https://github.com/intel/llvm/actions/runs/13763836130/job/38491973677?pr=17297#step:17:2471)ea64bda7cb
out[213] = 0x0 vs etalon = 0x5866b24b13854d5d
out[214] = 0x0 vs etalon = 0x69ff5b5c08915d04
out[215] = 0x0 vs etalon = 0x7d51f5d53b543acf
out[216] = 0x0 vs etalon = 0x101891b607bbca13
out[217] = 0x0 vs etalon = 0x41f6e6e915f7bf50
out[218] = 0x0 vs etalon = 0x48d5ffa95f8153de
out[219] = 0x0 vs etalon = 0x25ee121b079cc5fb
out[220] = 0x0 vs etalon = 0x2beda75f3b651194
out[221] = 0x0 vs etalon = 0x7029a90f007b352a
out[222] = 0x0 vs etalon = 0x56abc4bb0bf68b34
out[223] = 0x0 vs etalon = 0x104c31250efef19c
 FAILED
FAILED

--
Command Output (stderr):
--
RUN: at line 10: env ONEAPI_DEVICE_SELECTOR=level_zero:gpu  /__w/llvm/llvm/build-e2e/ESIMD/unified_memory_api/Output/prefetch_acc_stateful_pvc.cpp.tmp.out
+ env ONEAPI_DEVICE_SELECTOR=level_zero:gpu /__w/llvm/llvm/build-e2e/ESIMD/unified_memory_api/Output/prefetch_acc_stateful_pvc.cpp.tmp.out

--

[sarnex]

@fineg74 Any ideas on this one? If I remember correctly you worked on prefetch. Thanks

[fineg74]

Given the fact that ESIMD wasn't touched for ages it is most likely a driver issue