Fix those LIT tests that require newer GPU driver after move to Gen12

Signed-off-by: Vyacheslav N Klochkov <vyacheslav.n.klochkov@intel.com>

Author: v-klochkov

sycl/test-e2e/ESIMD/api/slm_gather_scatter_heavy.cpp

@@ -444,9 +444,10 @@ int main() {
   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".
-  if (!esimd_test::minLinuxDriver(q, "1.3.26816") ||
-      !esimd_test::minWindowsDriver(q, "1.3.26816")) {
+  // 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;
   }

sycl/test-e2e/ESIMD/esimd_test_utils.hpp

@@ -621,26 +621,78 @@ template <typename... ArgT> void printTestLabel(queue Q, ArgT &&...Args) {
   std::cout << std::endl;
 }
 
-// Not Linux: return true;
-// Linux:     return true iff GPU driver version >= MinDriver
-bool minLinuxDriver(queue Q, std::string MinDriver) {
-#ifdef __SYCL_RT_OS_LINUX
-  auto Driver = Q.get_device().get_info<sycl::info::device::driver_version>();
-  return Driver >= MinDriver;
-#else
-  return true;
-#endif
-}
+enum GPUDriverOS { Linux = 1, Windows = 2, LinuxAndWindows = 3 };
+
+/// This function returns true if it can detect the level-zero or opencl
+/// GPU driver and can determine that the current driver is same or newer
+/// than the one passed in \p RequiredVersion or \p WinOpenCLRequiredVersion.
+///
+/// Below are how driver versions look like:
+///   Linux/L0:       [1.3.26370]
+///   Linux/opencl:   [23.22.26370.18]
+///   Windows/L0:     [1.3.26370]
+///   Windows/opencl: [31.0.101.4502]
+///
+/// This function uses only the part of the driver identification:
+///   - the second half of the driver id on win/opencl, e.g. 101.4502";
+///   - the 5-digit id for 3 other platforms, e.g. 26370.
+///
+/// Note: For the previous & new driver version and their release dates
+/// for win/opencl see the link:
+/// https://www.intel.com/content/www/us/en/download/726609/intel-arc-iris-xe-graphics-whql-windows.html
+bool isGPUDriverGE(queue Q, GPUDriverOS OSCheck, std::string RequiredVersion,
+                   std::string WinOpenCLRequiredVersion = "") {
+  auto Dev = Q.get_device();
+  if (!Dev.is_gpu())
+    return false;
 
-// Not Windows: return true;
-// Windows:     return true iff GPU driver version >= MinDriver
-bool minWindowsDriver(queue Q, std::string MinDriver) {
-#ifdef __SYCL_RT_OS_WINDOWS
-  auto Driver = Q.get_device().get_info<sycl::info::device::driver_version>();
-  return Driver >= MinDriver;
-#else
-  return true;
+  bool IsLinux = false;
+#if defined(__SYCL_RT_OS_LINUX)
+  IsLinux = true;
+#elif !defined(__SYCL_RT_OS_WINDOWS)
+  return false;
 #endif
+
+  // A and B must have digits at the same positions.
+  // Otherwise, A and B symbols must be equal, e.g. both be equal to '.'.
+  auto verifyDriverVersionFormat = [](const std::string &A,
+                                      const std::string &B) {
+    if (A.size() != B.size())
+      throw std::runtime_error(
+          "Inconsistent expected & actual driver versions");
+    for (int I = 0; I < A.size(); I++) {
+      if ((A[I] >= '0' && A[I] <= '9' && !(B[I] >= '0' && B[I] <= '9')) &&
+          A[I] != B[I])
+        throw std::runtime_error(
+            "Inconsistent expected & actual driver versions");
+    }
+  };
+
+  auto BE = Q.get_backend();
+  int Length = 5;              // extract 5 digits for 3 or 4 platforms
+  int Start = 4;               // start of the driver id for 2 of 4 platforms
+  if (BE == backend::opencl) { // opencl has less-standard versioning
+    if (IsLinux) {
+      Start = 6;
+    } else {
+      Start = 5;
+      Length = 8;
+      RequiredVersion = WinOpenCLRequiredVersion;
+    }
+  }
+
+  bool IsGE = true;
+  if (IsLinux && (OSCheck & GPUDriverOS::Linux) ||
+      !IsLinux && (OSCheck & GPUDriverOS::Windows)) {
+    auto CurrentVersion = Dev.get_info<sycl::info::device::driver_version>();
+    CurrentVersion = CurrentVersion.substr(Start, Length);
+    verifyDriverVersionFormat(CurrentVersion, RequiredVersion);
+    std::cout << "RequiredVersion = " << RequiredVersion << ", Start=" << Start
+              << ", Length=" << Length << std::endl;
+    std::cout << "CurrentVersion = " << CurrentVersion << std::endl;
+    IsGE &= CurrentVersion >= RequiredVersion;
+  }
+  return IsGE;
 }
 
 } // namespace esimd_test

sycl/test-e2e/ESIMD/local_accessor_block_load_store.cpp

@@ -114,7 +114,9 @@ int main() {
   Pass &= test<int, 16, Align16>(Q);
   Pass &= test<float, 16, Align16>(Q);
 
-  if (Dev.has(aspect::fp16) && esimd_test::minLinuxDriver(Q, "1.3.26032"))
+  if (Dev.has(aspect::fp16) &&
+      esimd_test::isGPUDriverGE(Q, esimd_test::GPUDriverOS::LinuxAndWindows,
+                                "26032", "101.4502"))
     Pass &= test<sycl::half, 16, Align16>(Q);
 
   // Check SLM load/store with vector size that is not power of 2

sycl/test-e2e/ESIMD/slm_block_load_store.cpp

@@ -8,9 +8,6 @@
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 //
-// TODO: Enable the test when GPU driver is ready/fixed.
-// XFAIL: opencl || windows || gpu-intel-pvc
-// TODO: add support for local_accessors to esimd_emulator.
 // UNSUPPORTED: esimd_emulator
 
 // This test verifies usage of slm_block_load() and slm_block_store().
@@ -76,14 +73,15 @@ template <typename T, int VL, int Align = 16> bool test(queue Q) {
   }
 
   bool Pass = true;
+  int NumPrintedErorrs = 0;
   for (int I = 0; I < GlobalRange * VL; I++) {
     int GID = I / VL;
     int LID = GID % LocalRange;
     int VecElementIndex = I % VL;
 
     T Expected = GID * 100 + VecElementIndex;
     T Computed = Out[I];
-    if (Computed != Expected) {
+    if (Computed != Expected && ++NumPrintedErorrs < 16) {
       std::cout << "Error: Out[" << I << "]:" << Computed << " != " << Expected
                 << ":[expected]" << std::endl;
       Pass = false;
@@ -100,6 +98,15 @@ 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;
@@ -108,7 +115,7 @@ int main() {
   Pass &= test<int, 16, Align16>(Q);
   Pass &= test<float, 16, Align16>(Q);
 
-  if (Dev.has(aspect::fp16) && esimd_test::minLinuxDriver(Q, "1.3.26032"))
+  if (Dev.has(aspect::fp16))
     Pass &= test<sycl::half, 16, Align16>(Q);
 
   // Check SLM load/store with alignment smaller than 16-bytes.

sycl/test-e2e/InvokeSimd/Regression/slm_load_store.cpp

@@ -8,6 +8,8 @@
  * Test check basic support of local memory access in invoke_simd.
  */
 
+#include "../invoke_simd_utils.hpp"
+
 #include <sycl/ext/intel/esimd.hpp>
 #include <sycl/ext/oneapi/experimental/invoke_simd.hpp>
 #include <sycl/sycl.hpp>
@@ -76,6 +78,15 @@ int main(void) {
   auto Dev = Q.get_device();
   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;
 

sycl/test-e2e/InvokeSimd/invoke_simd_utils.hpp

@@ -0,0 +1,83 @@
+#pragma once
+
+#include <string>
+#include <sycl/ext/intel/esimd.hpp>
+#include <sycl/sycl.hpp>
+
+using namespace sycl::ext::oneapi::experimental;
+using namespace sycl;
+namespace esimd = sycl::ext::intel::esimd;
+
+enum GPUDriverOS { Linux = 1, Windows = 2, LinuxAndWindows = 3 };
+
+/// This function returns true if it can detect the level-zero or opencl
+/// GPU driver and can determine that the current driver is same or newer
+/// than the one passed in \p RequiredVersion or \p WinOpenCLRequiredVersion.
+///
+/// Below are how driver versions look like:
+///   Linux/L0:       [1.3.26370]
+///   Linux/opencl:   [23.22.26370.18]
+///   Windows/L0:     [1.3.26370]
+///   Windows/opencl: [31.0.101.4502]
+///
+/// This function uses only the part of the driver identification:
+///   - the second half of the driver id on win/opencl, e.g. 101.4502";
+///   - the 5-digit id for 3 other platforms, e.g. 26370.
+///
+/// Note: For the previous & new driver version and their release dates
+/// for win/opencl see the link:
+/// https://www.intel.com/content/www/us/en/download/726609/intel-arc-iris-xe-graphics-whql-windows.html
+bool isGPUDriverGE(queue Q, GPUDriverOS OSCheck, std::string RequiredVersion,
+                   std::string WinOpenCLRequiredVersion = "") {
+  auto Dev = Q.get_device();
+  if (!Dev.is_gpu())
+    return false;
+
+  bool IsLinux = false;
+#if defined(__SYCL_RT_OS_LINUX)
+  IsLinux = true;
+#elif !defined(__SYCL_RT_OS_WINDOWS)
+  return false;
+#endif
+
+  // A and B must have digits at the same positions.
+  // Otherwise, A and B symbols must be equal, e.g. both be equal to '.'.
+  auto verifyDriverVersionFormat = [](const std::string &A,
+                                      const std::string &B) {
+    if (A.size() != B.size())
+      throw std::runtime_error(
+          "Inconsistent expected & actual driver versions");
+    for (int I = 0; I < A.size(); I++) {
+      if ((A[I] >= '0' && A[I] <= '9' && !(B[I] >= '0' && B[I] <= '9')) &&
+          A[I] != B[I])
+        throw std::runtime_error(
+            "Inconsistent expected & actual driver versions");
+    }
+  };
+
+  auto BE = Q.get_backend();
+  int Length = 5;              // extract 5 digits for 3 or 4 platforms
+  int Start = 4;               // start of the driver id for 2 of 4 platforms
+  if (BE == backend::opencl) { // opencl has less-standard versioning
+    if (IsLinux) {
+      Start = 6;
+    } else {
+      Start = 5;
+      Length = 8;
+      RequiredVersion = WinOpenCLRequiredVersion;
+    }
+  }
+
+  bool IsGE = true;
+  if (IsLinux && (OSCheck & GPUDriverOS::Linux) ||
+      !IsLinux && (OSCheck & GPUDriverOS::Windows)) {
+    auto CurrentVersion = Dev.get_info<sycl::info::device::driver_version>();
+    CurrentVersion = CurrentVersion.substr(Start, Length);
+    verifyDriverVersionFormat(CurrentVersion, RequiredVersion);
+    std::cout << "RequiredVersion = " << RequiredVersion << ", Start=" << Start
+              << ", Length=" << Length << std::endl;
+    std::cout << "CurrentVersion = " << CurrentVersion << std::endl;
+    IsGE &= CurrentVersion >= RequiredVersion;
+  }
+  return IsGE;
+}