[SYCL][ABI-break] Promote guarded SYCL 2020 features and fix buffer reinterpret (#6541)

Promotes all SYCL 2020 features currently guarded by the SYCL2020_CONFORMANT_APIS macro.
Additionally, buffer::reinterpret is changed to correctly rebind the allocator. To accomplish this, std::allocator_traits is specialized for aligned_allocator.

Related change: intel/llvm-test-suite#1137

Author: steffenlarsen

sycl/CMakeLists.txt

@@ -30,7 +30,7 @@ set(SYCL_MINOR_VERSION 7)
 set(SYCL_PATCH_VERSION 0)
 # Don't forget to re-enable sycl_symbols_windows.dump once we leave ABI-breaking
 # window!
-set(SYCL_DEV_ABI_VERSION 4)
+set(SYCL_DEV_ABI_VERSION 5)
 if (SYCL_ADD_DEV_VERSION_POSTFIX)
   set(SYCL_VERSION_POSTFIX "-${SYCL_DEV_ABI_VERSION}")
 endif()

sycl/doc/PreprocessorMacros.md

@@ -55,24 +55,7 @@ This file describes macros that have effect on SYCL compiler and run-time.
 - **SYCL2020_CONFORMANT_APIS**
   This macro is used to comply with the SYCL 2020 specification, as some of the current 
   implementations may be widespread and not conform to it.
-  Description of what it changes:
-  1) According to spec, `backend_return_t` for opencl event 
-  should be `std::vector<cl_event>` instead of `cl_event`. Defining this macro 
-  will change the behavior of `sycl::get_native()` function and using types for 
-  next structs: `interop<backend::opencl, event>`, `BackendInput<backend::opencl, event>`, 
-  `BackendReturn<backend::opencl, event>` to be in line with the spec.
-  2) According to spec, `backend_return_t` for opencl buffer 
-  should be `std::vector<cl_mem>` instead of `cl_mem`. Defining this macro 
-  will change the behavior of `interop_handle::get_native_mem()` and `sycl::get_native()` functions
-  and using type for `BackendReturn<backend::opencl, buffer>` to be in line with the spec.
-  3) According to spec, `sycl::buffer_allocator` should be a template class taking a single
-  type parameter denoting the data type of the associated buffer. Likewise, `sycl::buffer`
-  with that take an allocator as a constructor argument should use
-  `sycl::buffer_allocator<std::remove_const_t<T>>`  by default, where `T` is the data type of
-  that buffer. Defining this macro will change the definition of `sycl::buffer_allocator` to
-  be templated and `sycl::buffer` will be using `sycl::buffer_allocator<std::remove_const_t<T>>`
-  by default, where `T` is the data type of that buffer, if it is not explicitly given an
-  allocator.
+  Defining this macro currently has no effect on the API.
 
 ## Version macros
 

sycl/include/sycl/backend.hpp

@@ -86,7 +86,6 @@ struct BufferInterop {
   }
 };
 
-#ifdef SYCL2020_CONFORMANT_APIS
 template <typename DataT, int Dimensions, typename AllocatorT>
 struct BufferInterop<backend::opencl, DataT, Dimensions, AllocatorT> {
   using ReturnType =
@@ -101,7 +100,6 @@ struct BufferInterop<backend::opencl, DataT, Dimensions, AllocatorT> {
     return ReturnValue;
   }
 };
-#endif
 
 template <backend BackendName, typename DataT, int Dimensions,
           typename AllocatorT>
@@ -143,30 +141,12 @@ auto get_native(const kernel_bundle<State> &Obj)
 }
 
 template <backend BackendName, typename DataT, int Dimensions,
-          typename AllocatorT,
-          std::enable_if_t<BackendName == backend::opencl> * = nullptr>
-#ifndef SYCL2020_CONFORMANT_APIS
-__SYCL_DEPRECATED(
-    "get_native<backend::opencl, buffer>, which return type "
-    "cl_mem is deprecated. According to SYCL 2020 spec, please define "
-    "SYCL2020_CONFORMANT_APIS and use vector<cl_mem> instead.")
-#endif
-auto get_native(const buffer<DataT, Dimensions, AllocatorT> &Obj)
-    -> backend_return_t<BackendName, buffer<DataT, Dimensions, AllocatorT>> {
-  return detail::get_native_buffer<BackendName>(Obj);
-}
-
-template <backend BackendName, typename DataT, int Dimensions,
-          typename AllocatorT,
-          std::enable_if_t<BackendName != backend::opencl> * = nullptr>
+          typename AllocatorT>
 auto get_native(const buffer<DataT, Dimensions, AllocatorT> &Obj)
     -> backend_return_t<BackendName, buffer<DataT, Dimensions, AllocatorT>> {
   return detail::get_native_buffer<BackendName>(Obj);
 }
 
-// define SYCL2020_CONFORMANT_APIS to correspond SYCL 2020 spec and return
-// vector<cl_event> from get_native instead of just cl_event
-#ifdef SYCL2020_CONFORMANT_APIS
 template <>
 inline backend_return_t<backend::opencl, event>
 get_native<backend::opencl, event>(const event &Obj) {
@@ -184,24 +164,6 @@ get_native<backend::opencl, event>(const event &Obj) {
   }
   return ReturnValue;
 }
-#else
-// Specialization for cl_event with deprecation message
-template <>
-__SYCL_DEPRECATED(
-    "get_native<backend::opencl, event>, which return type is "
-    "cl_event is deprecated. According to SYCL 2020 spec, please define "
-    "SYCL2020_CONFORMANT_APIS and use vector<cl_event> instead.")
-inline backend_return_t<backend::opencl, event> get_native<
-    backend::opencl, event>(const event &Obj) {
-  // TODO use SYCL 2020 exception when implemented
-  if (Obj.get_backend() != backend::opencl) {
-    throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
-                              PI_ERROR_INVALID_OPERATION);
-  }
-  return reinterpret_cast<
-      typename detail::interop<backend::opencl, event>::type>(Obj.getNative());
-}
-#endif
 
 // Native handle of an accessor should be accessed through interop_handler
 template <backend BackendName, typename DataT, int Dimensions,
@@ -334,7 +296,7 @@ typename std::enable_if<
 }
 
 template <backend Backend, typename T, int Dimensions = 1,
-          typename AllocatorT = detail::default_buffer_allocator<T>>
+          typename AllocatorT = buffer_allocator<std::remove_const_t<T>>>
 typename std::enable_if<detail::InteropFeatureSupportMap<Backend>::MakeBuffer ==
                                 true &&
                             Backend != backend::ext_oneapi_level_zero,

sycl/include/sycl/buffer.hpp

@@ -23,27 +23,11 @@ class handler;
 class queue;
 template <int dimensions> class range;
 
-// Guard SYCL 2020 buffer_allocator with template arguments behind the
-// SYCL2020_CONFORMANT_APIS macro.
-#ifdef SYCL2020_CONFORMANT_APIS
 template <typename DataT>
 using buffer_allocator = detail::sycl_memory_object_allocator<DataT>;
-#else
-using buffer_allocator = detail::sycl_memory_object_allocator<char>;
-#endif
 
 namespace detail {
 
-// Generalized implementation of the default allocator used by buffers.
-// TODO: When the SYCL 1.2.1 version of buffer_allocator is removed, this should
-//       be removed.
-#ifdef SYCL2020_CONFORMANT_APIS
-template <typename DataT>
-using default_buffer_allocator = buffer_allocator<std::remove_const_t<DataT>>;
-#else
-template <typename> using default_buffer_allocator = buffer_allocator;
-#endif
-
 template <typename T, int Dimensions, typename AllocatorT>
 buffer<T, Dimensions, AllocatorT, void>
 make_buffer_helper(pi_native_handle Handle, const context &Ctx, event Evt = {},
@@ -59,7 +43,7 @@ auto get_native_buffer(const buffer<DataT, Dimensions, Allocator, void> &Obj)
                         buffer<DataT, Dimensions, Allocator, void>>;
 
 template <backend Backend, typename DataT, int Dimensions,
-          typename AllocatorT = detail::default_buffer_allocator<DataT>>
+          typename AllocatorT = buffer_allocator<std::remove_const_t<DataT>>>
 struct BufferInterop;
 } // namespace detail
 
@@ -72,7 +56,7 @@ struct BufferInterop;
 ///
 /// \ingroup sycl_api
 template <typename T, int dimensions = 1,
-          typename AllocatorT = detail::default_buffer_allocator<T>,
+          typename AllocatorT = buffer_allocator<std::remove_const_t<T>>,
           typename __Enabled = typename detail::enable_if_t<(dimensions > 0) &&
                                                             (dimensions <= 3)>>
 class buffer {
@@ -492,7 +476,9 @@ class buffer {
   bool is_sub_buffer() const { return IsSubBuffer; }
 
   template <typename ReinterpretT, int ReinterpretDim>
-  buffer<ReinterpretT, ReinterpretDim, AllocatorT>
+  buffer<ReinterpretT, ReinterpretDim,
+         typename std::allocator_traits<AllocatorT>::template rebind_alloc<
+             ReinterpretT>>
   reinterpret(range<ReinterpretDim> reinterpretRange) const {
     if (sizeof(ReinterpretT) * reinterpretRange.size() != byte_size())
       throw sycl::invalid_object_error(
@@ -501,16 +487,22 @@ class buffer {
           "represented by the type and range of this SYCL buffer",
           PI_ERROR_INVALID_VALUE);
 
-    return buffer<ReinterpretT, ReinterpretDim, AllocatorT>(
+    return buffer<ReinterpretT, ReinterpretDim,
+                  typename std::allocator_traits<
+                      AllocatorT>::template rebind_alloc<ReinterpretT>>(
         impl, reinterpretRange, OffsetInBytes, IsSubBuffer);
   }
 
   template <typename ReinterpretT, int ReinterpretDim = dimensions>
   typename std::enable_if<
       (sizeof(ReinterpretT) == sizeof(T)) && (dimensions == ReinterpretDim),
-      buffer<ReinterpretT, ReinterpretDim, AllocatorT>>::type
+      buffer<ReinterpretT, ReinterpretDim,
+             typename std::allocator_traits<AllocatorT>::template rebind_alloc<
+                 ReinterpretT>>>::type
   reinterpret() const {
-    return buffer<ReinterpretT, ReinterpretDim, AllocatorT>(
+    return buffer<ReinterpretT, ReinterpretDim,
+                  typename std::allocator_traits<
+                      AllocatorT>::template rebind_alloc<ReinterpretT>>(
         impl, get_range(), OffsetInBytes, IsSubBuffer);
   }
 

sycl/include/sycl/detail/aligned_allocator.hpp

@@ -10,11 +10,11 @@
 
 #include <sycl/detail/common.hpp>
 #include <sycl/detail/os_util.hpp>
-#include <sycl/range.hpp>
 
 #include <cstdlib>
 #include <cstring>
 #include <memory>
+#include <type_traits>
 #include <vector>
 
 __SYCL_INLINE_NAMESPACE(cl) {
@@ -81,3 +81,62 @@ template <typename T> class aligned_allocator {
 } // namespace detail
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)
+
+namespace std {
+template <typename T>
+struct allocator_traits<sycl::detail::aligned_allocator<T>> {
+  using allocator_type = typename sycl::detail::aligned_allocator<T>;
+  using value_type = typename allocator_type::value_type;
+  using pointer = typename allocator_type::pointer;
+  using const_pointer = typename allocator_type::const_pointer;
+  using void_pointer =
+      typename std::pointer_traits<pointer>::template rebind<void>;
+  using const_void_pointer =
+      typename std::pointer_traits<pointer>::template rebind<const void>;
+  using difference_type =
+      typename std::pointer_traits<pointer>::difference_type;
+  using size_type = typename std::make_unsigned<difference_type>::type;
+  using propagate_on_container_copy_assignment = std::false_type;
+  using propagate_on_container_move_assignment = std::false_type;
+  using propagate_on_container_swap = std::false_type;
+  using is_always_equal = typename std::is_empty<allocator_type>::type;
+
+  template <typename U>
+  using rebind_alloc =
+      typename sycl::detail::aligned_allocator<T>::template rebind<U>::other;
+  template <typename U> using rebind_traits = allocator_traits<rebind_alloc<U>>;
+
+  static pointer allocate(allocator_type &Allocator, size_type NumElems) {
+    return Allocator.allocate(NumElems);
+  }
+
+  static pointer allocate(allocator_type &Allocator, size_type NumElems,
+                          const_void_pointer) {
+    // TODO: Utilize the locality hint argument.
+    return Allocator.allocate(NumElems);
+  }
+
+  static void deallocate(allocator_type &Allocator, pointer Ptr,
+                         size_type NumElems) {
+    Allocator.deallocate(Ptr, NumElems);
+  }
+
+  template <class U, class... ArgsT>
+  static void construct(allocator_type &Allocator, U *Ptr, ArgsT &&...Args) {
+    return Allocator.construct(Ptr, Args...);
+  }
+
+  template <class U> static void destroy(allocator_type &Allocator, U *Ptr) {
+    Allocator.destroy(Ptr);
+  }
+
+  static size_type max_size(const allocator_type &) noexcept {
+    return std::numeric_limits<size_type>::max() / sizeof(value_type);
+  }
+
+  static allocator_type
+  select_on_container_copy_construction(const allocator_type &Allocator) {
+    return Allocator;
+  }
+};
+} // namespace std

sycl/include/sycl/detail/backend_traits_opencl.hpp

@@ -84,17 +84,10 @@ struct BackendInput<backend::opencl, buffer<DataT, Dimensions, AllocatorT>> {
   using type = cl_mem;
 };
 
-#ifdef SYCL2020_CONFORMANT_APIS
 template <typename DataT, int Dimensions, typename AllocatorT>
 struct BackendReturn<backend::opencl, buffer<DataT, Dimensions, AllocatorT>> {
   using type = std::vector<cl_mem>;
 };
-#else
-template <typename DataT, int Dimensions, typename AllocatorT>
-struct BackendReturn<backend::opencl, buffer<DataT, Dimensions, AllocatorT>> {
-  using type = cl_mem;
-};
-#endif
 
 template <> struct BackendInput<backend::opencl, context> {
   using type = cl_context;
@@ -112,7 +105,6 @@ template <> struct BackendReturn<backend::opencl, device> {
   using type = cl_device_id;
 };
 
-#ifdef SYCL2020_CONFORMANT_APIS
 template <> struct interop<backend::opencl, event> {
   using type = std::vector<cl_event>;
   using value_type = cl_event;
@@ -125,17 +117,6 @@ template <> struct BackendReturn<backend::opencl, event> {
   using type = std::vector<cl_event>;
   using value_type = cl_event;
 };
-#else
-template <> struct interop<backend::opencl, event> {
-  using type = cl_event;
-};
-template <> struct BackendInput<backend::opencl, event> {
-  using type = cl_event;
-};
-template <> struct BackendReturn<backend::opencl, event> {
-  using type = cl_event;
-};
-#endif
 
 template <> struct BackendInput<backend::opencl, queue> {
   using type = cl_command_queue;

sycl/include/sycl/detail/sycl_mem_obj_allocator.hpp

@@ -8,12 +8,12 @@
 
 #pragma once
 
+#include <sycl/detail/aligned_allocator.hpp>
+
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 namespace detail {
 
-template <typename T> class aligned_allocator;
-
 template <typename DataT>
 using sycl_memory_object_allocator = aligned_allocator<DataT>;
 

sycl/include/sycl/ext/oneapi/backend/level_zero.hpp

@@ -193,7 +193,7 @@ inline kernel make_kernel<backend::ext_oneapi_level_zero>(
 
 // Specialization of sycl::make_buffer with event for Level-Zero backend.
 template <backend Backend, typename T, int Dimensions = 1,
-          typename AllocatorT = detail::default_buffer_allocator<T>>
+          typename AllocatorT = buffer_allocator<std::remove_const_t<T>>>
 typename std::enable_if<Backend == backend::ext_oneapi_level_zero,
                         buffer<T, Dimensions, AllocatorT>>::type
 make_buffer(
@@ -208,7 +208,7 @@ make_buffer(
 
 // Specialization of sycl::make_buffer for Level-Zero backend.
 template <backend Backend, typename T, int Dimensions = 1,
-          typename AllocatorT = detail::default_buffer_allocator<T>>
+          typename AllocatorT = buffer_allocator<std::remove_const_t<T>>>
 typename std::enable_if<Backend == backend::ext_oneapi_level_zero,
                         buffer<T, Dimensions, AllocatorT>>::type
 make_buffer(

sycl/test/abi/layout_buffer.cpp

@@ -79,7 +79,7 @@ void foo(sycl::buffer<int, 2>) {}
 // CHECK-NEXT:      | [sizeof=184, dsize=184, align=8,
 // CHECK-NEXT:      |  nvsize=184, nvalign=8]
 
-// CHECK: 0 | class sycl::buffer<int, 2, class sycl::detail::aligned_allocator<char>, void>
+// CHECK: 0 | class sycl::buffer<int, 2, class sycl::detail::aligned_allocator<int>, void>
 // CHECK-NEXT: 0 |   class std::shared_ptr<class sycl::detail::buffer_impl> impl
 // CHECK-NEXT: 0 |     class std::__shared_ptr<class sycl::detail::buffer_impl, __gnu_cxx::_S_atomic> (base)
 // CHECK-NEXT: 0 |       class std::__shared_ptr_access<class sycl::detail::buffer_impl, __gnu_cxx::_S_atomic, false, false> (base) (empty)

sycl/test/abi/user_mangling.cpp

@@ -49,7 +49,7 @@ void acc(sycl::accessor<int, 1, sycl::access::mode::read, sycl::access::target::
 // CHK-HOST: define dso_local void @_Z3accN2cl4sycl8accessorINS0_3vecIiLi4EEELi1ELNS0_6access4modeE1024ELNS4_6targetE2019ELNS4_11placeholderE0ENS0_3ext6oneapi22accessor_property_listIJEEEEE({{.*}})
 void acc(sycl::accessor<sycl::cl_int4, 1, sycl::access::mode::read, sycl::access::target::host_image>) {} 
 
-// CHK-HOST: define dso_local void @_Z3bufN2cl4sycl6bufferIiLi1ENS0_6detail17aligned_allocatorIcEEvEE({{.*}})
+// CHK-HOST: define dso_local void @_Z3bufN2cl4sycl6bufferIiLi1ENS0_6detail17aligned_allocatorIiEEvEE({{.*}})
 void buf(sycl::buffer<int>) {}
 
 // CHK-HOST: define dso_local void @_Z3ctxN2cl4sycl7contextE({{.*}})

sycl/test/regression/check_vector_of_opencl_event.cpp

@@ -1,4 +1,4 @@
-// RUN: %clangxx -fsycl -DSYCL2020_CONFORMANT_APIS %s -o %t.out
+// RUN: %clangxx -fsycl %s -o %t.out
 // RUN: %RUN_ON_HOST %t.out
 //
 //===----------------------------------------------------------------------===//

sycl/unittests/scheduler/NoHostUnifiedMemory.cpp

@@ -221,8 +221,8 @@ TEST_F(SchedulerTest, NoHostUnifiedMemory) {
     InteropPiContext = detail::getSyclObjImpl(InteropContext)->getHandleRef();
     auto BufI = std::make_shared<detail::buffer_impl>(
         detail::pi::cast<pi_native_handle>(MockInteropBuffer), Q.get_context(),
-        make_unique_ptr<detail::SYCLMemObjAllocatorHolder<
-            detail::default_buffer_allocator<char>, char>>(),
+        make_unique_ptr<
+            detail::SYCLMemObjAllocatorHolder<buffer_allocator<char>, char>>(),
         /* OwnNativeHandle */ true, event());
 
     detail::Requirement Req = getMockRequirement();