DPC++ Runtime
Runtime libraries for oneAPI DPC++
accessor_impl.hpp
Go to the documentation of this file.
1 //==------------ accessor_impl.hpp - SYCL standard header file -------------==//
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/id.hpp>
15 #include <CL/sycl/range.hpp>
16 #include <CL/sycl/stl.hpp>
17 
19 namespace sycl {
20 
21 namespace ext {
22 namespace intel {
23 namespace esimd {
24 namespace detail {
25 // Forward declare a "back-door" access class to support ESIMD.
26 class AccessorPrivateProxy;
27 } // namespace detail
28 } // namespace esimd
29 } // namespace intel
30 } // namespace ext
31 
32 namespace detail {
33 
34 class Command;
35 
36 // The class describes a requirement to access a SYCL memory object such as
37 // sycl::buffer and sycl::image. For example, each accessor used in a kernel,
38 // except one with access target "local", adds such requirement for the command
39 // group.
40 
41 template <int Dims> class AccessorImplDevice {
42 public:
43  AccessorImplDevice() = default;
45  range<Dims> MemoryRange)
46  : Offset(Offset), AccessRange(AccessRange), MemRange(MemoryRange) {}
47 
51 
52  bool operator==(const AccessorImplDevice &Rhs) const {
53  return (Offset == Rhs.Offset && AccessRange == Rhs.AccessRange &&
54  MemRange == Rhs.MemRange);
55  }
56 };
57 
58 template <int Dims> class LocalAccessorBaseDevice {
59 public:
61  : AccessRange(Size),
62  MemRange(InitializedVal<Dims, range>::template get<0>()) {}
63  // TODO: Actually we need only one field here, but currently compiler requires
64  // all of them.
68 
69  bool operator==(const LocalAccessorBaseDevice &Rhs) const {
70  return (AccessRange == Rhs.AccessRange);
71  }
72 };
73 
74 class __SYCL_EXPORT AccessorImplHost {
75 public:
76  AccessorImplHost(id<3> Offset, range<3> AccessRange, range<3> MemoryRange,
77  access::mode AccessMode, detail::SYCLMemObjI *SYCLMemObject,
78  int Dims, int ElemSize, int OffsetInBytes = 0,
79  bool IsSubBuffer = false, bool IsESIMDAcc = false)
80  : MOffset(Offset), MAccessRange(AccessRange), MMemoryRange(MemoryRange),
81  MAccessMode(AccessMode), MSYCLMemObj(SYCLMemObject), MDims(Dims),
82  MElemSize(ElemSize), MOffsetInBytes(OffsetInBytes),
83  MIsSubBuffer(IsSubBuffer), MIsESIMDAcc(IsESIMDAcc) {}
84 
86 
88  : MOffset(Other.MOffset), MAccessRange(Other.MAccessRange),
89  MMemoryRange(Other.MMemoryRange), MAccessMode(Other.MAccessMode),
90  MSYCLMemObj(Other.MSYCLMemObj), MDims(Other.MDims),
91  MElemSize(Other.MElemSize), MOffsetInBytes(Other.MOffsetInBytes),
92  MIsSubBuffer(Other.MIsSubBuffer), MIsESIMDAcc(Other.MIsESIMDAcc) {}
93 
94  // The resize method provides a way to change the size of the
95  // allocated memory and corresponding properties for the accessor.
96  // These are normally fixed for the accessor, but this capability
97  // is needed to support the stream class.
98  // Stream implementation creates an accessor with initial size for
99  // work item. But the number of work items is not available during
100  // stream construction. The resize method allows to update the accessor
101  // as the information becomes available to the handler.
102 
103  void resize(size_t GlobalSize);
104 
106  // The size of accessing region.
108  // The size of memory object this requirement is created for.
111 
113 
114  unsigned int MDims;
115  unsigned int MElemSize;
116  unsigned int MOffsetInBytes;
118 
119  void *MData = nullptr;
120 
121  Command *MBlockedCmd = nullptr;
122 
123  bool PerWI = false;
124 
125  // Outdated, leaving to preserve ABI.
126  // TODO: Remove during next major release.
128 };
129 
130 using AccessorImplPtr = std::shared_ptr<AccessorImplHost>;
131 
133 public:
134  AccessorBaseHost(id<3> Offset, range<3> AccessRange, range<3> MemoryRange,
135  access::mode AccessMode, detail::SYCLMemObjI *SYCLMemObject,
136  int Dims, int ElemSize, int OffsetInBytes = 0,
137  bool IsSubBuffer = false) {
138  impl = std::shared_ptr<AccessorImplHost>(new AccessorImplHost(
139  Offset, AccessRange, MemoryRange, AccessMode, SYCLMemObject, Dims,
140  ElemSize, OffsetInBytes, IsSubBuffer));
141  }
142 
143 protected:
144  id<3> &getOffset() { return impl->MOffset; }
145  range<3> &getAccessRange() { return impl->MAccessRange; }
146  range<3> &getMemoryRange() { return impl->MMemoryRange; }
147  void *getPtr() { return impl->MData; }
148  unsigned int getElemSize() const { return impl->MElemSize; }
149 
150  const id<3> &getOffset() const { return impl->MOffset; }
151  const range<3> &getAccessRange() const { return impl->MAccessRange; }
152  const range<3> &getMemoryRange() const { return impl->MMemoryRange; }
153  void *getPtr() const { return const_cast<void *>(impl->MData); }
154 
155  template <class Obj>
156  friend decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject);
157 
158  template <typename, int, access::mode, access::target, access::placeholder,
159  typename>
160  friend class accessor;
161 
163 
164 private:
165  friend class sycl::ext::intel::esimd::detail::AccessorPrivateProxy;
166 };
167 
168 class __SYCL_EXPORT LocalAccessorImplHost {
169 public:
170  // Allocate ElemSize more data to have sufficient padding to enforce
171  // alignment.
172  LocalAccessorImplHost(sycl::range<3> Size, int Dims, int ElemSize)
173  : MSize(Size), MDims(Dims), MElemSize(ElemSize),
174  MMem(Size[0] * Size[1] * Size[2] * ElemSize + ElemSize) {}
175 
177  int MDims;
179  std::vector<char> MMem;
180 };
181 
182 using LocalAccessorImplPtr = std::shared_ptr<LocalAccessorImplHost>;
183 
185 public:
186  LocalAccessorBaseHost(sycl::range<3> Size, int Dims, int ElemSize) {
187  impl = std::shared_ptr<LocalAccessorImplHost>(
188  new LocalAccessorImplHost(Size, Dims, ElemSize));
189  }
190  sycl::range<3> &getSize() { return impl->MSize; }
191  const sycl::range<3> &getSize() const { return impl->MSize; }
192  void *getPtr() {
193  // Const cast this in order to call the const getPtr.
194  return const_cast<const LocalAccessorBaseHost *>(this)->getPtr();
195  }
196  void *getPtr() const {
197  char *ptr = impl->MMem.data();
198 
199  // Align the pointer to MElemSize.
200  size_t val = reinterpret_cast<size_t>(ptr);
201  if (val % impl->MElemSize != 0) {
202  ptr += impl->MElemSize - val % impl->MElemSize;
203  }
204 
205  return ptr;
206  }
207 
208  int getNumOfDims() { return impl->MDims; }
209  int getElementSize() { return impl->MElemSize; }
210 
211 protected:
212  template <class Obj>
213  friend decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject);
214 
215  std::shared_ptr<LocalAccessorImplHost> impl;
216 };
217 
219 
220 void __SYCL_EXPORT addHostAccessorAndWait(Requirement *Req);
221 
222 #if __cplusplus >= 201703L
223 
224 template <typename MayBeTag1, typename MayBeTag2>
225 constexpr access::mode deduceAccessMode() {
226  // property_list = {} is not properly detected by deduction guide,
227  // when parameter is passed without curly braces: access(buffer, no_init)
228  // thus simplest approach is to check 2 last arguments for being a tag
229  if constexpr (std::is_same<MayBeTag1,
231  std::is_same<MayBeTag2,
233  return access::mode::read;
234  }
235 
236  if constexpr (std::is_same<MayBeTag1,
238  std::is_same<MayBeTag2,
240  return access::mode::write;
241  }
242 
243  if constexpr (
244  std::is_same<MayBeTag1,
245  mode_target_tag_t<access::mode::read,
246  access::target::constant_buffer>>::value ||
247  std::is_same<MayBeTag2,
248  mode_target_tag_t<access::mode::read,
249  access::target::constant_buffer>>::value) {
250  return access::mode::read;
251  }
252 
253  return access::mode::read_write;
254 }
255 
256 template <typename MayBeTag1, typename MayBeTag2>
257 constexpr access::target deduceAccessTarget(access::target defaultTarget) {
258  if constexpr (
259  std::is_same<MayBeTag1,
260  mode_target_tag_t<access::mode::read,
261  access::target::constant_buffer>>::value ||
262  std::is_same<MayBeTag2,
263  mode_target_tag_t<access::mode::read,
264  access::target::constant_buffer>>::value) {
265  return access::target::constant_buffer;
266  }
267 
268  return defaultTarget;
269 }
270 
271 #endif
272 
273 } // namespace detail
274 } // namespace sycl
275 } // __SYCL_INLINE_NAMESPACE(cl)
cl::sycl::detail::LocalAccessorBaseHost::getSize
sycl::range< 3 > & getSize()
Definition: accessor_impl.hpp:190
cl::sycl::detail::AccessorImplDevice::Offset
id< Dims > Offset
Definition: accessor_impl.hpp:48
cl::sycl::detail::AccessorImplHost::MDims
unsigned int MDims
Definition: accessor_impl.hpp:114
cl::sycl::detail::AccessorBaseHost::getPtr
void * getPtr() const
Definition: accessor_impl.hpp:153
cl::sycl::detail::LocalAccessorImplHost::MDims
int MDims
Definition: accessor_impl.hpp:177
cl::sycl::detail::AccessorBaseHost::AccessorBaseHost
AccessorBaseHost(id< 3 > Offset, range< 3 > AccessRange, range< 3 > MemoryRange, access::mode AccessMode, detail::SYCLMemObjI *SYCLMemObject, int Dims, int ElemSize, int OffsetInBytes=0, bool IsSubBuffer=false)
Definition: accessor_impl.hpp:134
sycl_mem_obj_i.hpp
cl::sycl::detail::LocalAccessorBaseHost::getPtr
void * getPtr() const
Definition: accessor_impl.hpp:196
cl::sycl::detail::LocalAccessorImplHost
Definition: accessor_impl.hpp:168
cl::sycl::detail::LocalAccessorBaseHost::getNumOfDims
int getNumOfDims()
Definition: accessor_impl.hpp:208
cl::sycl::detail::AccessorImplDevice
Definition: accessor_impl.hpp:41
cl::sycl::access::placeholder
placeholder
Definition: access.hpp:43
cl::sycl::detail::LocalAccessorImplPtr
std::shared_ptr< LocalAccessorImplHost > LocalAccessorImplPtr
Definition: accessor_impl.hpp:182
cl::sycl::detail::AccessorImplHost::MIsESIMDAcc
bool MIsESIMDAcc
Definition: accessor_impl.hpp:127
stl.hpp
cl::sycl::id< Dims >
cl::sycl::detail::LocalAccessorBaseDevice::AccessRange
range< Dims > AccessRange
Definition: accessor_impl.hpp:65
cl::sycl::detail::InitializedVal
Definition: common.hpp:227
sycl
Definition: invoke_simd.hpp:68
cl::sycl::detail::LocalAccessorBaseHost::getElementSize
int getElementSize()
Definition: accessor_impl.hpp:209
access.hpp
cl::sycl::detail::LocalAccessorBaseDevice
Definition: accessor_impl.hpp:58
cl::sycl::detail::AccessorImplDevice::AccessRange
range< Dims > AccessRange
Definition: accessor_impl.hpp:49
cl::sycl::detail::AccessorBaseHost::impl
AccessorImplPtr impl
Definition: accessor_impl.hpp:162
cl::sycl::detail::write
void write(GlobalBufAccessorT &GlobalFlushBuf, size_t FlushBufferSize, unsigned WIOffset, const char *Str, unsigned Len, unsigned Padding=0)
Definition: stream.hpp:110
cl::sycl::detail::LocalAccessorBaseHost
Definition: accessor_impl.hpp:184
id.hpp
cl::sycl::range< Dims >
cl::sycl::detail::LocalAccessorImplHost::MSize
sycl::range< 3 > MSize
Definition: accessor_impl.hpp:176
cl::sycl::detail::get
Definition: tuple.hpp:59
cl::sycl::detail::LocalAccessorImplHost::MMem
std::vector< char > MMem
Definition: accessor_impl.hpp:179
cl::sycl::detail::AccessorImplHost::MElemSize
unsigned int MElemSize
Definition: accessor_impl.hpp:115
export.hpp
cl::sycl::detail::LocalAccessorBaseDevice::MemRange
range< Dims > MemRange
Definition: accessor_impl.hpp:66
cl::sycl::detail::AccessorImplHost
Definition: accessor_impl.hpp:74
cl::sycl::detail::LocalAccessorBaseHost::getPtr
void * getPtr()
Definition: accessor_impl.hpp:192
cl::sycl::detail::AccessorImplHost::MOffsetInBytes
unsigned int MOffsetInBytes
Definition: accessor_impl.hpp:116
cl::sycl::detail::AccessorImplHost::MOffset
id< 3 > MOffset
Definition: accessor_impl.hpp:105
cl::sycl::detail::AccessorImplHost::MIsSubBuffer
bool MIsSubBuffer
Definition: accessor_impl.hpp:117
range.hpp
cl::sycl::detail::AccessorImplPtr
std::shared_ptr< AccessorImplHost > AccessorImplPtr
Definition: accessor_impl.hpp:130
cl::sycl::detail::AccessorBaseHost::getElemSize
unsigned int getElemSize() const
Definition: accessor_impl.hpp:148
cl::sycl::accessor
Buffer accessor.
Definition: accessor.hpp:223
cl::sycl::access::target
target
Definition: access.hpp:17
cl::sycl::detail::LocalAccessorBaseDevice::operator==
bool operator==(const LocalAccessorBaseDevice &Rhs) const
Definition: accessor_impl.hpp:69
cl::sycl::detail::Command
The Command class represents some action that needs to be performed on one or more memory objects.
Definition: commands.hpp:95
cl
We provide new interfaces for matrix muliply in this patch:
Definition: access.hpp:13
cl::sycl::detail::AccessorImplHost::MSYCLMemObj
detail::SYCLMemObjI * MSYCLMemObj
Definition: accessor_impl.hpp:112
cl::sycl::detail::AccessorImplHost::AccessorImplHost
AccessorImplHost(id< 3 > Offset, range< 3 > AccessRange, range< 3 > MemoryRange, access::mode AccessMode, detail::SYCLMemObjI *SYCLMemObject, int Dims, int ElemSize, int OffsetInBytes=0, bool IsSubBuffer=false, bool IsESIMDAcc=false)
Definition: accessor_impl.hpp:76
cl::sycl::detail::addHostAccessorAndWait
void addHostAccessorAndWait(Requirement *Req)
Definition: accessor_impl.cpp:36
cl::sycl::detail::AccessorImplDevice::AccessorImplDevice
AccessorImplDevice(id< Dims > Offset, range< Dims > AccessRange, range< Dims > MemoryRange)
Definition: accessor_impl.hpp:44
cl::sycl::detail::LocalAccessorImplHost::LocalAccessorImplHost
LocalAccessorImplHost(sycl::range< 3 > Size, int Dims, int ElemSize)
Definition: accessor_impl.hpp:172
cl::sycl::detail::AccessorBaseHost::getPtr
void * getPtr()
Definition: accessor_impl.hpp:147
cl::sycl::detail::AccessorImplHost::MAccessMode
access::mode MAccessMode
Definition: accessor_impl.hpp:110
cl::sycl::detail::getSyclObjImpl
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:204
cl::sycl::detail::LocalAccessorBaseDevice::Offset
id< Dims > Offset
Definition: accessor_impl.hpp:67
cl::sycl::detail::AccessorImplDevice::operator==
bool operator==(const AccessorImplDevice &Rhs) const
Definition: accessor_impl.hpp:52
cl::sycl::detail::AccessorImplHost::MMemoryRange
range< 3 > MMemoryRange
Definition: accessor_impl.hpp:109
cl::sycl::detail::AccessorImplHost::AccessorImplHost
AccessorImplHost(const AccessorImplHost &Other)
Definition: accessor_impl.hpp:87
cl::sycl::detail::AccessorBaseHost
Definition: accessor_impl.hpp:132
cl::sycl::detail::LocalAccessorBaseDevice::LocalAccessorBaseDevice
LocalAccessorBaseDevice(sycl::range< Dims > Size)
Definition: accessor_impl.hpp:60
cl::sycl::detail::LocalAccessorBaseHost::getSize
const sycl::range< 3 > & getSize() const
Definition: accessor_impl.hpp:191
cl::sycl::access::mode
mode
Definition: access.hpp:28
cl::sycl::detail::AccessorImplHost::MAccessRange
range< 3 > MAccessRange
Definition: accessor_impl.hpp:107
cl::sycl::mode_tag_t
Definition: access.hpp:67
cl::sycl::detail::SYCLMemObjI
Definition: sycl_mem_obj_i.hpp:28
cl::sycl::detail::AccessorImplDevice::MemRange
range< Dims > MemRange
Definition: accessor_impl.hpp:50
cl::sycl::detail::LocalAccessorBaseHost::impl
std::shared_ptr< LocalAccessorImplHost > impl
Definition: accessor_impl.hpp:215
cl::sycl::detail::LocalAccessorBaseHost::LocalAccessorBaseHost
LocalAccessorBaseHost(sycl::range< 3 > Size, int Dims, int ElemSize)
Definition: accessor_impl.hpp:186
__SYCL_INLINE_NAMESPACE
#define __SYCL_INLINE_NAMESPACE(X)
Definition: defines_elementary.hpp:12
cl::sycl::detail::LocalAccessorImplHost::MElemSize
int MElemSize
Definition: accessor_impl.hpp:178