[SYCL] Fix accessor construction from a buffer.

Allow accessor to be constructed from a buffer with a
non-default allocator.

Signed-off-by: Ilya Stepykin <ilya.stepykin@intel.com>

Author: Ilya Stepykin

sycl/include/CL/sycl/accessor.hpp

@@ -741,11 +741,11 @@ class accessor :
   using reference = DataT &;
   using const_reference = const DataT &;
 
-  template <int Dims = Dimensions>
-  accessor(
-      detail::enable_if_t<Dims == 0 && ((!IsPlaceH && IsHostBuf) ||
-                                (IsPlaceH && (IsGlobalBuf || IsConstantBuf))),
-                  buffer<DataT, 1>> &BufferRef)
+  template <int Dims = Dimensions, typename AllocatorT,
+            typename detail::enable_if_t<
+                Dims == 0 && ((!IsPlaceH && IsHostBuf) ||
+                (IsPlaceH && (IsGlobalBuf || IsConstantBuf)))>* = nullptr>
+  accessor(buffer<DataT, 1, AllocatorT> &BufferRef)
 #ifdef __SYCL_DEVICE_ONLY__
       : impl(id<AdjustedDim>(), BufferRef.get_range(), BufferRef.MemRange) {
 #else
@@ -762,11 +762,11 @@ class accessor :
 #endif
   }
 
-  template <int Dims = Dimensions>
-  accessor(
-      buffer<DataT, 1> &BufferRef,
-      detail::enable_if_t<Dims == 0 && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf)),
-                  handler> &CommandGroupHandler)
+  template <int Dims = Dimensions, typename AllocatorT>
+  accessor(buffer<DataT, 1, AllocatorT> &BufferRef,
+           detail::enable_if_t<Dims == 0 &&
+                               (!IsPlaceH && (IsGlobalBuf || IsConstantBuf)),
+                               handler> &CommandGroupHandler)
 #ifdef __SYCL_DEVICE_ONLY__
       : impl(id<AdjustedDim>(), BufferRef.get_range(), BufferRef.MemRange) {
   }
@@ -781,11 +781,12 @@ class accessor :
   }
 #endif
 
-  template <int Dims = Dimensions,
-            typename = detail::enable_if_t<
+  template <int Dims = Dimensions, typename AllocatorT,
+            typename detail::enable_if_t<
                 (Dims > 0) && ((!IsPlaceH && IsHostBuf) ||
-                               (IsPlaceH && (IsGlobalBuf || IsConstantBuf)))>>
-  accessor(buffer<DataT, Dimensions> &BufferRef)
+                               (IsPlaceH && (IsGlobalBuf || IsConstantBuf)))>
+                * = nullptr>
+  accessor(buffer<DataT, Dimensions, AllocatorT> &BufferRef)
 #ifdef __SYCL_DEVICE_ONLY__
       : impl(id<Dimensions>(), BufferRef.get_range(), BufferRef.MemRange) {
   }
@@ -803,10 +804,11 @@ class accessor :
   }
 #endif
 
-  template <int Dims = Dimensions,
+  template <int Dims = Dimensions, typename AllocatorT,
             typename = detail::enable_if_t<
                 (Dims > 0) && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf))>>
-  accessor(buffer<DataT, Dimensions> &BufferRef, handler &CommandGroupHandler)
+  accessor(buffer<DataT, Dimensions, AllocatorT> &BufferRef,
+           handler &CommandGroupHandler)
 #ifdef __SYCL_DEVICE_ONLY__
       : impl(id<AdjustedDim>(), BufferRef.get_range(), BufferRef.MemRange) {
   }
@@ -821,12 +823,12 @@ class accessor :
   }
 #endif
 
-  template <int Dims = Dimensions,
+  template <int Dims = Dimensions, typename AllocatorT,
             typename = detail::enable_if_t<
                 (Dims > 0) && ((!IsPlaceH && IsHostBuf) ||
                                (IsPlaceH && (IsGlobalBuf || IsConstantBuf)))>>
-  accessor(buffer<DataT, Dimensions> &BufferRef, range<Dimensions> AccessRange,
-           id<Dimensions> AccessOffset = {})
+  accessor(buffer<DataT, Dimensions, AllocatorT> &BufferRef,
+           range<Dimensions> AccessRange, id<Dimensions> AccessOffset = {})
 #ifdef __SYCL_DEVICE_ONLY__
       : impl(AccessOffset, AccessRange, BufferRef.MemRange) {
   }
@@ -843,11 +845,12 @@ class accessor :
   }
 #endif
 
-  template <int Dims = Dimensions,
+  template <int Dims = Dimensions, typename AllocatorT,
             typename = detail::enable_if_t<
                 (Dims > 0) && (!IsPlaceH && (IsGlobalBuf || IsConstantBuf))>>
-  accessor(buffer<DataT, Dimensions> &BufferRef, handler &CommandGroupHandler,
-           range<Dimensions> AccessRange, id<Dimensions> AccessOffset = {})
+  accessor(buffer<DataT, Dimensions, AllocatorT> &BufferRef,
+           handler &CommandGroupHandler, range<Dimensions> AccessRange,
+           id<Dimensions> AccessOffset = {})
 #ifdef __SYCL_DEVICE_ONLY__
       : impl(AccessOffset, AccessRange, BufferRef.MemRange) {
   }

sycl/test/basic_tests/accessor/accessor.cpp

@@ -379,4 +379,52 @@ int main() {
       return 1;
     }
   }
+
+  {
+    // Call every available accessor's constructor to ensure that they work with
+    // a buffer with a non-default allocator.
+    int data[] = {1, 2, 3};
+
+    using allocator_type = std::allocator<int>;
+
+    sycl::buffer<int, 1, allocator_type> buf1(&data[0], sycl::range<1>(1),
+                                              allocator_type{});
+    sycl::buffer<int, 1, allocator_type> buf2(&data[1], sycl::range<1>(1),
+                                              allocator_type{});
+    sycl::buffer<int, 1, allocator_type> buf3(&data[2], sycl::range<1>(1),
+                                              allocator_type{});
+
+    sycl::queue queue;
+    queue.submit([&](sycl::handler &cgh) {
+      sycl::accessor<int, 0, sycl::access::mode::read_write,
+                     sycl::access::target::global_buffer>
+          acc1(buf1, cgh);
+      sycl::accessor<int, 1, sycl::access::mode::read_write,
+                     sycl::access::target::global_buffer>
+          acc2(buf2, cgh);
+      sycl::accessor<int, 1, sycl::access::mode::read_write,
+                     sycl::access::target::global_buffer>
+          acc3(buf3, cgh, sycl::range<1>(1));
+
+      cgh.single_task<class acc_alloc_buf>([=]() {
+        acc1 *= 2;
+        acc2[0] *= 2;
+        acc3[0] *= 2;
+      });
+    });
+
+    sycl::accessor<int, 0, sycl::access::mode::read,
+                   sycl::access::target::host_buffer>
+        acc4(buf1);
+    sycl::accessor<int, 1, sycl::access::mode::read,
+                   sycl::access::target::host_buffer>
+        acc5(buf2);
+    sycl::accessor<int, 1, sycl::access::mode::read,
+                   sycl::access::target::host_buffer>
+        acc6(buf3, sycl::range<1>(1));
+
+    assert(acc4 == 2);
+    assert(acc5[0] == 4);
+    assert(acc6[0] == 6);
+  }
 }