[SYCL][CUDA][HIP] Fix PI version reporting

sycl/plugins/cuda/pi_cuda.cpp

@@ -1708,8 +1708,22 @@ pi_result cuda_piDeviceGetInfo(pi_device device, pi_device_info param_name,
                    device->get_reference_count());
   }
   case PI_DEVICE_INFO_VERSION: {
+    std::stringstream s;
+    int major;
+    sycl::detail::pi::assertion(
+        cuDeviceGetAttribute(&major,
+                             CU_DEVICE_ATTRIBUTE_COMPUTE_CAPABILITY_MAJOR,
+                             device->get()) == CUDA_SUCCESS);
+    s << major;
+
+    int minor;
+    sycl::detail::pi::assertion(
+        cuDeviceGetAttribute(&minor,
+                             CU_DEVICE_ATTRIBUTE_COMPUTE_CAPABILITY_MAJOR,
+                             device->get()) == CUDA_SUCCESS);
+    s << "." << minor;
     return getInfo(param_value_size, param_value, param_value_size_ret,
-                   "PI 0.0");
+                   s.str().c_str());
   }
   case PI_DEVICE_INFO_OPENCL_C_VERSION: {
     return getInfo(param_value_size, param_value, param_value_size_ret, "");

sycl/plugins/hip/pi_hip.cpp

@@ -1663,8 +1663,21 @@ pi_result hip_piDeviceGetInfo(pi_device device, pi_device_info param_name,
                    device->get_reference_count());
   }
   case PI_DEVICE_INFO_VERSION: {
+    std::stringstream s;
+
+    hipDeviceProp_t props;
+    sycl::detail::pi::assertion(hipGetDeviceProperties(&props, device->get()) ==
+                                hipSuccess);
+#if defined(__HIP_PLATFORM_NVIDIA__)
+    s << props.major << "." << props.minor;
+#elif defined(__HIP_PLATFORM_AMD__)
+    s << props.gcnArchName;
+#else
+#error("Must define exactly one of __HIP_PLATFORM_AMD__ or __HIP_PLATFORM_NVIDIA__");
+#endif
+
     return getInfo(param_value_size, param_value, param_value_size_ret,
-                   "PI 0.0");
+                   s.str().c_str());
   }
   case PI_DEVICE_INFO_OPENCL_C_VERSION: {
     return getInfo(param_value_size, param_value, param_value_size_ret, "");

sycl/source/detail/device_info.hpp

@@ -205,27 +205,11 @@ struct get_device_info_impl<std::vector<info::fp_config>, Param> {
   }
 };
 
-// Specialization for OpenCL version, splits the string returned by OpenCL
+// Specialization for device version
 template <> struct get_device_info_impl<std::string, info::device::version> {
   static std::string get(RT::PiDevice dev, const plugin &Plugin) {
-    std::string result = get_device_info_string(
-        dev, PiInfoCode<info::device::version>::value, Plugin);
-
-    // Extract OpenCL version from the returned string.
-    // For example, for the string "OpenCL 2.1 (Build 0)"
-    // return '2.1'.
-    auto dotPos = result.find('.');
-    if (dotPos == std::string::npos)
-      return result;
-
-    auto leftPos = result.rfind(' ', dotPos);
-    if (leftPos == std::string::npos)
-      leftPos = 0;
-    else
-      leftPos++;
-
-    auto rightPos = result.find(' ', dotPos);
-    return result.substr(leftPos, rightPos - leftPos);
+    return get_device_info_string(dev, PiInfoCode<info::device::version>::value,
+                                  Plugin);
   }
 };
 

sycl/source/detail/error_handling/error_handling.cpp

@@ -55,7 +55,8 @@ void handleInvalidWorkGroupSize(const device_impl &DeviceImpl, pi_kernel Kernel,
   bool IsLevelZero = false;   // Backend is any OneAPI Level 0 version
   auto Backend = Platform.get_backend();
   if (Backend == sycl::backend::opencl) {
-    std::string VersionString = DeviceImpl.get_info<info::device::version>();
+    std::string VersionString =
+        DeviceImpl.get_info<info::device::version>().substr(7, 3);
     IsOpenCL = true;
     IsOpenCLV1x = (VersionString.find("1.") == 0);
     IsOpenCLVGE20 =

sycl/test-e2e/GroupAlgorithm/SYCL2020/support.h

@@ -11,11 +11,10 @@ bool isSupportedDevice(device D) {
 
   if (PlatformName.find("OpenCL") != std::string::npos) {
     std::string Version = D.get_info<info::device::version>();
-    size_t Offset = Version.find("OpenCL");
-    if (Offset == std::string::npos)
-      return false;
-    Version = Version.substr(Offset + 7, 3);
-    if (Version >= std::string("2.0"))
+
+    // Group collectives are mandatory in OpenCL 2.0 but optional in 3.0.
+    Version = Version.substr(7, 3);
+    if (Version >= "2.0" && Version < "3.0")
       return true;
   }
 

sycl/test-e2e/GroupAlgorithm/support.h

@@ -15,11 +15,10 @@ bool isSupportedDevice(device D) {
 
   if (PlatformName.find("OpenCL") != std::string::npos) {
     std::string Version = D.get_info<sycl::info::device::version>();
-    size_t Offset = Version.find("OpenCL");
-    if (Offset == std::string::npos)
-      return false;
-    Version = Version.substr(Offset + 7, 3);
-    if (Version >= std::string("2.0"))
+
+    // Group collectives are mandatory in OpenCL 2.0 but optional in 3.0.
+    Version = Version.substr(7, 3);
+    if (Version >= "2.0" && Version < "3.0")
       return true;
   }
 

sycl/test-e2e/SubGroup/helper.hpp

@@ -169,5 +169,19 @@ bool core_sg_supported(const device &Device) {
   auto Vec = Device.get_info<info::device::extensions>();
   if (std::find(Vec.begin(), Vec.end(), "cl_khr_subgroups") != std::end(Vec))
     return true;
-  return Device.get_info<info::device::version>() >= "2.1";
+
+  if (std::find(Vec.begin(), Vec.end(), "cl_intel_subgroups") != std::end(Vec))
+    return true;
+
+  if (Device.get_backend() == sycl::backend::opencl) {
+    // Extract the numerical version from the version string, OpenCL version
+    // string have the format "OpenCL <major>.<minor> <vendor specific data>".
+    std::string ver = Device.get_info<info::device::version>().substr(7, 3);
+
+    // cl_khr_subgroups was core in OpenCL 2.1 and 2.2, but went back to
+    // optional in 3.0
+    return ver >= "2.1" && ver < "3.0";
+  }
+
+  return false;
 }