[SYCL] Introduce backend macros

[vladimirlaz]

According to SYCL spec: backend macro should be specified for all supported
backends (defined in the Khronos group). It also refers to extension macros
(aka feature test macro) for other backends introduced by a vendor. Khronos
group defines only OpenCL backend and other backends supported by DPC++
should be treated as extensions.

The following macros are defined when the compiler is built with the
corresponding backend. OpenCL and Level_Zero backends are always
available, others are enabled on build time:

• OpenCL backed: SYCL_BACKEND_OPENCL
• Level Zero backend: SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO ()
• CUDA backend: SYCL_EXT_ONEAPI_BACKEND_CUDA
• ESIMD_CPU backend: SYCL_EXT_INTEL_BACKEND_ESIMD_EMULATOR
• HIP backend: SYCL_EXT_ONEAPI_BACKEND_HIP
  Extra libs may be required to link DPC++ application and run it.

backend::esimd_cpu is renamed to backend::ext_intel_esimd_emulator. Old
naming was marked as deprecated.

[s-kanaev]

LGTM

[vladimirlaz]

@gmlueck, @jbrodman, @Pennycook could you please have a look?

[gmlueck]

It looks like the backends cuda, esimd_cpu, and hip are only conditionally supported if they are enabled in the "configure.py" script. Therefore, shouldn't these macros only be conditionally defined?

[vladimirlaz]

cuda, esimd_cpu, and hip are only conditionally supported if they a

We do not remove support for BEs in headers depending on the build configuration. So we headers contain all BE specific code. that is why I decided to add these macros unconditionally.

[gmlueck]

We do not remove support for BEs in headers depending on the build configuration. So we headers contain all BE specific code. that is why I decided to add these macros unconditionally.

Conformant SYCL code might look something like this:

void conditionallyDoCudaThing(queue q) {
#ifdef SYCL_EXT_ONEAPI_BACKEND_CUDA
  if (q.get_backend() == backend::cuda) {
    q.single_task([=] {
      // Calls that are legal in a CUDA kernel
    });
  }
#endif
}

What will happen if DPC++ is configured without CUDA support? Will that code still compile and run without errors? Obviously, the kernel won't be submitted in this case because q.get_backend() == backend::cuda will evaluate to false. However, the compiler must not report any errors and the runtime must not throw any exceptions or fail even though DPC++ is not configured with CUDA support. Will this be the case even though SYCL_EXT_ONEAPI_BACKEND_CUDA is defined?

[vladimirlaz]

We do not remove support for BEs in headers depending on the build configuration. So we headers contain all BE specific code. that is why I decided to add these macros unconditionally.

Conformant SYCL code might look something like this:

void conditionallyDoCudaThing(queue q) {
#ifdef SYCL_EXT_ONEAPI_BACKEND_CUDA
  if (q.get_backend() == backend::cuda) {
    q.single_task([=] {
      // Calls that are legal in a CUDA kernel
    });
  }
#endif
}

What will happen if DPC++ is configured without CUDA support? Will that code still compile and run without errors? Obviously, the kernel won't be submitted in this case because q.get_backend() == backend::cuda will evaluate to false. However, the compiler must not report any errors and the runtime must not throw any exceptions or fail even though DPC++ is not configured with CUDA support. Will this be the case even though SYCL_EXT_ONEAPI_BACKEND_CUDA is defined?

For CUDA support where was the explicit request to make DPC++ ready for CUDA (implemented in #3811). If user drop have lib_pi_cud.so and libclc-builtin libraries to the regular compiler it can be used to work with CUDA BE.

• @pvchupin

[gmlueck]

For CUDA support where was the explicit request to make DPC++ ready for CUDA (implemented in #3811). If user drop have lib_pi_cud.so and libclc-builtin libraries to the regular compiler it can be used to work with CUDA BE.

Let's consider the DPCPP production build. I think CUDA support is not configured for that build, correct? I presume the libraries you mention are also not included in that build. Will the example code I show above compile without errors in that case?

[vladimirlaz]

Let's consider the DPCPP production build. I think CUDA support is not configured for that build, correct? I presume the libraries you mention are also not included in that build. Will the example code I show above compile without errors in that case?

It will be successfully compiled. On the link step there will be an error due to missed built-in libraries.

[gmlueck]

It will be successfully compiled. On the link step there will be an error due to missed built-in libraries.

This is the sort of thing I was concerned about. This would be solved if we defined the macros conditionally based on the source configuration. Is that difficult?

[vladimirlaz]

It will be successfully compiled. On the link step there will be an error due to missed built-in libraries.

This is the sort of thing I was concerned about. This would be solved if we defined the macros conditionally based on the source configuration. Is that difficult?

We can enable macros conditionally during the build of the compiler (there are no technical difficulties here). But this approach contradicts the request implemented in the scope of #3811.
@pvchupin does the support of CUDA backend is still required in the regular build (when a customer drops libpi_cuda.so and libclc in the compiler)?

[pvchupin]

It will be successfully compiled. On the link step there will be an error due to missed built-in libraries.

This is the sort of thing I was concerned about. This would be solved if we defined the macros conditionally based on the source configuration. Is that difficult?

We can enable macros conditionally during the build of the compiler (there are no technical difficulties here). But this approach contradicts the request implemented in the scope of #3811. @pvchupin does the support of CUDA backend is still required in the regular build (when a customer drops libpi_cuda.so and libclc in the compiler)?

I don't see any contradiction. Let's chat offline...

[gmlueck]

We were discussing the name of the SYCL_EXT_INTEL_BACKEND_ESIMD_CPU macro offline with @kbobrovs, and we think the current name is confusing. Suggested alternatives are:

• SYCL_EXT_INTEL_BACKEND_ESIMD_EMULATOR, or
• SYCL_EXT_INTEL_BACKEND_ESIMD_EMU

Can we change the name to one of those? Possibly, it makes sense to rename the enumerator in sycl::backend also to ext_intel_esimd_emulator (or ext_intel_esimd_emu) in this PR also, so they are in sync?

[vladimirlaz]

We were discussing the name of the SYCL_EXT_INTEL_BACKEND_ESIMD_CPU macro offline with @kbobrovs, and we think the current name is confusing. Suggested alternatives are:

• SYCL_EXT_INTEL_BACKEND_ESIMD_EMULATOR, or
• SYCL_EXT_INTEL_BACKEND_ESIMD_EMU

Can we change the name to one of those? Possibly, it makes sense to rename the enumerator in sycl::backend also to ext_intel_esimd_emulator (or ext_intel_esimd_emu) in this PR also, so they are in sync?

I will apply suggested changes later this week

[kbobrovs]

Let's use full SYCL_EXT_INTEL_BACKEND_ESIMD_EMULATOR for the ESIMD emulator macro.

[vladimirlaz]

@kbobrovs, @gmlueck, could you please review?

[gmlueck]

LGTM

[bader]

buildbot/configure.py looks good to me (not sure if ESIMD_CPU -> EMULATOR renaming directly related to backend macos though).