28 template <
int dimensions>
class range;
30 template <
typename DataT>
33 template <
typename DataT,
int Dimensions, access::mode AccessMode>
36 template <
typename T,
int Dimensions,
typename AllocatorT,
typename Enable>
39 namespace ext::oneapi {
47 template <
typename T,
int Dimensions,
typename AllocatorT>
50 bool OwnNativeHandle =
true) {
59 buffer<DataT, Dimensions, Allocator, void>>;
62 typename AllocatorT = buffer_allocator<std::remove_const_t<DataT>>>
69 std::unique_ptr<detail::SYCLMemObjAllocator> Allocator);
71 buffer_plain(
void *HostData,
size_t SizeInBytes,
size_t RequiredAlign,
73 std::unique_ptr<detail::SYCLMemObjAllocator> Allocator);
75 buffer_plain(
const void *HostData,
size_t SizeInBytes,
size_t RequiredAlign,
77 std::unique_ptr<detail::SYCLMemObjAllocator> Allocator);
79 buffer_plain(
const std::shared_ptr<const void> &HostData,
80 const size_t SizeInBytes,
size_t RequiredAlign,
82 std::unique_ptr<detail::SYCLMemObjAllocator> Allocator,
88 const size_t SizeInBytes,
size_t RequiredAlign,
90 std::unique_ptr<detail::SYCLMemObjAllocator> Allocator,
94 std::unique_ptr<detail::SYCLMemObjAllocator> Allocator,
95 bool OwnNativeHandle,
event AvailableEvent);
97 buffer_plain(
const std::shared_ptr<detail::buffer_impl> &impl) : impl(impl) {}
99 void set_final_data_internal();
101 void set_final_data_internal(
102 const std::function<
void(
const std::function<
void(
void *
const Ptr)> &)>
105 void set_write_back(
bool NeedWriteBack);
108 void *UserObj,
const void *HostObj,
109 const void *Type, uint32_t Dim,
110 uint32_t ElemType,
size_t Range[3]);
112 template <
typename propertyT>
bool has_property() const noexcept;
114 template <typename propertyT> propertyT
get_property() const;
122 void addOrReplaceAccessorProperties(const
property_list &PropertyList);
124 size_t getSize() const;
126 void handleRelease() const;
141 template <typename T,
int dimensions = 1,
143 typename __Enabled = typename detail::
enable_if_t<(dimensions > 0) &&
149 static_assert(!std::is_same<T, std::string>::value,
150 "'std::string' is not a device copyable type");
160 template <
class Container>
164 decltype(std::declval<Container>().data())> (*)[],
165 const T (*)[]>::value>,
166 decltype(std::declval<Container>().size())>;
169 std::is_convertible<typename std::iterator_traits<It>::iterator_category,
170 std::input_iterator_tag>::value>;
171 template <
typename ItA,
typename ItB>
173 std::is_same<ItA, ItB>::value && !std::is_const<ItA>::value, ItA>;
184 : buffer_plain(bufferRange.
size() *
sizeof(
T),
190 CodeLoc, (
void *)impl.get(),
nullptr, (
const void *)
typeid(
T).name(),
191 dimensions,
sizeof(
T), rangeToArray(Range).data());
198 bufferRange.
size() *
sizeof(
T),
204 CodeLoc, (
void *)impl.get(),
nullptr, (
const void *)
typeid(
T).name(),
205 dimensions,
sizeof(
T), rangeToArray(Range).data());
211 : buffer_plain(hostData, bufferRange.
size() *
sizeof(
T),
217 CodeLoc, (
void *)impl.get(), hostData, (
const void *)
typeid(
T).name(),
218 dimensions,
sizeof(
T), rangeToArray(Range).data());
225 hostData, bufferRange.
size() *
sizeof(
T),
231 CodeLoc, (
void *)impl.get(), hostData, (
const void *)
typeid(
T).name(),
232 dimensions,
sizeof(
T), rangeToArray(Range).data());
235 template <
typename _T = T>
240 : buffer_plain(hostData, bufferRange.
size() *
sizeof(
T),
246 CodeLoc, (
void *)impl.get(), hostData, (
const void *)
typeid(
T).name(),
247 dimensions,
sizeof(
T), rangeToArray(Range).data());
250 template <
typename _T = T>
256 hostData, bufferRange.
size() *
sizeof(
T),
262 CodeLoc, (
void *)impl.get(), hostData, (
const void *)
typeid(
T).name(),
263 dimensions,
sizeof(
T), rangeToArray(Range).data());
266 buffer(
const std::shared_ptr<T> &hostData,
271 hostData, bufferRange.
size() *
sizeof(
T),
275 std::is_const<T>::value),
278 CodeLoc, (
void *)impl.get(), (
void *)hostData.get(),
279 (
const void *)
typeid(
T).name(), dimensions,
sizeof(
T),
280 rangeToArray(Range).data());
283 buffer(
const std::shared_ptr<T[]> &hostData,
288 hostData, bufferRange.
size() *
sizeof(
T),
292 std::is_const<T>::value),
295 CodeLoc, (
void *)impl.get(), (
void *)hostData.get(),
296 (
const void *)
typeid(
T).name(), dimensions,
sizeof(
T),
297 rangeToArray(Range).data());
300 buffer(
const std::shared_ptr<T> &hostData,
305 hostData, bufferRange.
size() *
sizeof(
T),
308 std::is_const<T>::value),
311 CodeLoc, (
void *)impl.get(), (
void *)hostData.get(),
312 (
const void *)
typeid(
T).name(), dimensions,
sizeof(
T),
313 rangeToArray(Range).data());
316 buffer(
const std::shared_ptr<T[]> &hostData,
321 hostData, bufferRange.
size() *
sizeof(
T),
324 std::is_const<T>::value),
327 CodeLoc, (
void *)impl.get(), (
void *)hostData.get(),
328 (
const void *)
typeid(
T).name(), dimensions,
sizeof(
T),
329 rangeToArray(Range).data());
332 template <
class InputIterator,
int N = dimensions,
333 typename = EnableIfOneDimension<N>,
334 typename = EnableIfItInputIterator<InputIterator>>
335 buffer(InputIterator first, InputIterator last, AllocatorT allocator,
340 [first, last](
void *ToPtr) {
344 using IteratorValueType =
346 using IteratorNonConstValueType =
348 using IteratorPointerToNonConstValueType =
350 std::copy(first, last,
351 static_cast<IteratorPointerToNonConstValueType
>(ToPtr));
353 std::distance(first, last) *
sizeof(T),
357 detail::iterator_to_const_type_t<InputIterator>::value),
358 Range(range<1>(std::distance(first, last))) {
359 size_t r[3] = {Range[0], 0, 0};
361 (
const void *)
typeid(T).name(),
362 dimensions,
sizeof(T), r);
365 template <
class InputIterator,
int N = dimensions,
366 typename = EnableIfOneDimension<N>,
367 typename = EnableIfItInputIterator<InputIterator>>
368 buffer(InputIterator first, InputIterator last,
373 [first, last](
void *ToPtr) {
377 using IteratorValueType =
379 using IteratorNonConstValueType =
381 using IteratorPointerToNonConstValueType =
383 std::copy(first, last,
384 static_cast<IteratorPointerToNonConstValueType
>(ToPtr));
386 std::distance(first, last) *
sizeof(T),
389 detail::iterator_to_const_type_t<InputIterator>::value),
390 Range(range<1>(std::distance(first, last))) {
391 size_t r[3] = {Range[0], 0, 0};
393 (
const void *)
typeid(T).name(),
394 dimensions,
sizeof(T), r);
398 template <
class Container,
int N = dimensions,
399 typename = EnableIfOneDimension<N>,
400 typename = EnableIfContiguous<Container>>
401 buffer(Container &container, AllocatorT allocator,
405 container.data(), container.size() *
sizeof(
T),
410 size_t r[3] = {Range[0], 0, 0};
412 CodeLoc, (
void *)impl.get(), container.data(),
413 (
const void *)
typeid(T).name(), dimensions,
sizeof(T), r);
417 template <
class Container,
int N = dimensions,
418 typename = EnableIfOneDimension<N>,
419 typename = EnableIfContiguous<Container>>
422 :
buffer(container, {}, propList, CodeLoc) {}
427 : buffer_plain(b.impl), Range(subRange),
428 OffsetInBytes(getOffsetInBytes<T>(baseIndex, b.Range)),
431 CodeLoc, (
void *)impl.get(), impl.get(), (
const void *)
typeid(T).name(),
432 dimensions,
sizeof(T), rangeToArray(Range).data());
434 if (b.is_sub_buffer())
435 throw sycl::invalid_object_error(
436 "Cannot create sub buffer from sub buffer.", PI_ERROR_INVALID_VALUE);
437 if (isOutOfBounds(baseIndex, subRange, b.Range))
438 throw sycl::invalid_object_error(
439 "Requested sub-buffer size exceeds the size of the parent buffer",
440 PI_ERROR_INVALID_VALUE);
441 if (!isContiguousRegion(baseIndex, subRange, b.Range))
442 throw sycl::invalid_object_error(
443 "Requested sub-buffer region is not contiguous",
444 PI_ERROR_INVALID_VALUE);
449 : buffer_plain(rhs.impl), Range(rhs.Range),
450 OffsetInBytes(rhs.OffsetInBytes), IsSubBuffer(rhs.IsSubBuffer) {
452 CodeLoc, (
void *)impl.get(), impl.get(), (
const void *)
typeid(T).name(),
453 dimensions,
sizeof(T), rangeToArray(Range).data());
458 : buffer_plain(
std::move(rhs.impl)), Range(rhs.Range),
459 OffsetInBytes(rhs.OffsetInBytes), IsSubBuffer(rhs.IsSubBuffer) {
461 CodeLoc, (
void *)impl.get(), impl.get(), (
const void *)
typeid(T).name(),
462 dimensions,
sizeof(T), rangeToArray(Range).data());
482 size_t get_count()
const {
return size(); }
483 size_t size() const noexcept {
return Range.size(); }
486 "get_size() is deprecated, please use byte_size() instead")
487 size_t get_size()
const {
return byte_size(); }
488 size_t byte_size() const noexcept {
return size() *
sizeof(T); }
491 return buffer_plain::get_allocator_internal()
492 ->template getAllocator<AllocatorT>();
495 template <access::mode Mode, access::target Target = access::target::device>
496 accessor<
T, dimensions, Mode, Target, access::placeholder::false_t,
501 return accessor<T, dimensions, Mode, Target, access::placeholder::false_t,
503 *
this, CommandGroupHandler, {}, CodeLoc);
506 template <access::mode mode>
508 "use get_host_access instead")
513 accessor_property_list<>> get_access(const detail::code_location
515 detail::code_location::
517 return accessor<T, dimensions,
mode, access::target::host_buffer,
518 access::placeholder::false_t,
522 template <access::mode mode, access::target target = access::target::device>
529 if (isOutOfBounds(accessOffset, accessRange, this->Range))
530 throw sycl::invalid_object_error(
531 "Requested accessor would exceed the bounds of the buffer",
532 PI_ERROR_INVALID_VALUE);
536 *
this, commandGroupHandler, accessRange, accessOffset, {}, CodeLoc);
539 template <access::mode mode>
541 "use get_host_access instead")
543 T, dimensions,
mode, access::
target::host_buffer,
546 accessor_property_list<>> get_access(
range<dimensions> accessRange,
547 id<dimensions> accessOffset = {},
550 detail::code_location::
552 if (isOutOfBounds(accessOffset, accessRange, this->Range))
553 throw sycl::invalid_object_error(
554 "Requested accessor would exceed the bounds of the buffer",
555 PI_ERROR_INVALID_VALUE);
557 return accessor<
T, dimensions,
mode, access::target::host_buffer,
558 access::placeholder::false_t,
560 *
this, accessRange, accessOffset, {}, CodeLoc);
563 template <
typename... Ts>
auto get_access(Ts... args) {
567 template <
typename... Ts>
568 auto get_access(
handler &commandGroupHandler, Ts... args) {
569 return accessor{*
this, commandGroupHandler, args...};
572 template <
typename... Ts>
auto get_host_access(Ts... args) {
576 template <
typename... Ts>
577 auto get_host_access(
handler &commandGroupHandler, Ts... args) {
581 template <
typename Destination = std::
nullptr_t>
582 void set_final_data(Destination finalData =
nullptr) {
583 this->set_final_data_internal(finalData);
586 void set_final_data_internal(std::nullptr_t) {
587 buffer_plain::set_final_data_internal();
590 template <
template <
typename WeakT>
class WeakPtrT,
typename WeakT>
592 std::is_convertible<WeakPtrT<WeakT>, std::weak_ptr<WeakT>>::value>
593 set_final_data_internal(WeakPtrT<WeakT> FinalData) {
594 std::weak_ptr<WeakT> TempFinalData(FinalData);
595 this->set_final_data_internal(TempFinalData);
598 template <
typename WeakT>
599 void set_final_data_internal(std::weak_ptr<WeakT> FinalData) {
600 buffer_plain::set_final_data_internal(
601 [FinalData](
const std::function<
void(
void *
const Ptr)> &F) {
602 if (std::shared_ptr<WeakT> LockedFinalData = FinalData.lock())
603 F(LockedFinalData.get());
607 template <
typename Destination>
609 set_final_data_internal(Destination FinalData) {
611 buffer_plain::set_final_data_internal();
613 buffer_plain::set_final_data_internal(
614 [FinalData](
const std::function<
void(
void *
const Ptr)> &F) {
619 template <
typename Destination>
621 set_final_data_internal(Destination FinalData) {
622 const size_t Size = size();
623 buffer_plain::set_final_data_internal(
624 [FinalData, Size](
const std::function<
void(
void *
const Ptr)> &F) {
629 std::unique_ptr<DestinationValueT[]> ContiguousStorage(
630 new DestinationValueT[Size]);
631 F(ContiguousStorage.get());
632 std::copy(ContiguousStorage.get(), ContiguousStorage.get() + Size,
637 void set_final_data(std::nullptr_t) {
638 buffer_plain::set_final_data_internal();
641 void set_write_back(
bool flag =
true) { buffer_plain::set_write_back(flag); }
643 bool is_sub_buffer()
const {
return IsSubBuffer; }
645 template <
typename Re
interpretT,
int Re
interpretDim>
646 buffer<ReinterpretT, ReinterpretDim,
647 typename std::allocator_traits<AllocatorT>::template rebind_alloc<
648 std::remove_const_t<ReinterpretT>>>
650 if (
sizeof(ReinterpretT) * reinterpretRange.
size() != byte_size())
651 throw sycl::invalid_object_error(
652 "Total size in bytes represented by the type and range of the "
653 "reinterpreted SYCL buffer does not equal the total size in bytes "
654 "represented by the type and range of this SYCL buffer",
655 PI_ERROR_INVALID_VALUE);
657 return buffer<ReinterpretT, ReinterpretDim,
658 typename std::allocator_traits<AllocatorT>::
659 template rebind_alloc<std::remove_const_t<ReinterpretT>>>(
660 impl, reinterpretRange, OffsetInBytes, IsSubBuffer);
663 template <
typename Re
interpretT,
int Re
interpretDim = dimensions>
664 typename std::enable_if<
665 (
sizeof(ReinterpretT) ==
sizeof(
T)) && (dimensions == ReinterpretDim),
666 buffer<ReinterpretT, ReinterpretDim,
667 typename std::allocator_traits<AllocatorT>::template rebind_alloc<
668 std::remove_const_t<ReinterpretT>>>>::type
669 reinterpret()
const {
670 return buffer<ReinterpretT, ReinterpretDim,
671 typename std::allocator_traits<AllocatorT>::
672 template rebind_alloc<std::remove_const_t<ReinterpretT>>>(
673 impl, get_range(), OffsetInBytes, IsSubBuffer);
676 template <
typename Re
interpretT,
int Re
interpretDim = dimensions>
677 typename std::enable_if<
678 (ReinterpretDim == 1) && ((dimensions != ReinterpretDim) ||
679 (
sizeof(ReinterpretT) !=
sizeof(T))),
681 reinterpret()
const {
682 long sz = byte_size();
683 if (sz %
sizeof(ReinterpretT) != 0)
684 throw sycl::invalid_object_error(
685 "Total byte size of buffer is not evenly divisible by the size of "
686 "the reinterpreted type",
687 PI_ERROR_INVALID_VALUE);
690 impl,
range<1>{sz /
sizeof(ReinterpretT)}, OffsetInBytes, IsSubBuffer);
693 template <
typename propertyT>
bool has_property() const noexcept {
694 return buffer_plain::template has_property<propertyT>();
697 template <
typename propertyT> propertyT
get_property()
const {
698 return buffer_plain::template get_property<propertyT>();
705 bool outOfBounds =
false;
706 for (
int i = 0; i < dimensions; ++i)
707 outOfBounds |= newRange[i] + offset[i] > parentRange[i];
715 template <
typename A,
int dims,
typename C,
typename Enable>
720 template <
typename HT,
int HDims,
typename HAllocT>
728 size_t OffsetInBytes = 0;
729 bool IsSubBuffer =
false;
732 template <
int N = dimensions,
typename = EnableIfOneDimension<N>>
734 bool OwnNativeHandle,
event AvailableEvent = {},
737 MemObject, SyclContext,
739 OwnNativeHandle, std::move(AvailableEvent)),
742 Range[0] = buffer_plain::getSize() /
sizeof(T);
744 CodeLoc, (
void *)impl.get(), &MemObject, (
const void *)
typeid(T).name(),
745 dimensions,
sizeof(T), rangeToArray(Range).data());
748 void addOrReplaceAccessorProperties(
const property_list &PropertyList) {
749 buffer_plain::addOrReplaceAccessorProperties(PropertyList);
753 buffer_plain::deleteAccProps(Kind);
757 buffer(
const std::shared_ptr<detail::buffer_impl> &Impl,
758 range<dimensions> reinterpretRange,
size_t reinterpretOffset,
760 const detail::code_location CodeLoc = detail::code_location::current())
761 : buffer_plain(Impl), Range(reinterpretRange),
762 OffsetInBytes(reinterpretOffset), IsSubBuffer(isSubBuffer) {
764 CodeLoc, (
void *)impl.get(), Impl.get(), (
const void *)
typeid(T).name(),
765 dimensions,
sizeof(T), rangeToArray(Range).data());
768 template <
typename Type,
int N>
769 size_t getOffsetInBytes(
const id<N> &offset,
const range<N> &range) {
773 bool isContiguousRegion(
const id<1> &,
const range<1> &,
const range<1> &) {
778 bool isContiguousRegion(
const id<2> &offset,
const range<2> &newRange,
779 const range<2> &parentRange) {
787 return newRange[0] == 1;
788 return newRange[1] == parentRange[1];
791 bool isContiguousRegion(
const id<3> &offset,
const range<3> &newRange,
792 const range<3> &parentRange) {
802 return newRange[0] == 1 && newRange[1] == 1;
804 return newRange[0] == 1 && newRange[2] == parentRange[2];
805 return newRange[1] == parentRange[1] && newRange[2] == parentRange[2];
811 const buffer<DataT, Dimensions, Allocator, void> &Obj)
813 buffer<DataT, Dimensions, Allocator, void>>;
815 template <backend BackendName>
816 backend_return_t<BackendName, buffer<T, dimensions, AllocatorT>>
818 auto NativeHandles = buffer_plain::getNativeVector(BackendName);
819 return detail::BufferInterop<BackendName, T, dimensions,
820 AllocatorT>::GetNativeObjs(NativeHandles);
824 #ifdef __cpp_deduction_guides
825 template <
class InputIterator,
class AllocatorT>
826 buffer(InputIterator, InputIterator, AllocatorT,
const property_list & = {})
827 -> buffer<
typename std::iterator_traits<InputIterator>::value_type, 1,
829 template <
class InputIterator>
830 buffer(InputIterator, InputIterator,
const property_list & = {})
831 -> buffer<
typename std::iterator_traits<InputIterator>::value_type, 1>;
832 template <
class Container,
class AllocatorT>
833 buffer(Container &, AllocatorT,
const property_list & = {})
834 -> buffer<typename Container::value_type, 1, AllocatorT>;
835 template <
class Container>
836 buffer(Container &,
const property_list & = {})
837 -> buffer<typename Container::value_type, 1>;
838 template <
class T,
int dimensions,
class AllocatorT>
839 buffer(
const T *,
const range<dimensions> &, AllocatorT,
840 const property_list & = {}) -> buffer<T, dimensions, AllocatorT>;
841 template <
class T,
int dimensions>
842 buffer(
const T *,
const range<dimensions> &,
const property_list & = {})
843 -> buffer<T, dimensions>;
844 #endif // __cpp_deduction_guides
850 template <
typename T,
int dimensions,
typename AllocatorT>
851 struct hash<
sycl::buffer<T, dimensions, AllocatorT>> {
852 size_t operator()(
const sycl::buffer<T, dimensions, AllocatorT> &b)
const {
853 return hash<std::shared_ptr<sycl::detail::buffer_impl>>()(