DPC++ Runtime
Runtime libraries for oneAPI DPC++
accessor_image.hpp
Go to the documentation of this file.
1 #pragma once
2 
3 #include <sycl/accessor.hpp>
5 #include <sycl/device.hpp>
6 #include <sycl/image.hpp>
7 
8 #ifdef __SYCL_DEVICE_ONLY__
10 #endif
11 
12 namespace sycl {
13 inline namespace _V1 {
14 namespace detail {
15 template <int Dim, typename T> struct IsValidCoordDataT;
16 template <typename T> struct IsValidCoordDataT<1, T> {
17  constexpr static bool value = detail::is_contained<
19 };
20 template <typename T> struct IsValidCoordDataT<2, T> {
21  constexpr static bool value = detail::is_contained<
23  vec<opencl::cl_float, 2>>>::type::value;
24 };
25 template <typename T> struct IsValidCoordDataT<3, T> {
26  constexpr static bool value = detail::is_contained<
28  vec<opencl::cl_float, 4>>>::type::value;
29 };
30 
31 template <int Dim, typename T> struct IsValidUnsampledCoord2020DataT;
32 template <typename T> struct IsValidUnsampledCoord2020DataT<1, T> {
33  constexpr static bool value = std::is_same_v<T, int>;
34 };
35 template <typename T> struct IsValidUnsampledCoord2020DataT<2, T> {
36  constexpr static bool value = std::is_same_v<T, int2>;
37 };
38 template <typename T> struct IsValidUnsampledCoord2020DataT<3, T> {
39  constexpr static bool value = std::is_same_v<T, int4>;
40 };
41 
42 template <int Dim, typename T> struct IsValidSampledCoord2020DataT;
43 template <typename T> struct IsValidSampledCoord2020DataT<1, T> {
44  constexpr static bool value = std::is_same_v<T, float>;
45 };
46 template <typename T> struct IsValidSampledCoord2020DataT<2, T> {
47  constexpr static bool value = std::is_same_v<T, float2>;
48 };
49 template <typename T> struct IsValidSampledCoord2020DataT<3, T> {
50  constexpr static bool value = std::is_same_v<T, float4>;
51 };
52 
53 void __SYCL_EXPORT unsampledImageConstructorNotification(
54  void *ImageObj, void *AccessorObj,
55  const std::optional<image_target> &Target, access::mode Mode,
56  const void *Type, uint32_t ElemSize, const code_location &CodeLoc);
57 
58 void __SYCL_EXPORT sampledImageConstructorNotification(
59  void *ImageObj, void *AccessorObj,
60  const std::optional<image_target> &Target, const void *Type,
61  uint32_t ElemSize, const code_location &CodeLoc);
62 
66  std::shared_ptr<UnsampledImageAccessorImplHost>;
68  std::shared_ptr<SampledImageAccessorImplHost>;
69 
70 void __SYCL_EXPORT
72 void __SYCL_EXPORT
74 
75 class __SYCL_EXPORT UnsampledImageAccessorBaseHost {
76 protected:
78  : impl{Impl} {}
79 
80 public:
82  void *SYCLMemObject, int Dims, int ElemSize,
83  id<3> Pitch, image_channel_type ChannelType,
84  image_channel_order ChannelOrder,
85  const property_list &PropertyList = {});
86  const sycl::range<3> &getSize() const;
87  void *getMemoryObject() const;
88  detail::AccHostDataT &getAccData();
89  void *getPtr();
90  void *getPtr() const;
91  int getNumOfDims() const;
92  int getElementSize() const;
93  id<3> getPitch() const;
94  image_channel_type getChannelType() const;
95  image_channel_order getChannelOrder() const;
96  const property_list &getPropList() const;
97 
98 protected:
99  template <class Obj>
100  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
101 
102  template <class T>
103  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
104 
106 
107  // The function references helper methods required by GDB pretty-printers
109 #ifndef NDEBUG
110  const auto *this_const = this;
111  (void)getSize();
112  (void)this_const->getSize();
113  (void)getPtr();
114  (void)this_const->getPtr();
115 #endif
116  }
117 
118 #ifndef __SYCL_DEVICE_ONLY__
119  // Reads a pixel of the underlying image at the specified coordinate. It is
120  // the responsibility of the caller to ensure that the coordinate type is
121  // valid.
122  template <typename DataT, typename CoordT>
123  DataT read(const CoordT &Coords) const noexcept {
127  return imageReadSamplerHostImpl<CoordT, DataT>(
128  Coords, Smpl, getSize(), getPitch(), getChannelType(),
129  getChannelOrder(), getPtr(), getElementSize());
130  }
131 
132  // Writes to a pixel of the underlying image at the specified coordinate. It
133  // is the responsibility of the caller to ensure that the coordinate type is
134  // valid.
135  template <typename DataT, typename CoordT>
136  void write(const CoordT &Coords, const DataT &Color) const {
137  imageWriteHostImpl(Coords, Color, getPitch(), getElementSize(),
138  getChannelType(), getChannelOrder(), getPtr());
139  }
140 #endif
141 };
142 
143 class __SYCL_EXPORT SampledImageAccessorBaseHost {
144 protected:
146  : impl{Impl} {}
147 
148 public:
149  SampledImageAccessorBaseHost(sycl::range<3> Size, void *SYCLMemObject,
150  int Dims, int ElemSize, id<3> Pitch,
151  image_channel_type ChannelType,
152  image_channel_order ChannelOrder,
153  image_sampler Sampler,
154  const property_list &PropertyList = {});
155  const sycl::range<3> &getSize() const;
156  void *getMemoryObject() const;
157  detail::AccHostDataT &getAccData();
158  void *getPtr();
159  void *getPtr() const;
160  int getNumOfDims() const;
161  int getElementSize() const;
162  id<3> getPitch() const;
163  image_channel_type getChannelType() const;
164  image_channel_order getChannelOrder() const;
165  image_sampler getSampler() const;
166  const property_list &getPropList() const;
167 
168 protected:
169  template <class Obj>
170  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
171 
172  template <class T>
173  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
174 
176 
177  // The function references helper methods required by GDB pretty-printers
179 #ifndef NDEBUG
180  const auto *this_const = this;
181  (void)getSize();
182  (void)this_const->getSize();
183  (void)getPtr();
184  (void)this_const->getPtr();
185 #endif
186  }
187 
188 #ifndef __SYCL_DEVICE_ONLY__
189  // Reads a pixel of the underlying image at the specified coordinate. It is
190  // the responsibility of the caller to ensure that the coordinate type is
191  // valid.
192  template <typename DataT, typename CoordT>
193  DataT read(const CoordT &Coords) const {
194  return imageReadSamplerHostImpl<CoordT, DataT>(
195  Coords, getSampler(), getSize(), getPitch(), getChannelType(),
196  getChannelOrder(), getPtr(), getElementSize());
197  }
198 #endif
199 };
200 
201 template <typename DataT, int Dimensions, access::mode AccessMode,
203 class __image_array_slice__;
204 
205 // Image accessor
206 template <typename DataT, int Dimensions, access::mode AccessMode,
209 #ifndef __SYCL_DEVICE_ONLY__
210  : public detail::AccessorBaseHost {
211  size_t MImageCount;
212  image_channel_order MImgChannelOrder;
213  image_channel_type MImgChannelType;
214 #else
215 {
216 
217  using OCLImageTy = typename detail::opencl_image_type<Dimensions, AccessMode,
218  AccessTarget>::type;
219  OCLImageTy MImageObj;
220  char MPadding[sizeof(detail::AccessorBaseHost) +
221  sizeof(size_t /*MImageCount*/) + sizeof(image_channel_order) +
222  sizeof(image_channel_type) - sizeof(OCLImageTy)];
223 
224 protected:
225  void imageAccessorInit(OCLImageTy Image) { MImageObj = Image; }
226 
227 private:
228 #endif
229  template <typename T1, int T2, access::mode T3, access::placeholder T4>
230  friend class __image_array_slice__;
231 
232  constexpr static bool IsHostImageAcc =
233  (AccessTarget == access::target::host_image);
234 
235  constexpr static bool IsImageAcc = (AccessTarget == access::target::image);
236 
237  constexpr static bool IsImageArrayAcc =
238  (AccessTarget == access::target::image_array);
239 
240  constexpr static bool IsImageAccessWriteOnly =
243 
244  constexpr static bool IsImageAccessAnyWrite =
245  (IsImageAccessWriteOnly || AccessMode == access::mode::read_write);
246 
247  constexpr static bool IsImageAccessReadOnly =
249 
250  constexpr static bool IsImageAccessAnyRead =
251  (IsImageAccessReadOnly || AccessMode == access::mode::read_write);
252 
253  static_assert(std::is_same_v<DataT, vec<opencl::cl_int, 4>> ||
254  std::is_same_v<DataT, vec<opencl::cl_uint, 4>> ||
255  std::is_same_v<DataT, vec<opencl::cl_float, 4>> ||
256  std::is_same_v<DataT, vec<opencl::cl_half, 4>>,
257  "The data type of an image accessor must be only cl_int4, "
258  "cl_uint4, cl_float4 or cl_half4 from SYCL namespace");
259 
260  static_assert(IsImageAcc || IsHostImageAcc || IsImageArrayAcc,
261  "Expected image type");
262 
264  "Expected false as Placeholder value for image accessor.");
265 
266  static_assert(
267  ((IsImageAcc || IsImageArrayAcc) &&
268  (IsImageAccessWriteOnly || IsImageAccessReadOnly)) ||
269  (IsHostImageAcc && (IsImageAccessAnyWrite || IsImageAccessAnyRead)),
270  "Access modes can be only read/write/discard_write for image/image_array "
271  "target accessor, or they can be only "
272  "read/write/discard_write/read_write for host_image target accessor.");
273 
274  static_assert(Dimensions > 0 && Dimensions <= 3,
275  "Dimensions can be 1/2/3 for image accessor.");
276 
277 #ifdef __SYCL_DEVICE_ONLY__
278 
279  sycl::vec<int, Dimensions> getRangeInternal() const {
280  return __invoke_ImageQuerySize<sycl::vec<int, Dimensions>, OCLImageTy>(
281  MImageObj);
282  }
283 
284  size_t getElementSize() const {
285  int ChannelType = __invoke_ImageQueryFormat<int, OCLImageTy>(MImageObj);
286  int ChannelOrder = __invoke_ImageQueryOrder<int, OCLImageTy>(MImageObj);
287  int ElementSize = getSPIRVElementSize(ChannelType, ChannelOrder);
288  return ElementSize;
289  }
290 
291 #else
292 
293  sycl::vec<int, Dimensions> getRangeInternal() const {
294  // TODO: Implement for host.
295  throw sycl::exception(
297  "image::getRangeInternal() is not implemented for host");
298  return sycl::vec<int, Dimensions>{1};
299  }
300 
301 #endif
302 
303 #ifndef __SYCL_DEVICE_ONLY__
304 protected:
306 #endif // __SYCL_DEVICE_ONLY__
307 
308 private:
310 
311 #ifdef __SYCL_DEVICE_ONLY__
312  const OCLImageTy getNativeImageObj() const { return MImageObj; }
313 #endif // __SYCL_DEVICE_ONLY__
314 
315 public:
316  using value_type = DataT;
317  using reference = DataT &;
318  using const_reference = const DataT &;
319 
320  // image_accessor Constructors.
321 
322 #ifdef __SYCL_DEVICE_ONLY__
323  // Default constructor for objects later initialized with __init member.
324  image_accessor() {}
325 #endif
326 
327  // Available only when: accessTarget == access::target::host_image
328  // template <typename AllocatorT>
329  // accessor(image<dimensions, AllocatorT> &imageRef);
330  template <
331  typename AllocatorT, int Dims = Dimensions,
332  typename = std::enable_if_t<(Dims > 0 && Dims <= 3) && IsHostImageAcc>>
333  image_accessor(image<Dims, AllocatorT> &ImageRef, int ImageElementSize)
334 #ifdef __SYCL_DEVICE_ONLY__
335  {
336  (void)ImageRef;
337  (void)ImageElementSize;
338  // No implementation needed for device. The constructor is only called by
339  // host.
340  }
341 #else
342  : AccessorBaseHost({ImageRef.getRowPitch(), ImageRef.getSlicePitch(), 0},
343  detail::convertToArrayOfN<3, 1>(ImageRef.get_range()),
344  detail::convertToArrayOfN<3, 1>(ImageRef.get_range()),
345  AccessMode, detail::getSyclObjImpl(ImageRef).get(),
346  Dimensions, ImageElementSize, size_t(0)),
347  MImageCount(ImageRef.size()),
348  MImgChannelOrder(ImageRef.getChannelOrder()),
349  MImgChannelType(ImageRef.getChannelType()) {
351  }
352 #endif
353 
354  // Available only when: accessTarget == access::target::image
355  // template <typename AllocatorT>
356  // accessor(image<dimensions, AllocatorT> &imageRef,
357  // handler &commandGroupHandlerRef);
358  template <typename AllocatorT, int Dims = Dimensions,
359  typename = std::enable_if_t<(Dims > 0 && Dims <= 3) && IsImageAcc>>
361  handler &CommandGroupHandlerRef, int ImageElementSize)
362 #ifdef __SYCL_DEVICE_ONLY__
363  {
364  (void)ImageRef;
365  (void)CommandGroupHandlerRef;
366  (void)ImageElementSize;
367  // No implementation needed for device. The constructor is only called by
368  // host.
369  }
370 #else
371  : AccessorBaseHost({ImageRef.getRowPitch(), ImageRef.getSlicePitch(), 0},
372  detail::convertToArrayOfN<3, 1>(ImageRef.get_range()),
373  detail::convertToArrayOfN<3, 1>(ImageRef.get_range()),
374  AccessMode, detail::getSyclObjImpl(ImageRef).get(),
375  Dimensions, ImageElementSize, size_t(0)),
376  MImageCount(ImageRef.size()),
377  MImgChannelOrder(ImageRef.getChannelOrder()),
378  MImgChannelType(ImageRef.getChannelType()) {
379 
380  device Device = getDeviceFromHandler(CommandGroupHandlerRef);
381  if (!Device.has(aspect::ext_intel_legacy_image))
382  throw sycl::exception(
383  sycl::errc::feature_not_supported,
384  "SYCL 1.2.1 images are not supported by this device.");
385  }
386 #endif
387 
388  /* -- common interface members -- */
389 
390  // operator == and != need to be defined only for host application as per the
391  // SYCL spec 1.2.1
392 #ifndef __SYCL_DEVICE_ONLY__
393  bool operator==(const image_accessor &Rhs) const { return Rhs.impl == impl; }
394 #else
395  // The operator with __SYCL_DEVICE_ONLY__ need to be declared for compilation
396  // of host application with device compiler.
397  // Usage of this operator inside the kernel code will give a runtime failure.
398  bool operator==(const image_accessor &Rhs) const;
399 #endif
400 
401  bool operator!=(const image_accessor &Rhs) const { return !(Rhs == *this); }
402 
403  // get_count() method : Returns the number of elements of the SYCL image this
404  // SYCL accessor is accessing.
405  //
406  // get_range() method : Returns a range object which represents the number of
407  // elements of dataT per dimension that this accessor may access.
408  // The range object returned must equal to the range of the image this
409  // accessor is associated with.
410 
411 #ifdef __SYCL_DEVICE_ONLY__
412 
413  __SYCL2020_DEPRECATED("get_count() is deprecated, please use size() instead")
414  size_t get_count() const { return size(); }
415  size_t size() const noexcept { return get_range<Dimensions>().size(); }
416 
417  template <int Dims = Dimensions, typename = std::enable_if_t<Dims == 1>>
418  range<1> get_range() const {
419  int Range = getRangeInternal();
420  return range<1>(Range);
421  }
422  template <int Dims = Dimensions, typename = std::enable_if_t<Dims == 2>>
423  range<2> get_range() const {
424  int2 Range = getRangeInternal();
425  return range<2>(Range[0], Range[1]);
426  }
427  template <int Dims = Dimensions, typename = std::enable_if_t<Dims == 3>>
428  range<3> get_range() const {
429  int3 Range = getRangeInternal();
430  return range<3>(Range[0], Range[1], Range[2]);
431  }
432 
433 #else
434  __SYCL2020_DEPRECATED("get_count() is deprecated, please use size() instead")
435  size_t get_count() const { return size(); };
436  size_t size() const noexcept { return MImageCount; };
437 
438  template <int Dims = Dimensions, typename = std::enable_if_t<(Dims > 0)>>
440  return detail::convertToArrayOfN<Dims, 1>(getAccessRange());
441  }
442 
443 #endif
444 
445  // Available only when:
446  // (accessTarget == access::target::image && accessMode == access::mode::read)
447  // || (accessTarget == access::target::host_image && ( accessMode ==
448  // access::mode::read || accessMode == access::mode::read_write))
449  template <typename CoordT, int Dims = Dimensions,
450  typename = std::enable_if_t<
451  (Dims > 0) && (IsValidCoordDataT<Dims, CoordT>::value) &&
452  (detail::is_genint_v<CoordT>) &&
453  ((IsImageAcc && IsImageAccessReadOnly) ||
454  (IsHostImageAcc && IsImageAccessAnyRead))>>
455  DataT read(const CoordT &Coords) const {
456 #ifdef __SYCL_DEVICE_ONLY__
457  return __invoke__ImageRead<DataT, OCLImageTy, CoordT>(MImageObj, Coords);
458 #else
461  return read<CoordT, Dims>(Coords, Smpl);
462 #endif
463  }
464 
465  // Available only when:
466  // (accessTarget == access::target::image && accessMode == access::mode::read)
467  // || (accessTarget == access::target::host_image && ( accessMode ==
468  // access::mode::read || accessMode == access::mode::read_write))
469  template <typename CoordT, int Dims = Dimensions,
470  typename = std::enable_if_t<
471  (Dims > 0) && (IsValidCoordDataT<Dims, CoordT>::value) &&
472  ((IsImageAcc && IsImageAccessReadOnly) ||
473  (IsHostImageAcc && IsImageAccessAnyRead))>>
474  DataT read(const CoordT &Coords, const sampler &Smpl) const {
475 #ifdef __SYCL_DEVICE_ONLY__
476  return __invoke__ImageReadSampler<DataT, OCLImageTy, CoordT>(
477  MImageObj, Coords, Smpl.impl.m_Sampler);
478 #else
479  return imageReadSamplerHostImpl<CoordT, DataT>(
480  Coords, Smpl, getAccessRange() /*Image Range*/,
481  getOffset() /*Image Pitch*/, MImgChannelType, MImgChannelOrder,
482  AccessorBaseHost::getPtr() /*ptr to image*/,
484 #endif
485  }
486 
487  // Available only when:
488  // (accessTarget == access::target::image && (accessMode ==
489  // access::mode::write || accessMode == access::mode::discard_write)) ||
490  // (accessTarget == access::target::host_image && (accessMode ==
491  // access::mode::write || accessMode == access::mode::discard_write ||
492  // accessMode == access::mode::read_write))
493  template <
494  typename CoordT, int Dims = Dimensions,
495  typename = std::enable_if_t<(Dims > 0) && (detail::is_genint_v<CoordT>) &&
497  ((IsImageAcc && IsImageAccessWriteOnly) ||
498  (IsHostImageAcc && IsImageAccessAnyWrite))>>
499  void write(const CoordT &Coords, const DataT &Color) const {
500 #ifdef __SYCL_DEVICE_ONLY__
501  __invoke__ImageWrite<OCLImageTy, CoordT, DataT>(MImageObj, Coords, Color);
502 #else
503  imageWriteHostImpl(Coords, Color, getOffset() /*ImagePitch*/,
504  AccessorBaseHost::getElemSize(), MImgChannelType,
505  MImgChannelOrder,
506  AccessorBaseHost::getPtr() /*Ptr to Image*/);
507 #endif
508  }
509 };
510 
511 template <typename DataT, int Dimensions, access::mode AccessMode,
514 
515  static_assert(Dimensions < 3,
516  "Image slice cannot have more then 2 dimensions");
517 
518  constexpr static int AdjustedDims = (Dimensions == 2) ? 4 : Dimensions + 1;
519 
520  template <typename CoordT, typename CoordElemType = get_elem_type_t<CoordT>>
522  getAdjustedCoords(const CoordT &Coords) const {
523  CoordElemType LastCoord = 0;
524 
525  if (std::is_same<float, CoordElemType>::value) {
526  sycl::vec<int, Dimensions + 1> Size = MBaseAcc.getRangeInternal();
527  LastCoord =
528  MIdx / static_cast<float>(Size.template swizzle<Dimensions>());
529  } else {
530  LastCoord = MIdx;
531  }
532 
533  sycl::vec<CoordElemType, Dimensions> LeftoverCoords{LastCoord};
534  sycl::vec<CoordElemType, AdjustedDims> AdjustedCoords{Coords,
535  LeftoverCoords};
536  return AdjustedCoords;
537  }
538 
539 public:
543  BaseAcc,
544  size_t Idx)
545  : MBaseAcc(BaseAcc), MIdx(Idx) {}
546 
547  template <typename CoordT, int Dims = Dimensions,
548  typename = std::enable_if_t<
550  DataT read(const CoordT &Coords) const {
551  return MBaseAcc.read(getAdjustedCoords(Coords));
552  }
553 
554  template <typename CoordT, int Dims = Dimensions,
555  typename = std::enable_if_t<(Dims > 0) &&
557  DataT read(const CoordT &Coords, const sampler &Smpl) const {
558  return MBaseAcc.read(getAdjustedCoords(Coords), Smpl);
559  }
560 
561  template <typename CoordT, int Dims = Dimensions,
562  typename = std::enable_if_t<(Dims > 0) &&
564  void write(const CoordT &Coords, const DataT &Color) const {
565  return MBaseAcc.write(getAdjustedCoords(Coords), Color);
566  }
567 
568 #ifdef __SYCL_DEVICE_ONLY__
569  __SYCL2020_DEPRECATED("get_count() is deprecated, please use size() instead")
570  size_t get_count() const { return size(); }
571  size_t size() const noexcept { return get_range<Dimensions>().size(); }
572 
573  template <int Dims = Dimensions, typename = std::enable_if_t<Dims == 1>>
574  range<1> get_range() const {
575  int2 Count = MBaseAcc.getRangeInternal();
576  return range<1>(Count.x());
577  }
578  template <int Dims = Dimensions, typename = std::enable_if_t<Dims == 2>>
579  range<2> get_range() const {
580  int3 Count = MBaseAcc.getRangeInternal();
581  return range<2>(Count.x(), Count.y());
582  }
583 
584 #else
585 
586  __SYCL2020_DEPRECATED("get_count() is deprecated, please use size() instead")
587  size_t get_count() const { return size(); }
588  size_t size() const noexcept {
589  return MBaseAcc.MImageCount / MBaseAcc.getAccessRange()[Dimensions];
590  }
591 
592  template <int Dims = Dimensions,
593  typename = std::enable_if_t<(Dims == 1 || Dims == 2)>>
595  return detail::convertToArrayOfN<Dims, 1>(MBaseAcc.getAccessRange());
596  }
597 
598 #endif
599 
600 private:
601  size_t MIdx;
604  MBaseAcc;
605 };
606 
607 } // namespace detail
613 template <typename DataT, int Dimensions, access::mode AccessMode,
616  DataT, Dimensions, AccessMode, access::target::image, IsPlaceholder>
617  : public detail::image_accessor<DataT, Dimensions, AccessMode,
618  access::target::image, IsPlaceholder>,
619  public detail::OwnerLessBase<
620  accessor<DataT, Dimensions, AccessMode, access::target::image,
621  IsPlaceholder>> {
622 private:
623  accessor(const detail::AccessorImplPtr &Impl)
624  : detail::image_accessor<DataT, Dimensions, AccessMode,
626 
627 public:
628  template <typename AllocatorT>
630  handler &CommandGroupHandler)
631  : detail::image_accessor<DataT, Dimensions, AccessMode,
633  Image, CommandGroupHandler, Image.getElementSize()) {
634 #ifndef __SYCL_DEVICE_ONLY__
635  detail::associateWithHandler(CommandGroupHandler, this,
637 #endif
638  }
639 
640  template <typename AllocatorT>
642  handler &CommandGroupHandler, const property_list &propList)
643  : detail::image_accessor<DataT, Dimensions, AccessMode,
645  Image, CommandGroupHandler, Image.getElementSize()) {
646  (void)propList;
647 #ifndef __SYCL_DEVICE_ONLY__
648  detail::associateWithHandler(CommandGroupHandler, this,
650 #endif
651  }
652 #ifdef __SYCL_DEVICE_ONLY__
653 private:
654  using OCLImageTy =
655  typename detail::opencl_image_type<Dimensions, AccessMode,
656  access::target::image>::type;
657 
658  // Front End requires this method to be defined in the accessor class.
659  // It does not call the base class's init method.
660  void __init(OCLImageTy Image) { this->imageAccessorInit(Image); }
661 
662  // __init variant used by the device compiler for ESIMD kernels.
663  void __init_esimd(OCLImageTy Image) { this->imageAccessorInit(Image); }
664 
665 public:
666  // Default constructor for objects later initialized with __init member.
667  accessor() = default;
668 #endif
669 };
670 
678 template <typename DataT, int Dimensions, access::mode AccessMode,
681  access::target::host_image, IsPlaceholder>
682  : public detail::image_accessor<DataT, Dimensions, AccessMode,
683  access::target::host_image, IsPlaceholder>,
684  public detail::OwnerLessBase<
685  accessor<DataT, Dimensions, AccessMode, access::target::host_image,
686  IsPlaceholder>> {
687 public:
688  template <typename AllocatorT>
690  : detail::image_accessor<DataT, Dimensions, AccessMode,
691  access::target::host_image, IsPlaceholder>(
692  Image, Image.getElementSize()) {}
693 
694  template <typename AllocatorT>
696  const property_list &propList)
697  : detail::image_accessor<DataT, Dimensions, AccessMode,
698  access::target::host_image, IsPlaceholder>(
699  Image, Image.getElementSize()) {
700  (void)propList;
701  }
702 };
703 
712 template <typename DataT, int Dimensions, access::mode AccessMode,
715  DataT, Dimensions, AccessMode, access::target::image_array, IsPlaceholder>
716  : public detail::image_accessor<DataT, Dimensions + 1, AccessMode,
717  access::target::image, IsPlaceholder>,
718  public detail::OwnerLessBase<
719  accessor<DataT, Dimensions, AccessMode, access::target::image_array,
720  IsPlaceholder>> {
721 #ifdef __SYCL_DEVICE_ONLY__
722 private:
723  using OCLImageTy =
724  typename detail::opencl_image_type<Dimensions + 1, AccessMode,
725  access::target::image>::type;
726 
727  // Front End requires this method to be defined in the accessor class.
728  // It does not call the base class's init method.
729  void __init(OCLImageTy Image) { this->imageAccessorInit(Image); }
730 
731  // __init variant used by the device compiler for ESIMD kernels.
732  void __init_esimd(OCLImageTy Image) { this->imageAccessorInit(Image); }
733 
734 public:
735  // Default constructor for objects later initialized with __init member.
736  accessor() = default;
737 #endif
738 public:
739  template <typename AllocatorT>
741  handler &CommandGroupHandler)
742  : detail::image_accessor<DataT, Dimensions + 1, AccessMode,
743  access::target::image, IsPlaceholder>(
744  Image, CommandGroupHandler, Image.getElementSize()) {
745 #ifndef __SYCL_DEVICE_ONLY__
746  detail::associateWithHandler(CommandGroupHandler, this,
748 #endif
749  }
750 
751  template <typename AllocatorT>
753  handler &CommandGroupHandler, const property_list &propList)
754  : detail::image_accessor<DataT, Dimensions + 1, AccessMode,
755  access::target::image, IsPlaceholder>(
756  Image, CommandGroupHandler, Image.getElementSize()) {
757  (void)propList;
758 #ifndef __SYCL_DEVICE_ONLY__
759  detail::associateWithHandler(CommandGroupHandler, this,
761 #endif
762  }
763 
764  detail::__image_array_slice__<DataT, Dimensions, AccessMode, IsPlaceholder>
765  operator[](size_t Index) const {
766  return detail::__image_array_slice__<DataT, Dimensions, AccessMode,
767  IsPlaceholder>(*this, Index);
768  }
769 };
770 
771 // SYCL 2020 image accessors
772 
773 template <typename DataT, int Dimensions, access_mode AccessMode,
774  image_target AccessTarget = image_target::device>
776 #ifndef __SYCL_DEVICE_ONLY__
778 #endif // __SYCL_DEVICE_ONLY__
779  public detail::OwnerLessBase<
780  unsampled_image_accessor<DataT, Dimensions, AccessMode, AccessTarget>> {
781  static_assert(std::is_same_v<DataT, int4> || std::is_same_v<DataT, uint4> ||
782  std::is_same_v<DataT, float4> ||
783  std::is_same_v<DataT, half4>,
784  "The data type of an image accessor must be only int4, "
785  "uint4, float4 or half4 from SYCL namespace");
786  static_assert(AccessMode == access_mode::read ||
788  "Access mode must be either read or write.");
789 
790 #ifdef __SYCL_DEVICE_ONLY__
791  char MPadding[sizeof(detail::UnsampledImageAccessorBaseHost)];
792 #else
794 #endif // __SYCL_DEVICE_ONLY__
795 
796 public:
797  using value_type = typename std::conditional<AccessMode == access_mode::read,
798  const DataT, DataT>::type;
800  using const_reference = const DataT &;
801 
802  template <typename AllocatorT>
805  handler &CommandGroupHandlerRef, const property_list &PropList = {},
807 #ifdef __SYCL_DEVICE_ONLY__
808  {
809  (void)ImageRef;
810  (void)CommandGroupHandlerRef;
811  (void)PropList;
812  (void)CodeLoc;
813  }
814 #else
815  : host_base_class(detail::convertToArrayOfN<3, 1>(ImageRef.get_range()),
816  AccessMode, detail::getSyclObjImpl(ImageRef).get(),
817  Dimensions, ImageRef.getElementSize(),
818  {ImageRef.getRowPitch(), ImageRef.getSlicePitch(), 0},
819  ImageRef.getChannelType(), ImageRef.getChannelOrder(),
820  PropList) {
821  device Device = detail::getDeviceFromHandler(CommandGroupHandlerRef);
822  // Avoid aspect::image warning.
823  aspect ImageAspect = aspect::image;
824  if (AccessTarget == image_target::device && !Device.has(ImageAspect))
825  throw sycl::exception(
826  sycl::make_error_code(sycl::errc::feature_not_supported),
827  "Device associated with command group handler does not have "
828  "aspect::image.");
829 
831  detail::getSyclObjImpl(ImageRef).get(), this->impl.get(), AccessTarget,
832  AccessMode, (const void *)typeid(DataT).name(), sizeof(DataT), CodeLoc);
833  detail::associateWithHandler(CommandGroupHandlerRef, this, AccessTarget);
834  GDBMethodsAnchor();
835  }
836 #endif // __SYCL_DEVICE_ONLY__
837 
838  /* -- common interface members -- */
839 
841 
843 
845  operator=(const unsampled_image_accessor &Rhs) = default;
846 
848 
850 
851 #ifdef __SYCL_DEVICE_ONLY__
852  bool operator==(const unsampled_image_accessor &Rhs) const;
853 #else
854  bool operator==(const unsampled_image_accessor &Rhs) const {
855  return Rhs.impl == impl;
856  }
857 #endif // __SYCL_DEVICE_ONLY__
858 
859  bool operator!=(const unsampled_image_accessor &Rhs) const {
860  return !(Rhs == *this);
861  }
862 
863  /* -- property interface members -- */
864  template <typename Property> bool has_property() const noexcept {
865 #ifndef __SYCL_DEVICE_ONLY__
866  return getPropList().template has_property<Property>();
867 #else
868  return false;
869 #endif
870  }
871  template <typename Property> Property get_property() const {
872 #ifndef __SYCL_DEVICE_ONLY__
873  return getPropList().template get_property<Property>();
874 #else
875  return Property();
876 #endif
877  }
878 
879  size_t size() const noexcept {
880 #ifdef __SYCL_DEVICE_ONLY__
881  // Currently not reachable on device.
882  return 0;
883 #else
884  return host_base_class::getSize().size();
885 #endif // __SYCL_DEVICE_ONLY__
886  }
887 
888  /* Available only when: AccessMode == access_mode::read
889  if Dimensions == 1, CoordT = int
890  if Dimensions == 2, CoordT = int2
891  if Dimensions == 3, CoordT = int4 */
892  template <typename CoordT,
893  typename = std::enable_if_t<AccessMode == access_mode::read &&
895  Dimensions, CoordT>::value>>
896  DataT read(const CoordT &Coords) const noexcept {
897 #ifdef __SYCL_DEVICE_ONLY__
898  // Currently not reachable on device.
899  std::ignore = Coords;
900  return {0, 0, 0, 0};
901 #else
902  return host_base_class::read<DataT>(Coords);
903 #endif // __SYCL_DEVICE_ONLY__
904  }
905 
906  /* Available only when: AccessMode == access_mode::write
907  if Dimensions == 1, CoordT = int
908  if Dimensions == 2, CoordT = int2
909  if Dimensions == 3, CoordT = int4 */
910  template <typename CoordT,
911  typename = std::enable_if_t<AccessMode == access_mode::write &&
913  Dimensions, CoordT>::value>>
914  void write(const CoordT &Coords, const DataT &Color) const {
915 #ifdef __SYCL_DEVICE_ONLY__
916  // Currently not reachable on device.
917  std::ignore = Coords;
918  std::ignore = Color;
919 #else
920  host_base_class::write<DataT>(Coords, Color);
921 #endif // __SYCL_DEVICE_ONLY__
922  }
923 
924 private:
926 #ifndef __SYCL_DEVICE_ONLY__
927  : host_base_class{Impl}
928 #endif // __SYCL_DEVICE_ONLY__
929  {
930  std::ignore = Impl;
931  }
932 
933  template <class Obj>
934  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
935 
936  template <class T>
937  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
938 };
939 
940 template <typename DataT, int Dimensions = 1,
942  (std::is_const_v<DataT> ? access_mode::read
946  public detail::OwnerLessBase<
948  static_assert(std::is_same_v<DataT, int4> || std::is_same_v<DataT, uint4> ||
949  std::is_same_v<DataT, float4> ||
950  std::is_same_v<DataT, half4>,
951  "The data type of an image accessor must be only int4, "
952  "uint4, float4 or half4 from SYCL namespace");
953 
955 
956 public:
957  using value_type = typename std::conditional<AccessMode == access_mode::read,
958  const DataT, DataT>::type;
960  using const_reference = const DataT &;
961 
962  template <typename AllocatorT>
965  const property_list &PropList = {},
967  : base_class(detail::convertToArrayOfN<3, 1>(ImageRef.get_range()),
968  AccessMode, detail::getSyclObjImpl(ImageRef).get(),
969  Dimensions, ImageRef.getElementSize(),
970  {ImageRef.getRowPitch(), ImageRef.getSlicePitch(), 0},
971  ImageRef.getChannelType(), ImageRef.getChannelOrder(),
972  PropList) {
973  addHostUnsampledImageAccessorAndWait(base_class::impl.get());
974 
976  detail::getSyclObjImpl(ImageRef).get(), this->impl.get(), std::nullopt,
977  AccessMode, (const void *)typeid(DataT).name(), sizeof(DataT), CodeLoc);
978  }
979 
980  /* -- common interface members -- */
981 
983  default;
984 
986 
989 
992 
994 
995  bool operator==(const host_unsampled_image_accessor &Rhs) const {
996  return Rhs.impl == impl;
997  }
998  bool operator!=(const host_unsampled_image_accessor &Rhs) const {
999  return !(Rhs == *this);
1000  }
1001 
1002  /* -- property interface members -- */
1003  template <typename Property> bool has_property() const noexcept {
1004 #ifndef __SYCL_DEVICE_ONLY__
1005  return getPropList().template has_property<Property>();
1006 #else
1007  return false;
1008 #endif
1009  }
1010  template <typename Property> Property get_property() const {
1011 #ifndef __SYCL_DEVICE_ONLY__
1012  return getPropList().template get_property<Property>();
1013 #else
1014  return Property();
1015 #endif
1016  }
1017 
1018  size_t size() const noexcept { return base_class::getSize().size(); }
1019 
1020  /* Available only when: (AccessMode == access_mode::read ||
1021  AccessMode == access_mode::read_write)
1022  if Dimensions == 1, CoordT = int
1023  if Dimensions == 2, CoordT = int2
1024  if Dimensions == 3, CoordT = int4 */
1025  template <
1026  typename CoordT,
1027  typename = std::enable_if_t<
1031  DataT read(const CoordT &Coords) const noexcept
1032 #ifdef __SYCL_DEVICE_ONLY__
1033  ;
1034 #else
1035  {
1036  // Host implementation is only available in host code. Device is not allowed
1037  // to use host_unsampled_image_accessor.
1038  return base_class::read<DataT>(Coords);
1039  }
1040 #endif
1041 
1042  /* Available only when: (AccessMode == access_mode::write ||
1043  AccessMode == access_mode::read_write)
1044  if Dimensions == 1, CoordT = int
1045  if Dimensions == 2, CoordT = int2
1046  if Dimensions == 3, CoordT = int4 */
1047  template <
1048  typename CoordT,
1049  typename = std::enable_if_t<
1053  void write(const CoordT &Coords, const DataT &Color) const
1054 #ifdef __SYCL_DEVICE_ONLY__
1055  ;
1056 #else
1057  {
1058  // Host implementation is only available in host code. Device is not allowed
1059  // to use host_unsampled_image_accessor.
1060  base_class::write<DataT>(Coords, Color);
1061  }
1062 #endif
1063 
1064 private:
1067  : base_class{Impl} {}
1068 
1069  template <class Obj>
1070  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
1071 
1072  template <class T>
1073  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
1074 };
1075 
1076 template <typename DataT, int Dimensions,
1077  image_target AccessTarget = image_target::device>
1079 #ifndef __SYCL_DEVICE_ONLY__
1081 #endif // __SYCL_DEVICE_ONLY__
1082  public detail::OwnerLessBase<
1083  sampled_image_accessor<DataT, Dimensions, AccessTarget>> {
1084  static_assert(std::is_same_v<DataT, int4> || std::is_same_v<DataT, uint4> ||
1085  std::is_same_v<DataT, float4> ||
1086  std::is_same_v<DataT, half4>,
1087  "The data type of an image accessor must be only int4, "
1088  "uint4, float4 or half4 from SYCL namespace");
1089 
1090 #ifdef __SYCL_DEVICE_ONLY__
1091  char MPadding[sizeof(detail::SampledImageAccessorBaseHost)];
1092 #else
1094 #endif // __SYCL_DEVICE_ONLY__
1095 
1096 public:
1097  using value_type = const DataT;
1098  using reference = const DataT &;
1099  using const_reference = const DataT &;
1100 
1101  template <typename AllocatorT>
1104  handler &CommandGroupHandlerRef, const property_list &PropList = {},
1106 #ifdef __SYCL_DEVICE_ONLY__
1107  {
1108  (void)ImageRef;
1109  (void)CommandGroupHandlerRef;
1110  (void)PropList;
1111  (void)CodeLoc;
1112  }
1113 #else
1114  : host_base_class(detail::convertToArrayOfN<3, 1>(ImageRef.get_range()),
1115  detail::getSyclObjImpl(ImageRef).get(), Dimensions,
1116  ImageRef.getElementSize(),
1117  {ImageRef.getRowPitch(), ImageRef.getSlicePitch(), 0},
1118  ImageRef.getChannelType(), ImageRef.getChannelOrder(),
1119  ImageRef.getSampler(), PropList) {
1120  device Device = detail::getDeviceFromHandler(CommandGroupHandlerRef);
1121  // Avoid aspect::image warning.
1122  aspect ImageAspect = aspect::image;
1123  if (AccessTarget == image_target::device && !Device.has(ImageAspect))
1124  throw sycl::exception(
1125  sycl::make_error_code(sycl::errc::feature_not_supported),
1126  "Device associated with command group handler does not have "
1127  "aspect::image.");
1128 
1130  detail::getSyclObjImpl(ImageRef).get(), this->impl.get(), AccessTarget,
1131  (const void *)typeid(DataT).name(), sizeof(DataT), CodeLoc);
1132  detail::associateWithHandler(CommandGroupHandlerRef, this, AccessTarget);
1133  GDBMethodsAnchor();
1134  }
1135 #endif // __SYCL_DEVICE_ONLY__
1136 
1137  /* -- common interface members -- */
1138 
1140 
1142 
1144  operator=(const sampled_image_accessor &Rhs) = default;
1145 
1147 
1149 
1150 #ifdef __SYCL_DEVICE_ONLY__
1151  bool operator==(const sampled_image_accessor &Rhs) const;
1152 #else
1153  bool operator==(const sampled_image_accessor &Rhs) const {
1154  return Rhs.impl == impl;
1155  }
1156 #endif // __SYCL_DEVICE_ONLY__
1157 
1158  bool operator!=(const sampled_image_accessor &Rhs) const {
1159  return !(Rhs == *this);
1160  }
1161 
1162  /* -- property interface members -- */
1163  template <typename Property> bool has_property() const noexcept {
1164 #ifndef __SYCL_DEVICE_ONLY__
1165  return getPropList().template has_property<Property>();
1166 #else
1167  return false;
1168 #endif
1169  }
1170  template <typename Property> Property get_property() const {
1171 #ifndef __SYCL_DEVICE_ONLY__
1172  return getPropList().template get_property<Property>();
1173 #else
1174  return Property();
1175 #endif
1176  }
1177 
1178  size_t size() const noexcept {
1179 #ifdef __SYCL_DEVICE_ONLY__
1180  // Currently not reachable on device.
1181  return 0;
1182 #else
1183  return host_base_class::getSize().size();
1184 #endif // __SYCL_DEVICE_ONLY__
1185  }
1186 
1187  /* if Dimensions == 1, CoordT = float
1188  if Dimensions == 2, CoordT = float2
1189  if Dimensions == 3, CoordT = float4 */
1190  template <typename CoordT,
1191  typename = std::enable_if_t<detail::IsValidSampledCoord2020DataT<
1192  Dimensions, CoordT>::value>>
1193  DataT read(const CoordT &Coords) const noexcept {
1194 #ifdef __SYCL_DEVICE_ONLY__
1195  // Currently not reachable on device.
1196  std::ignore = Coords;
1197  return {0, 0, 0, 0};
1198 #else
1199  return host_base_class::read<DataT>(Coords);
1200 #endif // __SYCL_DEVICE_ONLY__
1201  }
1202 
1203 private:
1205 #ifndef __SYCL_DEVICE_ONLY__
1206  : host_base_class{Impl}
1207 #endif // __SYCL_DEVICE_ONLY__
1208  {
1209  std::ignore = Impl;
1210  }
1211 
1212  template <class Obj>
1213  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
1214 
1215  template <class T>
1216  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
1217 };
1218 
1219 template <typename DataT, int Dimensions>
1222  public detail::OwnerLessBase<
1223  host_sampled_image_accessor<DataT, Dimensions>> {
1224  static_assert(std::is_same_v<DataT, int4> || std::is_same_v<DataT, uint4> ||
1225  std::is_same_v<DataT, float4> ||
1226  std::is_same_v<DataT, half4>,
1227  "The data type of an image accessor must be only int4, "
1228  "uint4, float4 or half4 from SYCL namespace");
1229 
1231 
1232 public:
1233  using value_type = const DataT;
1234  using reference = const DataT &;
1235  using const_reference = const DataT &;
1236 
1237  template <typename AllocatorT>
1240  const property_list &PropList = {},
1242  : base_class(detail::convertToArrayOfN<3, 1>(ImageRef.get_range()),
1243  detail::getSyclObjImpl(ImageRef).get(), Dimensions,
1244  ImageRef.getElementSize(),
1245  {ImageRef.getRowPitch(), ImageRef.getSlicePitch(), 0},
1246  ImageRef.getChannelType(), ImageRef.getChannelOrder(),
1247  ImageRef.getSampler(), PropList) {
1248  addHostSampledImageAccessorAndWait(base_class::impl.get());
1249 
1251  detail::getSyclObjImpl(ImageRef).get(), this->impl.get(), std::nullopt,
1252  (const void *)typeid(DataT).name(), sizeof(DataT), CodeLoc);
1253  }
1254 
1255  /* -- common interface members -- */
1256 
1258 
1260 
1263 
1266 
1268 
1269  bool operator==(const host_sampled_image_accessor &Rhs) const {
1270  return Rhs.impl == impl;
1271  }
1272  bool operator!=(const host_sampled_image_accessor &Rhs) const {
1273  return !(Rhs == *this);
1274  }
1275 
1276  /* -- property interface members -- */
1277  template <typename Property> bool has_property() const noexcept {
1278 #ifndef __SYCL_DEVICE_ONLY__
1279  return getPropList().template has_property<Property>();
1280 #else
1281  return false;
1282 #endif
1283  }
1284  template <typename Property> Property get_property() const {
1285 #ifndef __SYCL_DEVICE_ONLY__
1286  return getPropList().template get_property<Property>();
1287 #else
1288  return Property();
1289 #endif
1290  }
1291 
1292  size_t size() const noexcept { return base_class::getSize().size(); }
1293 
1294  /* if Dimensions == 1, CoordT = float
1295  if Dimensions == 2, CoordT = float2
1296  if Dimensions == 3, CoordT = float4 */
1297  template <typename CoordT,
1298  typename = std::enable_if_t<detail::IsValidSampledCoord2020DataT<
1299  Dimensions, CoordT>::value>>
1300  DataT read(const CoordT &Coords) const
1301 #ifdef __SYCL_DEVICE_ONLY__
1302  ;
1303 #else
1304  {
1305  // Host implementation is only available in host code. Device is not allowed
1306  // to use host_sampled_image_accessor.
1307  return base_class::read<DataT>(Coords);
1308  }
1309 #endif
1310 
1311 private:
1313  : base_class{Impl} {}
1314 
1315  template <class Obj>
1316  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
1317 
1318  template <class T>
1319  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
1320 };
1321 
1322 } // namespace _V1
1323 } // namespace sycl
1324 
1325 namespace std {
1326 template <typename DataT, int Dimensions, sycl::access_mode AccessMode,
1327  sycl::image_target AccessTarget>
1328 struct hash<sycl::unsampled_image_accessor<DataT, Dimensions, AccessMode,
1329  AccessTarget>> {
1331  AccessTarget>;
1332 
1333  size_t operator()(const AccType &A) const {
1334 #ifdef __SYCL_DEVICE_ONLY__
1335  // Hash is not supported on DEVICE. Just return 0 here.
1336  (void)A;
1337  return 0;
1338 #else
1339  auto AccImplPtr = sycl::detail::getSyclObjImpl(A);
1340  return hash<decltype(AccImplPtr)>()(AccImplPtr);
1341 #endif
1342  }
1343 };
1344 
1345 template <typename DataT, int Dimensions, sycl::access_mode AccessMode>
1346 struct hash<
1347  sycl::host_unsampled_image_accessor<DataT, Dimensions, AccessMode>> {
1348  using AccType =
1350 
1351  size_t operator()(const AccType &A) const {
1352  auto AccImplPtr = sycl::detail::getSyclObjImpl(A);
1353  return hash<decltype(AccImplPtr)>()(AccImplPtr);
1354  }
1355 };
1356 
1357 template <typename DataT, int Dimensions, sycl::image_target AccessTarget>
1358 struct hash<sycl::sampled_image_accessor<DataT, Dimensions, AccessTarget>> {
1360 
1361  size_t operator()(const AccType &A) const {
1362 #ifdef __SYCL_DEVICE_ONLY__
1363  // Hash is not supported on DEVICE. Just return 0 here.
1364  (void)A;
1365  return 0;
1366 #else
1367  auto AccImplPtr = sycl::detail::getSyclObjImpl(A);
1368  return hash<decltype(AccImplPtr)>()(AccImplPtr);
1369 #endif
1370  }
1371 };
1372 
1373 template <typename DataT, int Dimensions>
1374 struct hash<sycl::host_sampled_image_accessor<DataT, Dimensions>> {
1376 
1377  size_t operator()(const AccType &A) const {
1378  auto AccImplPtr = sycl::detail::getSyclObjImpl(A);
1379  return hash<decltype(AccImplPtr)>()(AccImplPtr);
1380  }
1381 };
1382 
1383 } // namespace std
The file contains implementations of accessor class.
accessor(sycl::image< Dimensions, AllocatorT > &Image, const property_list &propList)
unsigned int getElemSize() const
Definition: accessor.cpp:87
AccessorBaseHost(const AccessorImplPtr &Impl)
Definition: accessor.hpp:518
DataT read(const CoordT &Coords) const
SampledImageAccessorBaseHost(const SampledImageAccessorImplPtr &Impl)
UnsampledImageAccessorBaseHost(const UnsampledImageAccessorImplPtr &Impl)
void write(const CoordT &Coords, const DataT &Color) const
DataT read(const CoordT &Coords) const noexcept
__SYCL2020_DEPRECATED("get_count() is deprecated, please use size() instead") size_t get_count() const
DataT read(const CoordT &Coords) const
void write(const CoordT &Coords, const DataT &Color) const
DataT read(const CoordT &Coords, const sampler &Smpl) const
__image_array_slice__(accessor< DataT, Dimensions, AccessMode, access::target::image_array, IsPlaceholder, ext::oneapi::accessor_property_list<>> BaseAcc, size_t Idx)
image_accessor(image< Dims, AllocatorT > &ImageRef, int ImageElementSize)
bool operator!=(const image_accessor &Rhs) const
__SYCL2020_DEPRECATED("get_count() is deprecated, please use size() instead") size_t get_count() const
friend class sycl::ext::intel::esimd::detail::AccessorPrivateProxy
void write(const CoordT &Coords, const DataT &Color) const
image_accessor(const AccessorImplPtr &Impl)
bool operator==(const image_accessor &Rhs) const
DataT read(const CoordT &Coords) const
DataT read(const CoordT &Coords, const sampler &Smpl) const
image_accessor(image< Dims, AllocatorT > &ImageRef, handler &CommandGroupHandlerRef, int ImageElementSize)
range< Dimensions > get_range() const
Definition: image.hpp:323
size_t size() const noexcept
Definition: image.hpp:335
size_t getElementSize() const
Definition: image.cpp:187
image_sampler getSampler() const noexcept
Definition: image.cpp:193
size_t getRowPitch() const
Definition: image.cpp:189
image_channel_order getChannelOrder() const
Definition: image.cpp:197
size_t getSlicePitch() const
Definition: image.cpp:191
image_channel_type getChannelType() const
Definition: image.cpp:201
The SYCL device class encapsulates a single SYCL device on which kernels may be executed.
Definition: device.hpp:64
bool has(aspect Aspect) const __SYCL_WARN_IMAGE_ASPECT(Aspect)
Indicates if the SYCL device has the given feature.
Definition: device.cpp:219
Command group handler class.
Definition: handler.hpp:462
bool operator!=(const host_sampled_image_accessor &Rhs) const
host_sampled_image_accessor(const host_sampled_image_accessor &Rhs)=default
DataT read(const CoordT &Coords) const
host_sampled_image_accessor & operator=(const host_sampled_image_accessor &Rhs)=default
host_sampled_image_accessor(host_sampled_image_accessor &&Rhs)=default
bool operator==(const host_sampled_image_accessor &Rhs) const
host_sampled_image_accessor(sampled_image< Dimensions, AllocatorT > &ImageRef, const property_list &PropList={}, const detail::code_location CodeLoc=detail::code_location::current())
host_sampled_image_accessor & operator=(host_sampled_image_accessor &&Rhs)=default
host_unsampled_image_accessor(unsampled_image< Dimensions, AllocatorT > &ImageRef, const property_list &PropList={}, const detail::code_location CodeLoc=detail::code_location::current())
void write(const CoordT &Coords, const DataT &Color) const
host_unsampled_image_accessor(const host_unsampled_image_accessor &Rhs)=default
host_unsampled_image_accessor(host_unsampled_image_accessor &&Rhs)=default
DataT read(const CoordT &Coords) const noexcept
bool operator==(const host_unsampled_image_accessor &Rhs) const
typename std::conditional< AccessMode==access_mode::read, const DataT, DataT >::type value_type
host_unsampled_image_accessor & operator=(host_unsampled_image_accessor &&Rhs)=default
host_unsampled_image_accessor & operator=(const host_unsampled_image_accessor &Rhs)=default
bool operator!=(const host_unsampled_image_accessor &Rhs) const
Defines a shared image data.
Definition: image.hpp:443
range< Dimensions > get_range() const
Definition: image.hpp:629
Objects of the property_list class are containers for the SYCL properties.
Defines the iteration domain of either a single work-group in a parallel dispatch,...
Definition: range.hpp:26
sampled_image_accessor(const sampled_image_accessor &Rhs)=default
bool operator==(const sampled_image_accessor &Rhs) const
sampled_image_accessor(sampled_image_accessor &&Rhs)=default
sampled_image_accessor & operator=(sampled_image_accessor &&Rhs)=default
sampled_image_accessor(sampled_image< Dimensions, AllocatorT > &ImageRef, handler &CommandGroupHandlerRef, const property_list &PropList={}, const detail::code_location CodeLoc=detail::code_location::current())
sampled_image_accessor & operator=(const sampled_image_accessor &Rhs)=default
DataT read(const CoordT &Coords) const noexcept
bool operator!=(const sampled_image_accessor &Rhs) const
DataT read(const CoordT &Coords) const noexcept
bool operator!=(const unsampled_image_accessor &Rhs) const
void write(const CoordT &Coords, const DataT &Color) const
unsampled_image_accessor & operator=(unsampled_image_accessor &&Rhs)=default
bool operator==(const unsampled_image_accessor &Rhs) const
unsampled_image_accessor & operator=(const unsampled_image_accessor &Rhs)=default
typename std::conditional< AccessMode==access_mode::read, const DataT, DataT >::type value_type
unsampled_image_accessor(const unsampled_image_accessor &Rhs)=default
unsampled_image_accessor(unsampled_image_accessor &&Rhs)=default
unsampled_image_accessor(unsampled_image< Dimensions, AllocatorT > &ImageRef, handler &CommandGroupHandlerRef, const property_list &PropList={}, const detail::code_location CodeLoc=detail::code_location::current())
class sycl::vec ///////////////////////// Provides a cross-patform vector class template that works e...
#define __SYCL_SPECIAL_CLASS
Definition: defines.hpp:29
#define __SYCL_EBO
Definition: common.hpp:175
class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor< DataT
Image accessors.
class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor accessor(buffer< DataT, Dimensions, AllocatorT >) -> accessor< DataT, Dimensions, access::mode::read_write, target::device, access::placeholder::true_t >
Buffer accessor.
void unsampledImageConstructorNotification(void *ImageObj, void *AccessorObj, const std::optional< image_target > &Target, access::mode Mode, const void *Type, uint32_t ElemSize, const code_location &CodeLoc)
void addHostUnsampledImageAccessorAndWait(UnsampledImageAccessorImplHost *Req)
void addHostAccessorAndWait(AccessorImplHost *Req)
device getDeviceFromHandler(handler &CommandGroupHandlerRef)
Definition: accessor.cpp:17
std::shared_ptr< UnsampledImageAccessorImplHost > UnsampledImageAccessorImplPtr
void imageWriteHostImpl(const CoordT &Coords, const WriteDataT &Color, id< 3 > ImgPitch, uint8_t ElementSize, image_channel_type ImgChannelType, image_channel_order ImgChannelOrder, void *BasePtr)
void addHostSampledImageAccessorAndWait(SampledImageAccessorImplHost *Req)
boost::mp11::mp_set_contains< TypeList, std::remove_cv_t< T > > is_contained
Definition: type_list.hpp:32
std::shared_ptr< SampledImageAccessorImplHost > SampledImageAccessorImplPtr
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: impl_utils.hpp:30
boost::mp11::mp_list< T... > type_list
Definition: type_list.hpp:22
T createSyclObjFromImpl(decltype(T::impl) ImplObj)
Definition: impl_utils.hpp:48
void sampledImageConstructorNotification(void *ImageObj, void *AccessorObj, const std::optional< image_target > &Target, const void *Type, uint32_t ElemSize, const code_location &CodeLoc)
void associateWithHandler(handler &, AccessorBaseHost *, access::target)
std::shared_ptr< AccessorImplHost > AccessorImplPtr
Definition: accessor.hpp:514
T & operator[](std::ptrdiff_t idx) const noexcept
access::mode access_mode
Definition: access.hpp:72
image_target
Definition: access.hpp:74
class __SYCL_EBO __SYCL_SPECIAL_CLASS Dimensions
constexpr mode_tag_t< access_mode::read_write > read_write
Definition: access.hpp:85
pointer get() const
Definition: multi_ptr.hpp:544
image_channel_order
Definition: image.hpp:56
class __SYCL_EBO __SYCL_SPECIAL_CLASS IsPlaceholder
class __SYCL_EBO __SYCL_SPECIAL_CLASS AccessMode
image_channel_type
Definition: image.hpp:74
std::error_code make_error_code(sycl::errc E) noexcept
Constructs an error code using e and sycl_category()
Definition: exception.cpp:93
Definition: access.hpp:18
_Abi const simd< _Tp, _Abi > & noexcept
Definition: simd.hpp:1324
static constexpr code_location current(const char *fileName=__CODELOC_FILE_NAME, const char *funcName=__CODELOC_FUNCTION, unsigned long lineNo=__CODELOC_LINE, unsigned long columnNo=__CODELOC_COLUMN) noexcept
Definition: common.hpp:68