Skip to content

[SYCL][E2E] Add support to filter out driver version #11427

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
aelovikov-intel merged 4 commits into from
Oct 5, 2023
Merged

[SYCL][E2E] Add support to filter out driver version #11427

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
4 commits
Select commit Hold shift + click to select a range
f4468c7
[SYCL][E2E] Add support to filter out driver version
aelovikov-intel Oct 4, 2023
aa0780f
Fix typo
aelovikov-intel Oct 4, 2023
c307e7b
Merge remote-tracking branch 'origin/sycl' into HEAD
aelovikov-intel Oct 5, 2023
65542e1
Copy REQUIRES-INTEL-DRIVER to 5 more tests
aelovikov-intel Oct 5, 2023
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
7 changes: 1 addition & 6 deletions sycl/test-e2e/ESIMD/accessor_local.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,3 +1,4 @@
// REQUIRES-INTEL-DRIVER: lin: 27202, win: 101.4677
Copy link
Contributor

@v-klochkov v-klochkov Oct 4, 2023 •
edited
Loading

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I agree that is really great feature. Thank you for adding it!

The only question from me: is this new filter completely disables the test or only '%{run}' portion of it?

Many tests are very useful as compile-only tests too. Otherwise, when the driver uplift comes, we may see not the expected runtime pass, but compilation fails caused by series of changes while the test was OFF.

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Oct 4, 2023

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

It skips completely. The reason I've done it is because compilation might be affected by the driver too (e.g., AOT compilation crash). I think that is pretty common for the Matrix project/tests (which triggered this PR).

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Oct 4, 2023

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Hm, thinking more about it, the argument above might not be true. However, I'd really hate to report PASS for something that hasn't been executed.

Copy link
Contributor

@v-klochkov v-klochkov Oct 4, 2023 •
edited
Loading

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

The old non-centralized checks in ESIMD monitored compile-time regressions, which was really good thing.

If only AOT compilation needs skip, then the best solution (IMO) would be to introduce %{build-aot} that is skipped if the driver doesn't meet the requirement. I have no idea though how difficult is to introduce %{build-aot}
So, if driver is too old:
%{build} - runs
%{build-aot} - skipped
%{run} - skipped

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Oct 4, 2023

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

How often have you caught the problems on a test with a runtime skip that could not be observed on other tests? Was it in CI or locally? If locally, was the new/good driver available to you?

Copy link
Contributor

@v-klochkov v-klochkov Oct 4, 2023 •
edited
Loading

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Cannot say how often. Perhaps it is the small chance to have overlooked regression.
Compile-time regression may happen when someone else does the changes, e.g. pulls from llvm.org and simply relies on CI, while CI might not catch it.

Don't consider my comments as blocking. The new centralized filter is great in general, even though it has some cons if it skips %{build}.

Copy link
Contributor

@sarnex sarnex Oct 4, 2023 •
edited
Loading

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

If we are worried about it the ESIMD team could find what tests are currently skipped in CI due to old driver and create compile-only tests for them if they have APIs that aren't locked down elsewhere.

Copy link
Contributor Author

@aelovikov-intel aelovikov-intel Oct 4, 2023

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I don't like implicit skips because the give you a false sense of all tests passing. UNSUPPORTED is much better in this regard. I think the right way to enable compilation is to create a fake second test that will compile-only the one disabled with this new markup:

# a.compile.cpp
// RUN: %{build} %S/a.cpp -o %t.out
#a.cpp
// REQUIRES-INTEL-DRIVER: ...
// NOTE: Remove a.compile.cpp once the above line is removed.
...

If we think a.cpp is important enough

// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
// This test verifies usage of local_accessor methods operator[]
Expand Down Expand Up @@ -112,12 +113,6 @@ int main() {
std::cout << "Running on " << Dev.get_info<sycl::info::device::name>()
<< ", Local memory size available : " << DeviceSLMSize << std::endl;

if (!isGPUDriverGE(Q, esimd_test::GPUDriverOS::LinuxAndWindows, "27202",
"101.4677")) {
std::cout << "Skipped. The test requires GPU driver 1.3.27202 or newer.\n";
return 0;
}

uint32_t LocalRange = 16;
uint32_t GlobalRange = LocalRange * 2; // 2 groups.

Expand Down
16 changes: 6 additions & 10 deletions sycl/test-e2e/ESIMD/api/slm_gather_scatter_heavy.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -9,7 +9,12 @@
// RUN: %{run} %t.out
//
// UNSUPPORTED: esimd_emulator

//
// GPU driver had an error in handling of SLM aligned block_loads/stores,
// which has been fixed only in "1.3.26816", and in win/opencl version going
// _after_ 101.4575.
// REQUIRES-INTEL-DRIVER: lin: 26816, win: 101.4576
//
// The test checks functionality of the slm gather/scatter ESIMD intrinsics.
// It varies element type, vector length and stride of gather/scatter operation.
// For simplicity of calculations, workgroup size (number of work items same
Expand Down Expand Up @@ -445,15 +450,6 @@ int main() {
auto dev = q.get_device();
esimd_test::printTestLabel(q);

// GPU driver had an error in handling of SLM aligned block_loads/stores,
// which has been fixed only in "1.3.26816", and in win/opencl version going
// _after_ 101.4575.
if (!esimd_test::isGPUDriverGE(q, esimd_test::GPUDriverOS::LinuxAndWindows,
"26816", "101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26816 or newer.\n";
return 0;
}

bool passed = true;
passed &= test_vl1<char, 3>(q);
passed &= test<char, 16, 3>(q);
Expand Down
10 changes: 1 addition & 9 deletions sycl/test-e2e/ESIMD/assert.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,3 +1,4 @@
// REQUIRES-INTEL-DRIVER: lin: 26816, win: 101.4576
// REQUIRES: linux && level_zero

// RUN: %{build} -DSYCL_FALLBACK_ASSERT=1 -o %t.out
Expand All @@ -19,15 +20,6 @@ using namespace sycl;
int main() {
queue Q(esimd_test::ESIMDSelector, esimd_test::createExceptionHandler());
esimd_test::printTestLabel(Q);
if (!esimd_test::isGPUDriverGE(Q, esimd_test::GPUDriverOS::LinuxAndWindows,
"26816", "101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26816 or newer.\n";
// Additionally, print expected messages to pass FileCheck checks below.
std::cerr << "Assert called: Id != 31 && \"assert message31\"\n";
std::cerr << "assert.cpp, Line 29, Function auto main()::(anonymous class)"
<< "::operator()(id<1>) const, gid(31, 0, 0), lid(3, 0, 0)\n";
return 0;
}

try {
Q.parallel_for(range<1>{100}, [=](id<1> Id) SYCL_ESIMD_KERNEL {
Expand Down
1 change: 1 addition & 0 deletions sycl/test-e2e/ESIMD/dword_local_accessor_atomic_smoke.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -7,6 +7,7 @@
//===----------------------------------------------------------------------===//
// This test checks DWORD local accessor atomic operations.
//===----------------------------------------------------------------------===//
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
//
Expand Down
1 change: 1 addition & 0 deletions sycl/test-e2e/ESIMD/dword_local_accessor_atomic_smoke_cmpxchg.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -7,6 +7,7 @@
//===----------------------------------------------------------------------===//
// This test checks DWORD local accessor cmpxchg atomic operations.
//===----------------------------------------------------------------------===//
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
//
Expand Down
1 change: 1 addition & 0 deletions sycl/test-e2e/ESIMD/dword_local_accessor_atomic_smoke_cmpxchg_scalar_off.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -9,6 +9,7 @@
// This test checks DWORD local accessor cmpxchg atomic operations with scalar
// offset.
//===----------------------------------------------------------------------===//
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
//
Expand Down
1 change: 1 addition & 0 deletions sycl/test-e2e/ESIMD/dword_local_accessor_atomic_smoke_scalar_off.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -7,6 +7,7 @@
//===----------------------------------------------------------------------===//
// This test checks DWORD local accessor atomic operations with scalar offset.
//===----------------------------------------------------------------------===//
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
//
Expand Down
8 changes: 1 addition & 7 deletions sycl/test-e2e/ESIMD/ext_math.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -5,6 +5,7 @@
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
//
//===----------------------------------------------------------------------===//
// REQUIRES-INTEL-DRIVER: lin: 27012, win: 101.4576
// DEFINE: %{mathflags} = %if cl_options %{/clang:-fno-fast-math%} %else %{-fno-fast-math%}
// RUN: %{build} -fsycl-device-code-split=per_kernel %{mathflags} -o %t.out
// RUN: %{run} %t.out
Expand Down Expand Up @@ -480,13 +481,6 @@ int main(void) {
queue Q(esimd_test::ESIMDSelector, esimd_test::createExceptionHandler());
esimd_test::printTestLabel(Q);
auto Dev = Q.get_device();
#ifndef SKIP_NEW_GPU_DRIVER_VERSION_CHECK
if (!esimd_test::isGPUDriverGE(Q, esimd_test::GPUDriverOS::LinuxAndWindows,
"27012", "101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.27012 or newer.\n";
return 0;
}
#endif

bool Pass = true;
#ifdef TEST_IEEE_DIV_REM
Expand Down
1 change: 1 addition & 0 deletions sycl/test-e2e/ESIMD/ext_math_saturate.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -5,6 +5,7 @@
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
//
//===----------------------------------------------------------------------===//
// REQUIRES-INTEL-DRIVER: lin: 27012, win: 101.4576
// DEFINE: %{mathflags} = %if cl_options %{/clang:-fno-fast-math%} %else %{-fno-fast-math%}
// RUN: %{build} -fsycl-device-code-split=per_kernel %{mathflags} -o %t.out
// RUN: %{run} %t.out
Expand Down
7 changes: 1 addition & 6 deletions sycl/test-e2e/ESIMD/local_accessor_block_load_store.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -5,6 +5,7 @@
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
//
//===----------------------------------------------------------------------===//
// REQUIRES-INTEL-DRIVER: lin: 27202, win: 101.4677
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
// This test verifies usage of block_load/block_store for local_accessor.
Expand Down Expand Up @@ -98,12 +99,6 @@ int main() {
auto DeviceSLMSize = Dev.get_info<sycl::info::device::local_mem_size>();
esimd_test::printTestLabel(Q, "Local memory size available", DeviceSLMSize);

if (!isGPUDriverGE(Q, esimd_test::GPUDriverOS::LinuxAndWindows, "27202",
"101.4677")) {
std::cout << "Skipped. The test requires GPU driver 1.3.27202 or newer.\n";
return 0;
}

constexpr size_t Align4 = 4;
constexpr size_t Align8 = 8;
constexpr size_t Align16 = 16;
Expand Down
7 changes: 2 additions & 5 deletions sycl/test-e2e/ESIMD/local_accessor_copy_to_from.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -5,6 +5,8 @@
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
//
//===----------------------------------------------------------------------===//
// REQUIRES-INTEL-DRIVER: lin: 27202, win: 101.4677
//
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
//
Expand Down Expand Up @@ -100,11 +102,6 @@ int main() {

auto dev = q.get_device();
std::cout << "Running on " << dev.get_info<info::device::name>() << "\n";
if (!isGPUDriverGE(q, esimd_test::GPUDriverOS::LinuxAndWindows, "27202",
"101.4677")) {
std::cout << "Skipped. The test requires GPU driver 1.3.27202 or newer.\n";
return 0;
}

bool passed = true;
passed &= test<char, 1>(q);
Expand Down
7 changes: 1 addition & 6 deletions sycl/test-e2e/ESIMD/local_accessor_gather_scatter.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -5,6 +5,7 @@
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
//
//===----------------------------------------------------------------------===//
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
//
Expand Down Expand Up @@ -100,12 +101,6 @@ int main(void) {
auto dev = q.get_device();
std::cout << "Running on " << dev.get_info<info::device::name>() << "\n";

if (!isGPUDriverGE(q, esimd_test::GPUDriverOS::LinuxAndWindows, "26690",
"101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26690 or newer.\n";
return 0;
}

bool passed = true;
passed &= test<char, 8, 1>(q);
passed &= test<char, 16, 3>(q);
Expand Down
7 changes: 1 addition & 6 deletions sycl/test-e2e/ESIMD/local_accessor_gather_scatter_rgba.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -8,6 +8,7 @@
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
// UNSUPPORTED: esimd_emulator
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
// The test checks functionality of the gather_rgba/scatter_rgba local
// accessor-based ESIMD intrinsics.

Expand Down Expand Up @@ -185,12 +186,6 @@ int main(void) {
std::cout << "Running on " << dev.get_info<sycl::info::device::name>()
<< "\n";

if (!isGPUDriverGE(q, esimd_test::GPUDriverOS::LinuxAndWindows, "26690",
"101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26690 or newer.\n";
return 0;
}

bool passed = true;
passed &= test<int, 8>(q);
passed &= test<int, 16>(q);
Expand Down
6 changes: 1 addition & 5 deletions sycl/test-e2e/ESIMD/lsc/local_accessor_atomic_smoke.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -8,6 +8,7 @@
// This test checks local accessor atomic operations.
//===----------------------------------------------------------------------===//
// REQUIRES: gpu-intel-pvc
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
//
Expand Down Expand Up @@ -623,11 +624,6 @@ int main(void) {

auto dev = q.get_device();
std::cout << "Running on " << dev.get_info<info::device::name>() << "\n";
if (!isGPUDriverGE(q, esimd_test::GPUDriverOS::LinuxAndWindows, "26690",
"101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26690 or newer.\n";
return 0;
}

bool passed = true;
#ifndef CMPXCHG_TEST
Expand Down
6 changes: 1 addition & 5 deletions sycl/test-e2e/ESIMD/lsc/lsc_local_accessor_block_load_store.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -6,6 +6,7 @@
//
//===----------------------------------------------------------------------===//
// REQUIRES: gpu-intel-pvc
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
//
Expand Down Expand Up @@ -97,11 +98,6 @@ int main() {
auto DeviceSLMSize = Dev.get_info<sycl::info::device::local_mem_size>();
std::cout << "Running on " << Dev.get_info<sycl::info::device::name>()
<< ", Local memory size available : " << DeviceSLMSize << std::endl;
if (!isGPUDriverGE(Q, esimd_test::GPUDriverOS::LinuxAndWindows, "26690",
"101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26690 or newer.\n";
return 0;
}

constexpr uint32_t LocalRange = 16;
constexpr uint32_t GlobalRange = LocalRange * 2; // 2 groups.
Expand Down
6 changes: 1 addition & 5 deletions sycl/test-e2e/ESIMD/lsc/lsc_local_accessor_gather_scatter.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -6,6 +6,7 @@
//
//===----------------------------------------------------------------------===//
// REQUIRES: gpu-intel-pvc
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out
//
Expand Down Expand Up @@ -101,11 +102,6 @@ int main(void) {

auto dev = q.get_device();
std::cout << "Running on " << dev.get_info<info::device::name>() << "\n";
if (!isGPUDriverGE(q, esimd_test::GPUDriverOS::LinuxAndWindows, "26690",
"101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26690 or newer.\n";
return 0;
}

bool passed = true;
passed &= test<char, 8, 1>(q);
Expand Down
15 changes: 6 additions & 9 deletions sycl/test-e2e/ESIMD/slm_block_load_store.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -9,6 +9,12 @@
// RUN: %{run} %t.out
//
// UNSUPPORTED: esimd_emulator
//
// GPU driver had an error in handling of SLM aligned block_loads/stores,
// which has been fixed only in "1.3.26816", and in win/opencl version going
// _after_ 101.4575.
// REQUIRES-INTEL-DRIVER: lin: 26816, win: 101.4576
//
// Even though the driver we use in CI is now new enough so this test doesn't
// skip itself, it still fails. Temporary XFAIL-ing it while it is being
// investigated.
Expand Down Expand Up @@ -103,15 +109,6 @@ int main() {
auto DeviceSLMSize = Dev.get_info<sycl::info::device::local_mem_size>();
esimd_test::printTestLabel(Q, "Local memory size available", DeviceSLMSize);

// GPU driver had an error in handling of SLM aligned block_loads/stores,
// which has been fixed only in "1.3.26816", and in win/opencl version going
// _after_ 101.4575.
if (!esimd_test::isGPUDriverGE(Q, esimd_test::GPUDriverOS::LinuxAndWindows,
"26816", "101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26816 or newer.\n";
return 0;
}

constexpr size_t Align4 = 4;
constexpr size_t Align8 = 8;
constexpr size_t Align16 = 16;
Expand Down
13 changes: 5 additions & 8 deletions sycl/test-e2e/InvokeSimd/Regression/slm_load_store.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,3 +1,8 @@
// GPU driver had an error in handling of SLM aligned block_loads/stores,
// which has been fixed only in "1.3.26816", and in win/opencl version going
// _after_ 101.4575.
// REQUIRES-INTEL-DRIVER: lin: 26816, win: 101.4576
//
// RUN: %{build} -fno-sycl-device-code-split-esimd -Xclang -fsycl-allow-func-ptr -o %t.out
// RUN: env IGC_VCSaveStackCallLinkage=1 IGC_VCDirectCallsOnly=1 %{run} %t.out
//
Expand Down Expand Up @@ -79,14 +84,6 @@ int main(void) {
std::cout << "Running on " << Dev.get_info<sycl::info::device::name>()
<< std::endl;

// GPU driver had an error in handling of SLM aligned block_loads/stores,
// which has been fixed only in "1.3.26816", and in win/opencl version going
// _after_ 101.4575.
if (!isGPUDriverGE(Q, GPUDriverOS::LinuxAndWindows, "26816", "101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26816 or newer.\n";
return 0;
}

auto DeviceSLMSize = Dev.get_info<sycl::info::device::local_mem_size>();
std::cout << "Local Memory Size: " << DeviceSLMSize << std::endl;

Expand Down
7 changes: 2 additions & 5 deletions sycl/test-e2e/InvokeSimd/Spec/simd_size/simd8.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,6 +1,8 @@
// Test not intended to run on PVC
// UNSUPPORTED: gpu-intel-pvc
//
// REQUIRES-INTEL-DRIVER: lin: 26690, win: 101.4576
//
// Check that full compilation works:
// RUN: %{build} -fno-sycl-device-code-split-esimd -Xclang -fsycl-allow-func-ptr -o %t.out
// RUN: env IGC_VCSaveStackCallLinkage=1 IGC_VCDirectCallsOnly=1 %{run} %t.out
Expand All @@ -26,11 +28,6 @@ int main(void) {
std::cout << "Running on " << dev.get_info<sycl::info::device::name>()
<< "\n";

if (!isGPUDriverGE(q, GPUDriverOS::LinuxAndWindows, "26690", "101.4576")) {
std::cout << "Skipped. The test requires GPU driver 1.3.26690 or newer.\n";
return 0;
}

bool passed = true;

// simd_size 8
Expand Down
Loading