DPC++ Runtime
Runtime libraries for oneAPI DPC++
buffer.hpp
Go to the documentation of this file.
1 //==----------- buffer.hpp --- SYCL buffer ---------------------------------==//
2 //
3 // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4 // See https://llvm.org/LICENSE.txt for license information.
5 // SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6 //
7 //===----------------------------------------------------------------------===//
8 
9 #pragma once
10 
14 #include <CL/sycl/exception.hpp>
16 #include <CL/sycl/stl.hpp>
18 
20 namespace sycl {
21 
22 class handler;
23 class queue;
24 template <int dimensions> class range;
25 
26 // Guard SYCL 2020 buffer_allocator with template arguments behind the
27 // SYCL2020_CONFORMANT_APIS macro.
28 #ifdef SYCL2020_CONFORMANT_APIS
29 template <typename DataT>
31 #else
33 #endif
34 
35 namespace detail {
36 
37 // Generalized implementation of the default allocator used by buffers.
38 // TODO: When the SYCL 1.2.1 version of buffer_allocator is removed, this should
39 // be removed.
40 #ifdef SYCL2020_CONFORMANT_APIS
41 template <typename DataT>
43 #else
44 template <typename> using default_buffer_allocator = buffer_allocator;
45 #endif
46 
47 template <typename T, int Dimensions, typename AllocatorT>
49 make_buffer_helper(pi_native_handle Handle, const context &Ctx, event Evt = {},
50  bool OwnNativeHandle = true) {
51  return buffer<T, Dimensions, AllocatorT, void>(Handle, Ctx, OwnNativeHandle,
52  Evt);
53 }
54 
55 template <backend BackendName, typename DataT, int Dimensions,
56  typename Allocator>
57 auto get_native_buffer(const buffer<DataT, Dimensions, Allocator, void> &Obj)
58  -> backend_return_t<BackendName,
59  buffer<DataT, Dimensions, Allocator, void>>;
60 
61 template <backend Backend, typename DataT, int Dimensions,
62  typename AllocatorT = detail::default_buffer_allocator<DataT>>
63 struct BufferInterop;
64 } // namespace detail
65 
74 template <typename T, int dimensions = 1,
75  typename AllocatorT = detail::default_buffer_allocator<T>,
76  typename __Enabled = typename detail::enable_if_t<(dimensions > 0) &&
77  (dimensions <= 3)>>
78 class buffer {
79  // TODO check is_device_copyable<T>::value after converting sycl::vec into a
80  // trivially copyable class.
81  static_assert(!std::is_same<T, std::string>::value,
82  "'std::string' is not a device copyable type");
83 
84 public:
85  using value_type = T;
86  using reference = value_type &;
87  using const_reference = const value_type &;
88  using allocator_type = AllocatorT;
89  template <int dims>
91  // using same requirement for contiguous container as std::span
92  template <class Container>
93  using EnableIfContiguous =
94  detail::void_t<detail::enable_if_t<std::is_convertible<
96  decltype(std::declval<Container>().data())> (*)[],
97  const T (*)[]>::value>,
98  decltype(std::declval<Container>().size())>;
99  template <class It>
101  std::is_convertible<typename std::iterator_traits<It>::iterator_category,
102  std::input_iterator_tag>::value>;
103  template <typename ItA, typename ItB>
105  std::is_same<ItA, ItB>::value && !std::is_const<ItA>::value, ItA>;
106 
107  std::array<size_t, 3> rangeToArray(range<3> &r) { return {r[0], r[1], r[2]}; }
108 
109  std::array<size_t, 3> rangeToArray(range<2> &r) { return {r[0], r[1], 0}; }
110 
111  std::array<size_t, 3> rangeToArray(range<1> &r) { return {r[0], 0, 0}; }
112 
113  buffer(const range<dimensions> &bufferRange,
114  const property_list &propList = {},
115  const detail::code_location CodeLoc = detail::code_location::current())
116  : Range(bufferRange) {
117  impl = std::make_shared<detail::buffer_impl>(
118  size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)), propList,
120  impl->constructorNotification(CodeLoc, (void *)impl.get(), nullptr,
121  (const void *)typeid(T).name(), dimensions,
122  sizeof(T), rangeToArray(Range).data());
123  }
124 
125  buffer(const range<dimensions> &bufferRange, AllocatorT allocator,
126  const property_list &propList = {},
127  const detail::code_location CodeLoc = detail::code_location::current())
128  : Range(bufferRange) {
129  impl = std::make_shared<detail::buffer_impl>(
130  size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)), propList,
132  allocator));
133  impl->constructorNotification(CodeLoc, (void *)impl.get(), nullptr,
134  (const void *)typeid(T).name(), dimensions,
135  sizeof(T), rangeToArray(Range).data());
136  }
137 
138  buffer(T *hostData, const range<dimensions> &bufferRange,
139  const property_list &propList = {},
140  const detail::code_location CodeLoc = detail::code_location::current())
141  : Range(bufferRange) {
142  impl = std::make_shared<detail::buffer_impl>(
143  hostData, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
144  propList,
146  impl->constructorNotification(CodeLoc, (void *)impl.get(), hostData,
147  (const void *)typeid(T).name(), dimensions,
148  sizeof(T), rangeToArray(Range).data());
149  }
150 
151  buffer(T *hostData, const range<dimensions> &bufferRange,
152  AllocatorT allocator, const property_list &propList = {},
153  const detail::code_location CodeLoc = detail::code_location::current())
154  : Range(bufferRange) {
155  impl = std::make_shared<detail::buffer_impl>(
156  hostData, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
157  propList,
159  allocator));
160  impl->constructorNotification(CodeLoc, (void *)impl.get(), hostData,
161  (const void *)typeid(T).name(), dimensions,
162  sizeof(T), rangeToArray(Range).data());
163  }
164 
165  template <typename _T = T>
167  const range<dimensions> &bufferRange,
168  const property_list &propList = {},
169  const detail::code_location CodeLoc = detail::code_location::current())
170  : Range(bufferRange) {
171  impl = std::make_shared<detail::buffer_impl>(
172  hostData, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
173  propList,
175  impl->constructorNotification(CodeLoc, (void *)impl.get(), hostData,
176  (const void *)typeid(T).name(), dimensions,
177  sizeof(T), rangeToArray(Range).data());
178  }
179 
180  template <typename _T = T>
182  const range<dimensions> &bufferRange, AllocatorT allocator,
183  const property_list &propList = {},
184  const detail::code_location CodeLoc = detail::code_location::current())
185  : Range(bufferRange) {
186  impl = std::make_shared<detail::buffer_impl>(
187  hostData, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
188  propList,
190  allocator));
191  impl->constructorNotification(CodeLoc, (void *)impl.get(), hostData,
192  (const void *)typeid(T).name(), dimensions,
193  sizeof(T), rangeToArray(Range).data());
194  }
195 
196  buffer(const std::shared_ptr<T> &hostData,
197  const range<dimensions> &bufferRange, AllocatorT allocator,
198  const property_list &propList = {},
199  const detail::code_location CodeLoc = detail::code_location::current())
200  : Range(bufferRange) {
201  impl = std::make_shared<detail::buffer_impl>(
202  hostData, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
203  propList,
205  allocator));
206  impl->constructorNotification(CodeLoc, (void *)impl.get(),
207  (void *)hostData.get(),
208  (const void *)typeid(T).name(), dimensions,
209  sizeof(T), rangeToArray(Range).data());
210  }
211 
212  buffer(const std::shared_ptr<T[]> &hostData,
213  const range<dimensions> &bufferRange, AllocatorT allocator,
214  const property_list &propList = {},
215  const detail::code_location CodeLoc = detail::code_location::current())
216  : Range(bufferRange) {
217  impl = std::make_shared<detail::buffer_impl>(
218  hostData, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
219  propList,
221  allocator));
222  impl->constructorNotification(CodeLoc, (void *)impl.get(),
223  (void *)hostData.get(),
224  (const void *)typeid(T).name(), dimensions,
225  sizeof(T), rangeToArray(Range).data());
226  }
227 
228  buffer(const std::shared_ptr<T> &hostData,
229  const range<dimensions> &bufferRange,
230  const property_list &propList = {},
231  const detail::code_location CodeLoc = detail::code_location::current())
232  : Range(bufferRange) {
233  impl = std::make_shared<detail::buffer_impl>(
234  hostData, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
235  propList,
237  impl->constructorNotification(CodeLoc, (void *)impl.get(),
238  (void *)hostData.get(),
239  (const void *)typeid(T).name(), dimensions,
240  sizeof(T), rangeToArray(Range).data());
241  }
242 
243  buffer(const std::shared_ptr<T[]> &hostData,
244  const range<dimensions> &bufferRange,
245  const property_list &propList = {},
246  const detail::code_location CodeLoc = detail::code_location::current())
247  : Range(bufferRange) {
248  impl = std::make_shared<detail::buffer_impl>(
249  hostData, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
250  propList,
252  impl->constructorNotification(CodeLoc, (void *)impl.get(),
253  (void *)hostData.get(),
254  (const void *)typeid(T).name(), dimensions,
255  sizeof(T), rangeToArray(Range).data());
256  }
257 
258  template <class InputIterator, int N = dimensions,
259  typename = EnableIfOneDimension<N>,
260  typename = EnableIfItInputIterator<InputIterator>>
261  buffer(InputIterator first, InputIterator last, AllocatorT allocator,
262  const property_list &propList = {},
263  const detail::code_location CodeLoc = detail::code_location::current())
264  : Range(range<1>(std::distance(first, last))) {
265  impl = std::make_shared<detail::buffer_impl>(
266  first, last, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
267  propList,
269  allocator));
270  size_t r[3] = {Range[0], 0, 0};
271  impl->constructorNotification(CodeLoc, (void *)impl.get(), &first,
272  (const void *)typeid(T).name(), dimensions,
273  sizeof(T), r);
274  }
275 
276  template <class InputIterator, int N = dimensions,
277  typename = EnableIfOneDimension<N>,
278  typename = EnableIfItInputIterator<InputIterator>>
279  buffer(InputIterator first, InputIterator last,
280  const property_list &propList = {},
281  const detail::code_location CodeLoc = detail::code_location::current())
282  : Range(range<1>(std::distance(first, last))) {
283  impl = std::make_shared<detail::buffer_impl>(
284  first, last, size() * sizeof(T), detail::getNextPowerOfTwo(sizeof(T)),
285  propList,
287  size_t r[3] = {Range[0], 0, 0};
288  impl->constructorNotification(CodeLoc, (void *)impl.get(), &first,
289  (const void *)typeid(T).name(), dimensions,
290  sizeof(T), r);
291  }
292 
293  // This constructor is a prototype for a future SYCL specification
294  template <class Container, int N = dimensions,
295  typename = EnableIfOneDimension<N>,
296  typename = EnableIfContiguous<Container>>
297  buffer(Container &container, AllocatorT allocator,
298  const property_list &propList = {},
299  const detail::code_location CodeLoc = detail::code_location::current())
300  : Range(range<1>(container.size())) {
301  impl = std::make_shared<detail::buffer_impl>(
302  container.data(), size() * sizeof(T),
303  detail::getNextPowerOfTwo(sizeof(T)), propList,
305  allocator));
306  size_t r[3] = {Range[0], 0, 0};
307  impl->constructorNotification(CodeLoc, (void *)impl.get(), container.data(),
308  (const void *)typeid(T).name(), dimensions,
309  sizeof(T), r);
310  }
311 
312  // This constructor is a prototype for a future SYCL specification
313  template <class Container, int N = dimensions,
314  typename = EnableIfOneDimension<N>,
315  typename = EnableIfContiguous<Container>>
316  buffer(Container &container, const property_list &propList = {},
317  const detail::code_location CodeLoc = detail::code_location::current())
318  : buffer(container, {}, propList, CodeLoc) {}
319 
321  const range<dimensions> &subRange,
322  const detail::code_location CodeLoc = detail::code_location::current())
323  : impl(b.impl), Range(subRange),
324  OffsetInBytes(getOffsetInBytes<T>(baseIndex, b.Range)),
325  IsSubBuffer(true) {
326  impl->constructorNotification(CodeLoc, (void *)impl.get(), impl.get(),
327  (const void *)typeid(T).name(), dimensions,
328  sizeof(T), rangeToArray(Range).data());
329 
330  if (b.is_sub_buffer())
331  throw cl::sycl::invalid_object_error(
332  "Cannot create sub buffer from sub buffer.", PI_ERROR_INVALID_VALUE);
333  if (isOutOfBounds(baseIndex, subRange, b.Range))
334  throw cl::sycl::invalid_object_error(
335  "Requested sub-buffer size exceeds the size of the parent buffer",
336  PI_ERROR_INVALID_VALUE);
337  if (!isContiguousRegion(baseIndex, subRange, b.Range))
338  throw cl::sycl::invalid_object_error(
339  "Requested sub-buffer region is not contiguous",
340  PI_ERROR_INVALID_VALUE);
341  }
342 
343 #ifdef __SYCL_INTERNAL_API
344  template <int N = dimensions, typename = EnableIfOneDimension<N>>
345  buffer(cl_mem MemObject, const context &SyclContext,
346  event AvailableEvent = {},
347  const detail::code_location CodeLoc = detail::code_location::current())
348  : Range{0} {
349 
350  impl = std::make_shared<detail::buffer_impl>(
351  detail::pi::cast<pi_native_handle>(MemObject), SyclContext,
352  make_unique_ptr<detail::SYCLMemObjAllocatorHolder<AllocatorT, T>>(),
353  /* OwnNativeHandle */ true, AvailableEvent);
354  Range[0] = impl->getSize() / sizeof(T);
355  impl->constructorNotification(CodeLoc, (void *)impl.get(), &MemObject,
356  (const void *)typeid(T).name(), dimensions,
357  sizeof(T), rangeToArray(Range).data());
358  }
359 #endif
360 
361  buffer(const buffer &rhs,
362  const detail::code_location CodeLoc = detail::code_location::current())
363  : impl(rhs.impl), Range(rhs.Range), OffsetInBytes(rhs.OffsetInBytes),
364  IsSubBuffer(rhs.IsSubBuffer) {
365  impl->constructorNotification(CodeLoc, (void *)impl.get(), impl.get(),
366  (const void *)typeid(T).name(), dimensions,
367  sizeof(T), rangeToArray(Range).data());
368  }
369 
370  buffer(buffer &&rhs,
371  const detail::code_location CodeLoc = detail::code_location::current())
372  : impl(std::move(rhs.impl)), Range(rhs.Range),
373  OffsetInBytes(rhs.OffsetInBytes), IsSubBuffer(rhs.IsSubBuffer) {
374  impl->constructorNotification(CodeLoc, (void *)impl.get(), impl.get(),
375  (const void *)typeid(T).name(), dimensions,
376  sizeof(T), rangeToArray(Range).data());
377  }
378 
379  buffer &operator=(const buffer &rhs) = default;
380 
381  buffer &operator=(buffer &&rhs) = default;
382 
383  ~buffer() = default;
384 
385  bool operator==(const buffer &rhs) const { return impl == rhs.impl; }
386 
387  bool operator!=(const buffer &rhs) const { return !(*this == rhs); }
388 
389  /* -- common interface members -- */
390 
391  /* -- property interface members -- */
392 
393  range<dimensions> get_range() const { return Range; }
394 
395  __SYCL2020_DEPRECATED("get_count() is deprecated, please use size() instead")
396  size_t get_count() const { return size(); }
397  size_t size() const noexcept { return Range.size(); }
398 
400  "get_size() is deprecated, please use byte_size() instead")
401  size_t get_size() const { return byte_size(); }
402  size_t byte_size() const noexcept { return size() * sizeof(T); }
403 
404  AllocatorT get_allocator() const {
405  return impl->template get_allocator<AllocatorT>();
406  }
407 
408  template <access::mode Mode, access::target Target = access::target::device>
409  accessor<T, dimensions, Mode, Target, access::placeholder::false_t,
412  handler &CommandGroupHandler,
413  const detail::code_location CodeLoc = detail::code_location::current()) {
414  return accessor<T, dimensions, Mode, Target, access::placeholder::false_t,
416  *this, CommandGroupHandler, {}, CodeLoc);
417  }
418 
419  template <access::mode mode>
420  accessor<T, dimensions, mode, access::target::host_buffer,
421  access::placeholder::false_t, ext::oneapi::accessor_property_list<>>
423  const detail::code_location CodeLoc = detail::code_location::current()) {
424  return accessor<T, dimensions, mode, access::target::host_buffer,
425  access::placeholder::false_t,
426  ext::oneapi::accessor_property_list<>>(*this, {}, CodeLoc);
427  }
428 
429  template <access::mode mode, access::target target = access::target::device>
430  accessor<T, dimensions, mode, target, access::placeholder::false_t,
433  handler &commandGroupHandler, range<dimensions> accessRange,
434  id<dimensions> accessOffset = {},
435  const detail::code_location CodeLoc = detail::code_location::current()) {
436  if (isOutOfBounds(accessOffset, accessRange, this->Range))
437  throw cl::sycl::invalid_object_error(
438  "Requested accessor would exceed the bounds of the buffer",
439  PI_ERROR_INVALID_VALUE);
440 
441  return accessor<T, dimensions, mode, target, access::placeholder::false_t,
443  *this, commandGroupHandler, accessRange, accessOffset, {}, CodeLoc);
444  }
445 
446  template <access::mode mode>
447  accessor<T, dimensions, mode, access::target::host_buffer,
448  access::placeholder::false_t, ext::oneapi::accessor_property_list<>>
450  range<dimensions> accessRange, id<dimensions> accessOffset = {},
451  const detail::code_location CodeLoc = detail::code_location::current()) {
452  if (isOutOfBounds(accessOffset, accessRange, this->Range))
453  throw cl::sycl::invalid_object_error(
454  "Requested accessor would exceed the bounds of the buffer",
455  PI_ERROR_INVALID_VALUE);
456 
457  return accessor<T, dimensions, mode, access::target::host_buffer,
458  access::placeholder::false_t,
460  *this, accessRange, accessOffset, {}, CodeLoc);
461  }
462 
463 #if __cplusplus >= 201703L
464 
465  template <typename... Ts> auto get_access(Ts... args) {
466  return accessor{*this, args...};
467  }
468 
469  template <typename... Ts>
470  auto get_access(handler &commandGroupHandler, Ts... args) {
471  return accessor{*this, commandGroupHandler, args...};
472  }
473 
474  template <typename... Ts> auto get_host_access(Ts... args) {
475  return host_accessor{*this, args...};
476  }
477 
478  template <typename... Ts>
479  auto get_host_access(handler &commandGroupHandler, Ts... args) {
480  return host_accessor{*this, commandGroupHandler, args...};
481  }
482 
483 #endif
484 
485  template <typename Destination = std::nullptr_t>
486  void set_final_data(Destination finalData = nullptr) {
487  impl->set_final_data(finalData);
488  }
489 
490  void set_write_back(bool flag = true) { impl->set_write_back(flag); }
491 
492  bool is_sub_buffer() const { return IsSubBuffer; }
493 
494  template <typename ReinterpretT, int ReinterpretDim>
496  reinterpret(range<ReinterpretDim> reinterpretRange) const {
497  if (sizeof(ReinterpretT) * reinterpretRange.size() != byte_size())
498  throw cl::sycl::invalid_object_error(
499  "Total size in bytes represented by the type and range of the "
500  "reinterpreted SYCL buffer does not equal the total size in bytes "
501  "represented by the type and range of this SYCL buffer",
502  PI_ERROR_INVALID_VALUE);
503 
505  impl, reinterpretRange, OffsetInBytes, IsSubBuffer);
506  }
507 
508  template <typename ReinterpretT, int ReinterpretDim = dimensions>
509  typename std::enable_if<
510  (sizeof(ReinterpretT) == sizeof(T)) && (dimensions == ReinterpretDim),
512  reinterpret() const {
514  impl, get_range(), OffsetInBytes, IsSubBuffer);
515  }
516 
517  template <typename ReinterpretT, int ReinterpretDim = dimensions>
518  typename std::enable_if<
519  (ReinterpretDim == 1) && ((dimensions != ReinterpretDim) ||
520  (sizeof(ReinterpretT) != sizeof(T))),
522  reinterpret() const {
523  long sz = byte_size();
524  if (sz % sizeof(ReinterpretT) != 0)
525  throw cl::sycl::invalid_object_error(
526  "Total byte size of buffer is not evenly divisible by the size of "
527  "the reinterpreted type",
528  PI_ERROR_INVALID_VALUE);
529 
531  impl, range<1>{sz / sizeof(ReinterpretT)}, OffsetInBytes, IsSubBuffer);
532  }
533 
534  template <typename propertyT> bool has_property() const {
535  return impl->template has_property<propertyT>();
536  }
537 
538  template <typename propertyT> propertyT get_property() const {
539  return impl->template get_property<propertyT>();
540  }
541 
542 protected:
543  bool isOutOfBounds(const id<dimensions> &offset,
544  const range<dimensions> &newRange,
545  const range<dimensions> &parentRange) {
546  bool outOfBounds = false;
547  for (int i = 0; i < dimensions; ++i)
548  outOfBounds |= newRange[i] + offset[i] > parentRange[i];
549 
550  return outOfBounds;
551  }
552 
553 private:
554  std::shared_ptr<detail::buffer_impl> impl;
555  template <class Obj>
556  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
557  template <typename A, int dims, typename C, typename Enable>
558  friend class buffer;
559  template <typename DataT, int dims, access::mode mode, access::target target,
560  access::placeholder isPlaceholder, typename PropertyListT>
561  friend class accessor;
562  template <typename HT, int HDims, typename HAllocT>
565  range<dimensions> Range;
566  // Offset field specifies the origin of the sub buffer inside the parent
567  // buffer
568  size_t OffsetInBytes = 0;
569  bool IsSubBuffer = false;
570 
571  // Interop constructor
572  template <int N = dimensions, typename = EnableIfOneDimension<N>>
573  buffer(pi_native_handle MemObject, const context &SyclContext,
574  bool OwnNativeHandle, event AvailableEvent = {},
575  const detail::code_location CodeLoc = detail::code_location::current())
576  : Range{0} {
577 
578  impl = std::make_shared<detail::buffer_impl>(
579  MemObject, SyclContext,
580  make_unique_ptr<detail::SYCLMemObjAllocatorHolder<AllocatorT, T>>(),
581  OwnNativeHandle, AvailableEvent);
582  Range[0] = impl->getSize() / sizeof(T);
583  impl->constructorNotification(CodeLoc, (void *)impl.get(), &MemObject,
584  (const void *)typeid(T).name(), dimensions,
585  sizeof(T), rangeToArray(Range).data());
586  }
587 
588  // Reinterpret contructor
589  buffer(std::shared_ptr<detail::buffer_impl> Impl,
590  range<dimensions> reinterpretRange, size_t reinterpretOffset,
591  bool isSubBuffer,
592  const detail::code_location CodeLoc = detail::code_location::current())
593  : impl(Impl), Range(reinterpretRange), OffsetInBytes(reinterpretOffset),
594  IsSubBuffer(isSubBuffer) {
595  impl->constructorNotification(CodeLoc, (void *)impl.get(), Impl.get(),
596  (const void *)typeid(T).name(), dimensions,
597  sizeof(T), rangeToArray(Range).data());
598  }
599 
600  template <typename Type, int N>
601  size_t getOffsetInBytes(const id<N> &offset, const range<N> &range) {
602  return detail::getLinearIndex(offset, range) * sizeof(Type);
603  }
604 
605  bool isContiguousRegion(const id<1> &, const range<1> &, const range<1> &) {
606  // 1D sub buffer always has contiguous region
607  return true;
608  }
609 
610  bool isContiguousRegion(const id<2> &offset, const range<2> &newRange,
611  const range<2> &parentRange) {
612  // For 2D sub buffer there are 2 cases:
613  // 1) Offset {Any, Any} | a piece of any line of a buffer
614  // Range {1, Any} |
615  // 2) Offset {Any, 0 } | any number of full lines
616  // Range {Any, Col} |
617  // where Col is a number of columns of original buffer
618  if (offset[1])
619  return newRange[0] == 1;
620  return newRange[1] == parentRange[1];
621  }
622 
623  bool isContiguousRegion(const id<3> &offset, const range<3> &newRange,
624  const range<3> &parentRange) {
625  // For 3D sub buffer there are 3 cases:
626  // 1) Offset {Any, Any, Any} | a piece of any line in any slice of a buffer
627  // Range {1, 1, Any} |
628  // 2) Offset {Any, Any, 0 } | any number of full lines in any slice
629  // Range {1, Any, Col} |
630  // 3) Offset {Any, 0, 0 } | any number of slices
631  // Range {Any, Row, Col} |
632  // where Row and Col are numbers of rows and columns of original buffer
633  if (offset[2])
634  return newRange[0] == 1 && newRange[1] == 1;
635  if (offset[1])
636  return newRange[0] == 1 && newRange[2] == parentRange[2];
637  return newRange[1] == parentRange[1] && newRange[2] == parentRange[2];
638  }
639 
640  template <backend BackendName, typename DataT, int Dimensions,
641  typename Allocator>
642  friend auto detail::get_native_buffer(
643  const buffer<DataT, Dimensions, Allocator, void> &Obj)
644  -> backend_return_t<BackendName,
645  buffer<DataT, Dimensions, Allocator, void>>;
646 
647  template <backend BackendName>
648  backend_return_t<BackendName, buffer<T, dimensions, AllocatorT>>
649  getNative() const {
650  auto NativeHandles = impl->getNativeVector(BackendName);
651  return detail::BufferInterop<BackendName, T, dimensions,
652  AllocatorT>::GetNativeObjs(NativeHandles);
653  }
654 };
655 
656 #ifdef __cpp_deduction_guides
657 template <class InputIterator, class AllocatorT>
658 buffer(InputIterator, InputIterator, AllocatorT, const property_list & = {})
659  -> buffer<typename std::iterator_traits<InputIterator>::value_type, 1,
660  AllocatorT>;
661 template <class InputIterator>
662 buffer(InputIterator, InputIterator, const property_list & = {})
663  -> buffer<typename std::iterator_traits<InputIterator>::value_type, 1>;
664 template <class Container, class AllocatorT>
665 buffer(Container &, AllocatorT, const property_list & = {})
666  -> buffer<typename Container::value_type, 1, AllocatorT>;
667 template <class Container>
668 buffer(Container &, const property_list & = {})
669  -> buffer<typename Container::value_type, 1>;
670 template <class T, int dimensions, class AllocatorT>
671 buffer(const T *, const range<dimensions> &, AllocatorT,
672  const property_list & = {}) -> buffer<T, dimensions, AllocatorT>;
673 template <class T, int dimensions>
674 buffer(const T *, const range<dimensions> &, const property_list & = {})
675  -> buffer<T, dimensions>;
676 #endif // __cpp_deduction_guides
677 
678 } // namespace sycl
679 } // __SYCL_INLINE_NAMESPACE(cl)
680 
681 namespace std {
682 template <typename T, int dimensions, typename AllocatorT>
683 struct hash<cl::sycl::buffer<T, dimensions, AllocatorT>> {
684  size_t
686  return hash<std::shared_ptr<cl::sycl::detail::buffer_impl>>()(
688  }
689 };
690 } // namespace std
cl::sycl::buffer::buffer
buffer(const std::shared_ptr< T > &hostData, const range< dimensions > &bufferRange, AllocatorT allocator, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:196
cl::sycl::buffer::has_property
bool has_property() const
Definition: buffer.hpp:534
cl::sycl::backend
backend
Definition: backend_types.hpp:21
cl::sycl::buffer< char, 1 >::EnableIfItInputIterator
detail::enable_if_t< std::is_convertible< typename std::iterator_traits< It >::iterator_category, std::input_iterator_tag >::value > EnableIfItInputIterator
Definition: buffer.hpp:102
cl::sycl::buffer::buffer
buffer(const range< dimensions > &bufferRange, AllocatorT allocator, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:125
property_list.hpp
cl::sycl::buffer::get_range
range< dimensions > get_range() const
Definition: buffer.hpp:393
cl::sycl::buffer::get_access
accessor< T, dimensions, mode, target, access::placeholder::false_t, ext::oneapi::accessor_property_list<> > get_access(handler &commandGroupHandler, range< dimensions > accessRange, id< dimensions > accessOffset={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:432
cl::sycl::detail::aligned_allocator
Definition: aligned_allocator.hpp:23
cl::sycl::buffer::buffer
buffer(const range< dimensions > &bufferRange, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:113
cl::sycl::detail::getLinearIndex
size_t getLinearIndex(const T< Dims > &Index, const U< Dims > &Range)
Definition: common.hpp:336
cl::sycl::buffer::set_final_data
void set_final_data(Destination finalData=nullptr)
Definition: buffer.hpp:486
cl::sycl::buffer::buffer
buffer(Container &container, AllocatorT allocator, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:297
cl::sycl::distance
float distance(T p0, T p1) __NOEXC
Definition: builtins.hpp:1011
cl::sycl::backend_return_t
typename backend_traits< Backend >::template return_type< SyclType > backend_return_t
Definition: backend.hpp:72
T
cl::sycl::buffer::rangeToArray
std::array< size_t, 3 > rangeToArray(range< 2 > &r)
Definition: buffer.hpp:109
cl::sycl::event
An event object can be used to synchronize memory transfers, enqueues of kernels and signaling barrie...
Definition: event.hpp:32
cl::sycl::access::placeholder
placeholder
Definition: access.hpp:43
stl.hpp
cl::sycl::id
A unique identifier of an item in an index space.
Definition: array.hpp:17
cl::sycl::buffer::buffer
buffer(InputIterator first, InputIterator last, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:279
cl::sycl::buffer_allocator
detail::sycl_memory_object_allocator< char > buffer_allocator
Definition: buffer.hpp:32
cl::sycl::detail::make_buffer_helper
buffer< T, Dimensions, AllocatorT, void > make_buffer_helper(pi_native_handle Handle, const context &Ctx, event Evt={}, bool OwnNativeHandle=true)
Definition: buffer.hpp:49
accessor_property_list.hpp
cl::sycl::buffer::__SYCL2020_DEPRECATED
__SYCL2020_DEPRECATED("get_size() is deprecated, please use byte_size() instead") size_t get_size() const
Definition: buffer.hpp:399
cl::sycl::detail::remove_pointer_t
typename remove_pointer< T >::type remove_pointer_t
Definition: type_traits.hpp:275
sycl
Definition: invoke_simd.hpp:68
cl::sycl::detail::void_t
void void_t
Definition: stl_type_traits.hpp:42
cl::sycl::property_list
Objects of the property_list class are containers for the SYCL properties.
Definition: property_list.hpp:26
cl::sycl::detail::code_location
Definition: common.hpp:54
cl::sycl::buffer
Defines a shared array that can be used by kernels in queues.
Definition: buffer.hpp:78
cl::sycl::buffer< char, 1 >::EnableIfOneDimension
typename detail::enable_if_t< 1==dims > EnableIfOneDimension
Definition: buffer.hpp:90
cl::sycl::range
Defines the iteration domain of either a single work-group in a parallel dispatch,...
Definition: buffer.hpp:24
cl::sycl::info::queue
queue
Definition: info_desc.hpp:229
cl::sycl::buffer::size
size_t size() const noexcept
Definition: buffer.hpp:397
stl_type_traits.hpp
cl::sycl::range::size
size_t size() const
Definition: range.hpp:50
cl::sycl::buffer::buffer
buffer(const std::shared_ptr< T[]> &hostData, const range< dimensions > &bufferRange, AllocatorT allocator, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:212
cl::sycl::buffer::buffer
buffer(EnableIfSameNonConstIterators< T, _T > const *hostData, const range< dimensions > &bufferRange, AllocatorT allocator, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:181
char
cl::sycl::buffer::buffer
buffer(Container &container, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:316
cl::sycl::buffer::buffer
buffer(const buffer &rhs, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:361
cl::sycl::buffer::__SYCL2020_DEPRECATED
__SYCL2020_DEPRECATED("get_count() is deprecated, please use size() instead") size_t get_count() const
Definition: buffer.hpp:395
cl::sycl::detail::get_native_buffer
auto get_native_buffer(const buffer< DataT, Dimensions, Allocator, void > &Obj) -> backend_return_t< BackendName, buffer< DataT, Dimensions, Allocator, void >>
cl::sycl::buffer::buffer
buffer(InputIterator first, InputIterator last, AllocatorT allocator, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:261
cl::sycl::accessor
Buffer accessor.
Definition: accessor.hpp:224
cl::sycl::access::target
target
Definition: access.hpp:17
cl
We provide new interfaces for matrix muliply in this patch:
Definition: access.hpp:13
cl::sycl::buffer::rangeToArray
std::array< size_t, 3 > rangeToArray(range< 3 > &r)
Definition: buffer.hpp:107
cl::sycl::buffer::buffer
buffer(T *hostData, const range< dimensions > &bufferRange, AllocatorT allocator, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:151
cl::sycl::image_channel_order::r
@ r
cl::sycl::buffer< char, 1 >::EnableIfContiguous
detail::void_t< detail::enable_if_t< std::is_convertible< detail::remove_pointer_t< decltype(std::declval< Container >().data())>(*)[], const char(*)[]>::value >, decltype(std::declval< Container >().size())> EnableIfContiguous
Definition: buffer.hpp:98
cl::sycl::buffer::operator!=
bool operator!=(const buffer &rhs) const
Definition: buffer.hpp:387
cl::sycl::buffer::get_allocator
AllocatorT get_allocator() const
Definition: buffer.hpp:404
pi_native_handle
uintptr_t pi_native_handle
Definition: pi.h:98
cl::sycl::detail::getNextPowerOfTwo
constexpr size_t getNextPowerOfTwo(size_t Var)
Definition: common.hpp:330
cl::sycl::handler
Command group handler class.
Definition: handler.hpp:362
std::hash< cl::sycl::buffer< T, dimensions, AllocatorT > >::operator()
size_t operator()(const cl::sycl::buffer< T, dimensions, AllocatorT > &b) const
Definition: buffer.hpp:685
cl::sycl::ext::oneapi::accessor_property_list
Objects of the accessor_property_list class are containers for the SYCL properties.
Definition: property_list.hpp:19
cl::sycl::buffer::buffer
buffer(EnableIfSameNonConstIterators< T, _T > const *hostData, const range< dimensions > &bufferRange, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:166
cl::sycl::buffer::get_access
accessor< T, dimensions, mode, access::target::host_buffer, access::placeholder::false_t, ext::oneapi::accessor_property_list<> > get_access(const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:422
cl::sycl::buffer::buffer
buffer(T *hostData, const range< dimensions > &bufferRange, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:138
cl::sycl::buffer::get_property
propertyT get_property() const
Definition: buffer.hpp:538
cl::sycl::detail::getSyclObjImpl
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:204
cl::sycl::buffer::buffer
buffer(const std::shared_ptr< T[]> &hostData, const range< dimensions > &bufferRange, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:243
exception.hpp
cl::sycl::buffer::is_sub_buffer
bool is_sub_buffer() const
Definition: buffer.hpp:492
std
Definition: accessor.hpp:2617
cl::sycl::buffer::rangeToArray
std::array< size_t, 3 > rangeToArray(range< 1 > &r)
Definition: buffer.hpp:111
cl::sycl::buffer::reinterpret
buffer< ReinterpretT, ReinterpretDim, AllocatorT > reinterpret(range< ReinterpretDim > reinterpretRange) const
Definition: buffer.hpp:496
cl::sycl::buffer::get_access
accessor< T, dimensions, Mode, Target, access::placeholder::false_t, ext::oneapi::accessor_property_list<> > get_access(handler &CommandGroupHandler, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:411
cl::sycl::buffer::set_write_back
void set_write_back(bool flag=true)
Definition: buffer.hpp:490
cl::sycl::buffer::operator==
bool operator==(const buffer &rhs) const
Definition: buffer.hpp:385
cl::sycl::context
The context class represents a SYCL context on which kernel functions may be executed.
Definition: context.hpp:38
cl::sycl::buffer::get_access
accessor< T, dimensions, mode, access::target::host_buffer, access::placeholder::false_t, ext::oneapi::accessor_property_list<> > get_access(range< dimensions > accessRange, id< dimensions > accessOffset={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:449
cl::sycl::buffer::buffer
buffer(buffer &&rhs, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:370
cl::sycl::detail::SYCLMemObjAllocatorHolder
Definition: sycl_mem_obj_allocator.hpp:38
common.hpp
cl::sycl::buffer::reinterpret
std::enable_if<(sizeof(ReinterpretT)==sizeof(T)) &&(dimensions==ReinterpretDim), buffer< ReinterpretT, ReinterpretDim, AllocatorT > >::type reinterpret() const
Definition: buffer.hpp:512
cl::sycl::access::mode
mode
Definition: access.hpp:28
cl::sycl::buffer::reinterpret
std::enable_if<(ReinterpretDim==1) &&((dimensions !=ReinterpretDim)||(sizeof(ReinterpretT) !=sizeof(T))), buffer< ReinterpretT, ReinterpretDim, AllocatorT > >::type reinterpret() const
Definition: buffer.hpp:522
cl::sycl::buffer::byte_size
size_t byte_size() const noexcept
Definition: buffer.hpp:402
cl::sycl::detail::enable_if_t
typename std::enable_if< B, T >::type enable_if_t
Definition: stl_type_traits.hpp:24
cl::sycl::buffer::buffer
buffer(buffer< T, dimensions, AllocatorT > &b, const id< dimensions > &baseIndex, const range< dimensions > &subRange, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:320
cl::sycl::buffer< char, 1 >::EnableIfSameNonConstIterators
typename detail::enable_if_t< std::is_same< ItA, ItB >::value &&!std::is_const< ItA >::value, ItA > EnableIfSameNonConstIterators
Definition: buffer.hpp:105
cl::sycl::Dimensions
Dimensions
Definition: backend.hpp:142
cl::sycl::buffer::isOutOfBounds
bool isOutOfBounds(const id< dimensions > &offset, const range< dimensions > &newRange, const range< dimensions > &parentRange)
Definition: buffer.hpp:543
cl::sycl::buffer::buffer
buffer(const std::shared_ptr< T > &hostData, const range< dimensions > &bufferRange, const property_list &propList={}, const detail::code_location CodeLoc=detail::code_location::current())
Definition: buffer.hpp:228
cl::sycl::make_unique_ptr
std::unique_ptr< T > make_unique_ptr(ArgsT &&... Args)
Definition: stl.hpp:51
cl::sycl::errc::accessor
@ accessor
__SYCL_INLINE_NAMESPACE
#define __SYCL_INLINE_NAMESPACE(X)
Definition: defines_elementary.hpp:12
buffer_impl.hpp