From 24571ce835feee154268da05046122fd198c57d1 Mon Sep 17 00:00:00 2001 From: Dmitriy Sobolev Date: Fri, 6 Dec 2024 21:58:58 +0000 Subject: [PATCH] Generalize the approach, introduce _ONEDPL_GENERIC_SYCL_LIBRARY macro --- .../dpcpp/parallel_backend_sycl_radix_sort.h | 2 +- .../dpcpp/parallel_backend_sycl_utils.h | 2 +- .../oneapi/dpl/pstl/hetero/dpcpp/sycl_defs.h | 65 ++++++++++--------- include/oneapi/dpl/pstl/utils.h | 2 +- 4 files changed, 38 insertions(+), 33 deletions(-) diff --git a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_radix_sort.h b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_radix_sort.h index b06ca8c9bea..d8b95a5808d 100644 --- a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_radix_sort.h +++ b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_radix_sort.h @@ -346,7 +346,7 @@ enum class __peer_prefix_algo template struct __peer_prefix_helper; -#define _ONEDPL_SYCL2020_SUBGROUP_BARRIER_PRESENT _ONEDPL_LIBSYCL_ZERO_OR_GE(50700) +#define _ONEDPL_SYCL2020_SUBGROUP_BARRIER_PRESENT (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50700) #if _ONEDPL_SYCL2020_SUBGROUP_BARRIER_PRESENT template diff --git a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_utils.h b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_utils.h index dd0a9e2816d..e98a34a212c 100644 --- a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_utils.h +++ b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_utils.h @@ -110,7 +110,7 @@ __supports_sub_group_size(const _ExecutionPolicy& __exec, std::size_t __target_s // 20201214 value corresponds to Intel(R) oneAPI C++ Compiler Classic 2021.1.2 Patch release #define _ONEDPL_SYCL2020_KERNEL_DEVICE_API_PRESENT \ - (__SYCL_COMPILER_VERSION > 20201214 || _ONEDPL_LIBSYCL_ZERO_OR_GE(50700)) + (_ONEDPL_GENERIC_SYCL_LIBRARY || __SYCL_COMPILER_VERSION > 20201214 || _ONEDPL_LIBSYCL_VERSION >= 50700) template ::std::size_t diff --git a/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_defs.h b/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_defs.h index 3f6241ffbf1..089ab4c7794 100644 --- a/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_defs.h +++ b/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_defs.h @@ -39,11 +39,10 @@ #if defined(__LIBSYCL_MAJOR_VERSION) && defined(__LIBSYCL_MINOR_VERSION) && defined(__LIBSYCL_PATCH_VERSION) # define _ONEDPL_LIBSYCL_VERSION \ (__LIBSYCL_MAJOR_VERSION * 10000 + __LIBSYCL_MINOR_VERSION * 100 + __LIBSYCL_PATCH_VERSION) -#else -# define _ONEDPL_LIBSYCL_VERSION 0 #endif -#define _ONEDPL_LIBSYCL_ZERO_OR_GE(_ONEDPL_LIBSYCL_VERSION) \ - (_ONEDPL_LIBSYCL_VERSION >= _ONEDPL_LIBSYCL_VERSION || _ONEDPL_LIBSYCL_VERSION == 0) +#if !defined(_ONEDPL_LIBSYCL_VERSION) +# define _ONEDPL_GENERIC_SYCL_LIBRARY 1 +#endif #if _ONEDPL_FPGA_DEVICE # if _ONEDPL_LIBSYCL_VERSION >= 50400 @@ -54,26 +53,32 @@ #endif // Macros to check the new SYCL features -#define _ONEDPL_SYCL2020_NO_INIT_PRESENT _ONEDPL_LIBSYCL_ZERO_OR_GE(50300) -#define _ONEDPL_SYCL2020_KERNEL_BUNDLE_PRESENT _ONEDPL_LIBSYCL_ZERO_OR_GE(50300) -#define _ONEDPL_SYCL2020_COLLECTIVES_PRESENT _ONEDPL_LIBSYCL_ZERO_OR_GE(50300) -#define _ONEDPL_SYCL2020_KNOWN_IDENTITY_PRESENT _ONEDPL_LIBSYCL_ZERO_OR_GE(50300) -#define _ONEDPL_SYCL2020_FUNCTIONAL_OBJECTS_PRESENT _ONEDPL_LIBSYCL_ZERO_OR_GE(50300) -#define _ONEDPL_SYCL2020_ATOMIC_REF_PRESENT _ONEDPL_LIBSYCL_ZERO_OR_GE(50500) -#define _ONEDPL_SYCL2020_PLACEHOLDER_HOST_ACCESSOR_DEPRECATED _ONEDPL_LIBSYCL_ZERO_OR_GE(60200) - -#define _ONEDPL_LIBSYCL_SUB_GROUP_MASK_PRESENT (SYCL_EXT_ONEAPI_SUB_GROUP_MASK >= 1) && \ - (_ONEDPL_LIBSYCL_VERSION >= 50700) +#define _ONEDPL_SYCL2020_NO_INIT_PRESENT \ + (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50300) +#define _ONEDPL_SYCL2020_KERNEL_BUNDLE_PRESENT \ + (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50300) +#define _ONEDPL_SYCL2020_COLLECTIVES_PRESENT \ + (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50300) +#define _ONEDPL_SYCL2020_KNOWN_IDENTITY_PRESENT \ + (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50300) +#define _ONEDPL_SYCL2020_FUNCTIONAL_OBJECTS_PRESENT \ + (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50300) +#define _ONEDPL_SYCL2020_ATOMIC_REF_PRESENT \ + (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50500) +#define _ONEDPL_SYCL2020_SUB_GROUP_PRESENT \ + (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50700) + +#define _ONEDPL_SYCL2020_PLACEHOLDER_HOST_ACCESSOR_DEPRECATED \ + (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 60200) + +#define _ONEDPL_LIBSYCL_SUB_GROUP_MASK_PRESENT (_ONEDPL_LIBSYCL_VERSION >= 50700) #define _ONEDPL_LIBSYCL_KNOWN_IDENTITY_PRESENT (_ONEDPL_LIBSYCL_VERSION == 50200) -#define _ONEDPL_SYCL_DEVICE_COPYABLE_SPECIALIZATION_BROKEN \ - (_ONEDPL_LIBSYCL_VERSION < 70100) && (_ONEDPL_LIBSYCL_VERSION != 0) +#define _ONEDPL_SYCL_DEVICE_COPYABLE_SPECIALIZATION_BROKEN (_ONEDPL_LIBSYCL_VERSION < 70100) // TODO: determine which compiler configurations provide subgroup load/store #define _ONEDPL_SYCL_SUB_GROUP_LOAD_STORE_PRESENT false -#define _ONEDPL_SYCL2020_SUB_GROUP_PRESENT _ONEDPL_LIBSYCL_ZERO_OR_GE(50700) - // Macro to check if we are compiling for SPIR-V devices. This macro must only be used within // SYCL kernels for determining SPIR-V compilation. Using this macro on the host may lead to incorrect behavior. #ifndef _ONEDPL_DETECT_SPIRV_COMPILATION // Check if overridden for testing @@ -84,7 +89,7 @@ # endif #endif // _ONEDPL_DETECT_SPIRV_COMPILATION -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(50300) +#if _ONEDPL_LIBSYCL_VERSION >= 50300 || _ONEDPL_GENERIC_SYCL_LIBRARY # define _ONEDPL_SYCL_REQD_SUB_GROUP_SIZE(SIZE) sycl::reqd_sub_group_size(SIZE) #else # define _ONEDPL_SYCL_REQD_SUB_GROUP_SIZE(SIZE) intel::reqd_sub_group_size(SIZE) @@ -101,7 +106,7 @@ // The unified future supporting USM memory and buffers is only supported after DPCPP 2023.1 // but not by 2023.2. -#if (_ONEDPL_LIBSYCL_ZERO_OR_GE(60100) && _ONEDPL_LIBSYCL_VERSION != 60200) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || (_ONEDPL_LIBSYCL_VERSION >= 50300 && _ONEDPL_LIBSYCL_VERSION != 60200) # define _ONEDPL_SYCL_UNIFIED_USM_BUFFER_PRESENT 1 #else # define _ONEDPL_SYCL_UNIFIED_USM_BUFFER_PRESENT 0 @@ -165,7 +170,7 @@ template constexpr auto __get_buffer_size(const _Buffer& __buffer) { -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(50300) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50300 return __buffer.size(); #else return __buffer.get_count(); @@ -176,7 +181,7 @@ template constexpr auto __get_accessor_size(const _Accessor& __accessor) { -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(50300) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50300 return __accessor.size(); #else return __accessor.get_count(); @@ -187,7 +192,7 @@ template constexpr void __group_barrier(_Item __item) { -#if 0 //_ONEDPL_LIBSYCL_ZERO_OR_GE(50300) +#if 0 //_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50300 //TODO: usage of sycl::group_barrier: probably, we have to revise SYCL parallel patterns which use a group_barrier. // 1) sycl::group_barrier() implementation is not ready // 2) sycl::group_barrier and sycl::item::group_barrier are not quite equivalent @@ -373,21 +378,21 @@ inline auto __fpga_selector() #endif // _ONEDPL_FPGA_DEVICE using __target = -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(50400) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50400 sycl::target; #else sycl::access::target; #endif constexpr __target __target_device = -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(50400) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50400 __target::device; #else __target::global_buffer; #endif constexpr __target __host_target = -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(60200) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 60200 __target::host_task; #else __target::host_buffer; @@ -395,7 +400,7 @@ constexpr __target __host_target = template using __buffer_allocator = -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(60000) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 60000 sycl::buffer_allocator<_DataT>; #else sycl::buffer_allocator; @@ -414,7 +419,7 @@ struct __atomic_ref : sycl::atomic<_AtomicType, _Space> template using __local_accessor = -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(60000) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 60000 sycl::local_accessor<_DataT, _Dimensions>; #else sycl::accessor<_DataT, _Dimensions, sycl::access::mode::read_write, __dpl_sycl::__target::local>; @@ -424,7 +429,7 @@ template auto __get_host_access(_Buf&& __buf) { -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(60200) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 60200 return ::std::forward<_Buf>(__buf).get_host_access(sycl::read_only); #else return ::std::forward<_Buf>(__buf).template get_access(); @@ -435,7 +440,7 @@ template auto __get_accessor_ptr(const _Acc& __acc) { -#if _ONEDPL_LIBSYCL_ZERO_OR_GE(70000) +#if _ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 70000 return __acc.template get_multi_ptr().get(); #else return __acc.get_pointer(); diff --git a/include/oneapi/dpl/pstl/utils.h b/include/oneapi/dpl/pstl/utils.h index f89855a293b..45e4d145a02 100644 --- a/include/oneapi/dpl/pstl/utils.h +++ b/include/oneapi/dpl/pstl/utils.h @@ -505,7 +505,7 @@ __dpl_bit_cast(const _Src& __src) noexcept { #if __cpp_lib_bit_cast >= 201806L return ::std::bit_cast<_Dst>(__src); -#elif _ONEDPL_BACKEND_SYCL && (_ONEDPL_LIBSYCL_VERSION >= 50300 || _ONEDPL_LIBSYCL_VERSION == 0) +#elif _ONEDPL_BACKEND_SYCL && (_ONEDPL_GENERIC_SYCL_LIBRARY || _ONEDPL_LIBSYCL_VERSION >= 50300) return sycl::bit_cast<_Dst>(__src); #elif __has_builtin(__builtin_bit_cast) return __builtin_bit_cast(_Dst, __src);