AMD GPU reporting SPIR support but an error occurs on execution

SYCL development
1

When running computecpp_info on a machine with an AMD processor the output shows the AMD GPU is supported. This is also reported when running clinfo.

For example:

--------------------------------------------------------------------------------
Device 0:

  Device is supported                     : UNTESTED - Device not tested on this OS
  CL_DEVICE_NAME                          : gfx904
  CL_DEVICE_VENDOR                        : Advanced Micro Devices, Inc.
  CL_DRIVER_VERSION                       : 2639.5 (PAL,HSAIL)
  CL_DEVICE_TYPE                          : CL_DEVICE_TYPE_GPU
--------------------------------------------------------------------------------

When executing ComputeCpp compiled binaries for this GPU errors will occur including:

  • cl::sycl::compile_program_error from create_program_for_kernel_impl()
  • terminate called after throwing an instance of
    ‘cl::sycl::detail::exception_implementation<(cl::sycl::detail::exception_types)9,
    cl::sycl::detail::exception_implementation<(cl::sycl::detail::exception_types)6, cl::sycl::exception> >’
2

While some AMD drivers are reporting support for SPIR or SPIR-V, which is necessary to execute ComputeCpp generated binaries, they do not actually support the SPIR instruction sets and the binaries will fail to run.

3

I have reported this on AMD forums countless times. 7-8 drivers have gone by which still report the presence of this extension erronously. In case you wish to use ComputeCpp with AMD GPUs, 18.2.3 is the last driver to support SPIR.

Also do note, that 18.2.3 still shows the CPU device, but in fact it’s not functioning. Trying to compile any OpenCL kernel results in internal compiler errors. CPU runtime last worked sometime around 17.7, but that is too old for many newer GPUs, such as my RX 580. Should you want to join the “conversation” on AMD forums, you can do so [here].

1 Like