More meaningful info::device::version for Intel GPUs

[al42and]

Is your feature request related to a problem? Please describe

We want an easy way to check the architecture of the device: Gen11, Xe-HPC, Xe-HPG, and so on, to select the best kernels based on earlier empirical tuning.

Per spec, sycl::device::get_info<sycl::info::device::version> "Returns a backend-defined device version."

AMD and NVIDIA plugins use sycl::info::device::version field to return the device architecture (like 8.6 or gfx90a, respectively), which looks like a reasonable interpretation of the standard.

However, OpenCL and LevelZero backends return the API versions (OpenCL 3.0 NEO or 1.3):

llvm/sycl/plugins/unified_runtime/ur/adapters/level_zero/device.cpp

Lines 279 to 283 in d430bef

	case UR_DEVICE_INFO_DRIVER_VERSION:
	case UR_DEVICE_INFO_BACKEND_RUNTIME_VERSION:
	return ReturnValue(Device->Platform->ZeDriverVersion.c_str());
	case UR_DEVICE_INFO_VERSION:
	return ReturnValue(Device->Platform->ZeDriverApiVersion.c_str());

This piece of information, IMO, better belongs in either platform::version or device::backend_version descriptors, while device::version should return something more device-specific.

Describe the solution you would like

Using releaseAcronyms or deviceAcronyms from the AOT compiler (https://github.com/intel/compute-runtime/blob/master/third_party/aot_config_headers/platforms.h) sounds like a good idea.

Describe alternatives you have considered

• A concatenation of family, release, and device, like xe:xe-hpg:dg2-g10.
• Hex value of PRODUCT_CONFIG (from the same file)
• Decoded value of PRODUCT_CONFIG (e.g., 11.0.0 for ICL / 0x02c00000)
• PCIe device ID (was previously reported as part of the device::name)

[dm-vodopyanov]

Hi @al42and, for identifying a device architecture, you can use sycl_ext_oneapi_device_architecture extension, it was recently implemented here in intel/llvm.

There are two host APIs:

• (1) New device descriptor: https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/experimental/sycl_ext_oneapi_device_architecture.asciidoc#new-device-descriptor
• (2) New member function of device class: https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/experimental/sycl_ext_oneapi_device_architecture.asciidoc#new-member-function-of-device-class

Example:

namespace syclex = sycl::ext::oneapi::experimental;

sycl::queue q;  
sycl::device dev = q.get_device();

// (1)
syclex::architecture arch = 
  dev.get_info<syclex::info::device::architecture>();

// (2)
bool is_pvc = dev.ext_oneapi_architecture_is(
  syclex::architecture::intel_gpu_pvc);

The full list of architectures is here: https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/experimental/sycl_ext_oneapi_device_architecture.asciidoc#new-enumeration-of-architectures, currently only Intel GPU architectures are supported in these host APIs.

[al42and]

Hi @dm-vodopyanov. That's a good option, but has some limitations. Perhaps my issue can be re-targeted as a wishlist for this extension. In no particular order:

1. Logging. No support for operator<<(std::ostream&, const syclex::architecture).
2. Comparison: No easy way to check things like "pre-Xe". The enum value can be converted to integer and compared like this, but that's not supported, is it?
3. Compatibility: a new Intel device (supported by compute-runtime) with older IntelLLVM (lacking the corresponding enum value). As long as it ingests SPIR-V, there is little need to limit ourselves to the known devices, is there?
4. Compatibility: supporting older compilers/runtime from the application code. As new hardware is released, the extension is updted, and the already long switch/case checking the hardware versions will get interspersed with PP macros checking for the availability of each enum value.

So, for a future Meteor Lake architecture (already somewhat supported in compute-runtime), instead of

bool is_pvc = dev.ext_oneapi_architecture_is(syclex::architecture::intel_gpu_pvc);

we will have something like:

#ifdef __SYCL_TARGET_INTEL_GPU_MTL__
bool is_mtl = dev.ext_oneapi_architecture_is(syclex::architecture::intel_gpu_mtl);
#else
bool is_mtl = false; // Or "true"; perhaps the compiler is slightly too old
#endif

[dm-vodopyanov]

Tagging @gmlueck for the awareness about wishlist for sycl_ext_oneapi_device_architecture's host APIs

[jinz2014]

users may see the warning message:

device_architecture_on_host.cpp:21:11: warning: 55 enumeration values not handled in switch: 'intel_gpu_bdw', 'intel_gpu_skl', 'intel_gpu_kbl'... [-Wswitch]

It is good to know the usage of the feature.