[SYCL] Introduce backend macros (#4636)

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

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

Author: vladimirlaz

buildbot/configure.py

@@ -29,7 +29,7 @@ def do_configure(args):
     libclc_targets_to_build = ''
     libclc_gen_remangled_variants = 'OFF'
     sycl_build_pi_cuda = 'OFF'
-    sycl_build_pi_esimd_cpu = 'OFF'
+    sycl_build_pi_esimd_emulator = 'OFF'
     sycl_build_pi_hip = 'OFF'
     sycl_build_pi_hip_platform = 'AMD'
     sycl_clang_extra_flags = ''
@@ -53,7 +53,7 @@ def do_configure(args):
         llvm_targets_to_build = 'ARM;AArch64'
 
     if args.enable_esimd_cpu_emulation:
-        sycl_build_pi_esimd_cpu = 'ON'
+        sycl_build_pi_esimd_emulator = 'ON'
 
     if args.cuda or args.hip:
         llvm_enable_projects += ';libclc'
@@ -128,7 +128,7 @@ def do_configure(args):
         "-DBUILD_SHARED_LIBS={}".format(llvm_build_shared_libs),
         "-DSYCL_ENABLE_XPTI_TRACING={}".format(sycl_enable_xpti_tracing),
         "-DLLVM_ENABLE_LLD={}".format(llvm_enable_lld),
-        "-DSYCL_BUILD_PI_ESIMD_CPU={}".format(sycl_build_pi_esimd_cpu),
+        "-DSYCL_BUILD_PI_ESIMD_EMULATOR={}".format(sycl_build_pi_esimd_emulator),
         "-DXPTI_ENABLE_WERROR={}".format(xpti_enable_werror),
         "-DSYCL_CLANG_EXTRA_FLAGS={}".format(sycl_clang_extra_flags)
     ]

sycl/.gitignore

@@ -1 +1,2 @@
 include/CL/sycl/version.hpp
+include/CL/sycl/feature_test.hpp

sycl/CMakeLists.txt

@@ -100,6 +100,8 @@ set(sycl_plugin_dir ${CMAKE_CURRENT_SOURCE_DIR}/plugins)
 string(TIMESTAMP __SYCL_COMPILER_VERSION "%Y%m%d")
 set(version_header "${sycl_inc_dir}/CL/sycl/version.hpp")
 configure_file("${version_header}.in" "${version_header}")
+set(feature_header "${sycl_inc_dir}/CL/sycl/feature_test.hpp")
+configure_file("${feature_header}.in" "${feature_header}")
 
 # This is workaround to detect changes (add or modify) in subtree which
 # are not detected by copy_directory command.
@@ -310,7 +312,7 @@ endif()
 # TODO : Remove 'if (NOT MSVC)' when CM_EMU supports Windows
 # environment
 if (NOT MSVC)
-  if (SYCL_BUILD_PI_ESIMD_CPU)
+  if (SYCL_BUILD_PI_ESIMD_EMULATOR)
     list(APPEND SYCL_TOOLCHAIN_DEPLOY_COMPONENTS pi_esimd_cpu libcmrt-headers)
     if (MSVC)
       list(APPEND SYCL_TOOLCHAIN_DEPLOY_COMPONENTS libcmrt-libs libcmrt-dlls)

sycl/include/CL/sycl/backend_types.hpp

@@ -26,7 +26,9 @@ enum class backend : char {
       ext_oneapi_level_zero,
   cuda = 3,
   all = 4,
-  esimd_cpu = 5,
+  ext_intel_esimd_emulator = 5,
+  esimd_cpu __SYCL2020_DEPRECATED("use 'ext_oneapi_esimd_emulator' instead") =
+      ext_intel_esimd_emulator,
   hip = 6,
 };
 
@@ -55,8 +57,8 @@ inline std::ostream &operator<<(std::ostream &Out, backend be) {
   case backend::cuda:
     Out << "cuda";
     break;
-  case backend::esimd_cpu:
-    Out << "esimd_cpu";
+  case backend::ext_intel_esimd_emulator:
+    Out << "ext_intel_esimd_emulator";
     break;
   case backend::hip:
     Out << "hip";

sycl/include/CL/sycl/feature_test.hpp renamed to sycl/include/CL/sycl/feature_test.hpp.in

@@ -10,6 +10,7 @@
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 
+#define SYCL_BACKEND_OPENCL 1
 // Feature test macro definitions
 
 // TODO: Move these feature-test macros to compiler driver.
@@ -29,6 +30,18 @@ namespace sycl {
 #define SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY 1
 #define SYCL_EXT_INTEL_USM_ADDRESS_SPACES 1
 #define SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO 1
+#cmakedefine01 SYCL_BUILD_PI_CUDA
+#if SYCL_BUILD_PI_CUDA
+#define SYCL_EXT_ONEAPI_BACKEND_CUDA 1
+#endif
+#cmakedefine01 SYCL_BUILD_PI_ESIMD_EMULATOR
+#if SYCL_BUILD_PI_ESIMD_EMULATOR
+#define SYCL_EXT_INTEL_BACKEND_ESIMD_EMULATOR 1
+#endif
+#cmakedefine01 SYCL_BUILD_PI_HIP
+#if SYCL_BUILD_PI_HIP
+#define SYCL_EXT_ONEAPI_BACKEND_HIP 1
+#endif
 
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)

sycl/include/sycl/ext/intel/experimental/esimd/emu/detail/esimdcpu_device_interface.hpp

@@ -75,7 +75,9 @@ ESIMDDeviceInterface *getESIMDDeviceInterface() {
   // tight loop)
   void *PIOpaqueData = nullptr;
 
-  PIOpaqueData = getPluginOpaqueData<cl::sycl::backend::esimd_cpu>(nullptr);
+  PIOpaqueData =
+      getPluginOpaqueData<cl::sycl::backend::ext_oneapi_esimd_emulator>(
+          nullptr);
 
   ESIMDEmuPluginOpaqueData *OpaqueData =
       reinterpret_cast<ESIMDEmuPluginOpaqueData *>(PIOpaqueData);

sycl/plugins/CMakeLists.txt

@@ -18,7 +18,7 @@ add_subdirectory(level_zero)
 # TODO : Remove 'if (NOT MSVC)' when CM_EMU supports Windows
 # environment
 if (NOT MSVC)
-  if (SYCL_BUILD_PI_ESIMD_CPU)
+  if (SYCL_BUILD_PI_ESIMD_EMULATOR)
     add_subdirectory(esimd_cpu)
   endif()
 endif()

sycl/source/detail/pi.cpp

@@ -74,7 +74,7 @@ void *getPluginOpaqueData(void *OpaqueDataParam) {
 }
 
 template __SYCL_EXPORT void *
-getPluginOpaqueData<cl::sycl::backend::esimd_cpu>(void *);
+getPluginOpaqueData<cl::sycl::backend::ext_intel_esimd_emulator>(void *);
 
 namespace pi {
 
@@ -504,7 +504,8 @@ template <backend BE> const plugin &getPlugin() {
 
 template __SYCL_EXPORT const plugin &getPlugin<backend::opencl>();
 template __SYCL_EXPORT const plugin &getPlugin<backend::level_zero>();
-template __SYCL_EXPORT const plugin &getPlugin<backend::esimd_cpu>();
+template __SYCL_EXPORT const plugin &
+getPlugin<backend::ext_intel_esimd_emulator>();
 
 // Report error and no return (keeps compiler from printing warnings).
 // TODO: Probably change that to throw a catchable exception,

sycl/test/basic_tests/exceptions-SYCL-2020.cpp

@@ -91,8 +91,10 @@ int main() {
   sycl::backend_traits<sycl::backend::cuda>::errc someCUDAErrCode{EC};
   sycl::errc_for<sycl::backend::cuda> anotherCUDAErrCode{EC};
   assert(someCUDAErrCode == anotherCUDAErrCode);
-  sycl::backend_traits<sycl::backend::esimd_cpu>::errc someESIMDErrCode{EC};
-  sycl::errc_for<sycl::backend::esimd_cpu> anotherESIMDErrCode{EC};
+  sycl::backend_traits<sycl::backend::ext_intel_esimd_emulator>::errc
+      someESIMDErrCode{EC};
+  sycl::errc_for<sycl::backend::ext_intel_esimd_emulator> anotherESIMDErrCode{
+      EC};
   assert(someESIMDErrCode == anotherESIMDErrCode);
   sycl::backend_traits<sycl::backend::hip>::errc someHIPErrCode{EC};
   sycl::errc_for<sycl::backend::hip> anotherHIPErrCode{EC};

sycl/test/extensions/macro.cpp

@@ -5,11 +5,23 @@
 #include <CL/sycl.hpp>
 #include <iostream>
 int main() {
+#if SYCL_BACKEND_OPENCL == 1
+  std::cout << "SYCL_BACKEND_OPENCL=1" << std::endl;
+#else
+  std::cerr << "SYCL_BACKEND_OPENCL!=1" << std::endl;
+  exit(1);
+#endif
 #if SYCL_EXT_ONEAPI_SUB_GROUP_MASK == 1
   std::cout << "SYCL_EXT_ONEAPI_SUB_GROUP_MASK=1" << std::endl;
 #else
   std::cerr << "SYCL_EXT_ONEAPI_SUB_GROUP_MASK!=1" << std::endl;
   exit(1);
+#endif
+#if SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO == 1
+  std::cout << "SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO=1" << std::endl;
+#else
+  std::cerr << "SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO!=1" << std::endl;
+  exit(1);
 #endif
   exit(0);
 }

sycl/test/extensions/macro_cuda.cpp

@@ -0,0 +1,15 @@
+// This test checks presence of macros for available extensions.
+// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %t.out
+// REQUIRES: cuda_be
+#include <CL/sycl.hpp>
+#include <iostream>
+int main() {
+#if SYCL_EXT_ONEAPI_BACKEND_CUDA == 1
+  std::cout << "SYCL_EXT_ONEAPI_BACKEND_CUDA=1" << std::endl;
+#else
+  std::cerr << "SYCL_EXT_ONEAPI_BACKEND_CUDA!=1" << std::endl;
+  exit(1);
+#endif
+  exit(0);
+}

sycl/test/extensions/macro_esimd_emulator.cpp

@@ -0,0 +1,15 @@
+// This test checks presence of macros for available extensions.
+// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %t.out
+// REQUIRES: esimd_emulator_be
+#include <CL/sycl.hpp>
+#include <iostream>
+int main() {
+#if SYCL_EXT_INTEL_BACKEND_ESIMD_EMULATOR == 1
+  std::cout << "SYCL_EXT_INTEL_BACKEND_ESIMD_EMULATOR=1" << std::endl;
+#else
+  std::cerr << "SYCL_EXT_INTEL_BACKEND_ESIMD_EMULATOR!=1" << std::endl;
+  exit(1);
+#endif
+  exit(0);
+}

sycl/test/extensions/macro_hip.cpp

@@ -0,0 +1,15 @@
+// This test checks presence of macros for available extensions.
+// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: %t.out
+// REQUIRES: hip_be
+#include <CL/sycl.hpp>
+#include <iostream>
+int main() {
+#if SYCL_EXT_ONEAPI_BACKEND_HIP == 1
+  std::cout << "SYCL_EXT_ONEAPI_BACKEND_HIP=1" << std::endl;
+#else
+  std::cerr << "SYCL_EXT_ONEAPI_BACKEND_HIP!=1" << std::endl;
+  exit(1);
+#endif
+  exit(0);
+}

sycl/test/lit.cfg.py

@@ -100,6 +100,15 @@
 lit_config.note("Triple: {}".format(triple))
 config.substitutions.append( ('%sycl_triple',  triple ) )
 
+if config.cuda_be == "ON":
+    config.available_features.add('cuda_be')
+
+if config.hip_be == "ON":
+    config.available_features.add('hip_be')
+
+if config.esimd_emulator_be == "ON":
+    config.available_features.add('esimd_emulator_be')
+
 if triple == 'nvptx64-nvidia-cuda':
     config.available_features.add('cuda')
 

sycl/test/lit.site.cfg.py.in

@@ -24,6 +24,9 @@ config.llvm_enable_projects = "@LLVM_ENABLE_PROJECTS@"
 config.sycl_threads_lib = '@SYCL_THREADS_LIB@'
 config.sycl_use_libcxx = '@SYCL_USE_LIBCXX@'
 config.extra_environment = lit_config.params.get("extra_environment", "@LIT_EXTRA_ENVIRONMENT@")
+config.cuda_be = '@SYCL_BUILD_PI_CUDA@'
+config.esimd_emulator_be = '@SYCL_BUILD_PI_ESIMD_EMULATOR@'
+config.hip_be = '@SYCL_BUILD_PI_HIP@'
 
 import lit.llvm
 lit.llvm.initialize(lit_config, config)

sycl/test/warnings/sycl_2020_deprecations.cpp

@@ -158,6 +158,10 @@ int main() {
   auto LevelZeroBackend = sycl::backend::level_zero;
   (void)LevelZeroBackend;
 
+  // expected-warning@+1{{'esimd_cpu' is deprecated: use 'ext_oneapi_esimd_emulator' instead}}
+  auto ESIMDCPUBackend = sycl::backend::esimd_cpu;
+  (void)ESIMDCPUBackend;
+
   sycl::half Val = 1.0f;
   // expected-warning@+1{{'bit_cast<unsigned short, sycl::detail::half_impl::half>' is deprecated: use 'sycl::bit_cast' instead}}
   auto BitCastRes = sycl::detail::bit_cast<unsigned short>(Val);