DPC++ Runtime
Runtime libraries for oneAPI DPC++
marray.hpp
Go to the documentation of this file.
1 //==----------------- marray.hpp --- Implements marray classes -------------==//
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 
11 #include <sycl/aliases.hpp>
12 #include <sycl/detail/common.hpp>
13 #include <sycl/detail/is_device_copyable.hpp>
14 #include <sycl/half_type.hpp>
15 
16 #include <array>
17 #include <cstddef>
18 #include <cstdint>
19 #include <type_traits>
20 #include <utility>
21 
22 namespace sycl {
23 inline namespace _V1 {
24 
25 template <typename DataT, std::size_t N> class marray;
26 
27 namespace detail {
28 
29 // Helper trait for counting the aggregate number of arguments in a type list,
30 // expanding marrays.
31 template <typename... Ts> struct GetMArrayArgsSize;
32 template <> struct GetMArrayArgsSize<> {
33  static constexpr std::size_t value = 0;
34 };
35 template <typename T, std::size_t N, typename... Ts>
36 struct GetMArrayArgsSize<marray<T, N>, Ts...> {
37  static constexpr std::size_t value = N + GetMArrayArgsSize<Ts...>::value;
38 };
39 template <typename T, typename... Ts> struct GetMArrayArgsSize<T, Ts...> {
40  static constexpr std::size_t value = 1 + GetMArrayArgsSize<Ts...>::value;
41 };
42 
43 } // namespace detail
44 
49 template <typename Type, std::size_t NumElements> class marray {
50  using DataT = Type;
51 
52 public:
53  using value_type = Type;
54  using reference = Type &;
55  using const_reference = const Type &;
56  using iterator = Type *;
57  using const_iterator = const Type *;
58 
59 private:
60  value_type MData[NumElements];
61 
62  // Trait for checking if an argument type is either convertible to the data
63  // type or an array of types convertible to the data type.
64  template <typename T>
65  struct IsSuitableArgType : std::is_convertible<T, DataT> {};
66  template <typename T, size_t N>
67  struct IsSuitableArgType<marray<T, N>> : std::is_convertible<T, DataT> {};
68 
69  // Trait for computing the conjunction of of IsSuitableArgType. The empty type
70  // list will trivially evaluate to true.
71  template <typename... ArgTN>
72  struct AllSuitableArgTypes : std::conjunction<IsSuitableArgType<ArgTN>...> {};
73 
74  // Utility trait for creating an std::array from an marray argument.
75  template <typename DataT, typename T, std::size_t... Is>
76  static constexpr std::array<DataT, sizeof...(Is)>
77  MArrayToArray(const marray<T, sizeof...(Is)> &A, std::index_sequence<Is...>) {
78  return {static_cast<DataT>(A.MData[Is])...};
79  }
80  template <typename DataT, typename T, std::size_t N>
81  static constexpr std::array<DataT, N>
82  FlattenMArrayArgHelper(const marray<T, N> &A) {
83  return MArrayToArray<DataT>(A, std::make_index_sequence<N>());
84  }
85  template <typename DataT, typename T>
86  static constexpr auto FlattenMArrayArgHelper(const T &A) {
87  return std::array<DataT, 1>{static_cast<DataT>(A)};
88  }
89  template <typename DataT, typename T> struct FlattenMArrayArg {
90  constexpr auto operator()(const T &A) const {
91  return FlattenMArrayArgHelper<DataT>(A);
92  }
93  };
94 
95  // Alias for shortening the marray arguments to array converter.
96  template <typename DataT, typename... ArgTN>
97  using MArrayArgArrayCreator =
98  detail::ArrayCreator<DataT, FlattenMArrayArg, ArgTN...>;
99 
100  // FIXME: Other marray specializations needs to be a friend to access MData.
101  // If the subscript operator is made constexpr this can be removed.
102  template <typename Type_, std::size_t NumElements_> friend class marray;
103 
104  constexpr void initialize_data(const Type &Arg) {
105  for (size_t i = 0; i < NumElements; ++i) {
106  MData[i] = Arg;
107  }
108  }
109 
110  template <size_t... Is>
111  constexpr marray(const std::array<DataT, NumElements> &Arr,
112  std::index_sequence<Is...>)
113  : MData{Arr[Is]...} {}
114 
115 public:
116  constexpr marray() : MData{} {}
117 
118  explicit constexpr marray(const Type &Arg) : MData{Arg} {
119  initialize_data(Arg);
120  }
121 
122  template <typename... ArgTN,
123  typename = std::enable_if_t<
124  AllSuitableArgTypes<ArgTN...>::value &&
125  detail::GetMArrayArgsSize<ArgTN...>::value == NumElements>>
126  constexpr marray(const ArgTN &...Args)
127  : marray{MArrayArgArrayCreator<DataT, ArgTN...>::Create(Args...),
128  std::make_index_sequence<NumElements>()} {}
129 
130  constexpr marray(const marray<Type, NumElements> &Rhs) = default;
131 
132  constexpr marray(marray<Type, NumElements> &&Rhs) = default;
133 
134  // Available only when: NumElements == 1
135  template <std::size_t Size = NumElements,
136  typename = std::enable_if_t<Size == 1>>
137  operator Type() const {
138  return MData[0];
139  }
140 
141  static constexpr std::size_t size() noexcept { return NumElements; }
142 
143  // subscript operator
144  reference operator[](std::size_t index) { return MData[index]; }
145 
146  const_reference operator[](std::size_t index) const { return MData[index]; }
147 
148  marray &operator=(const marray<Type, NumElements> &Rhs) = default;
149 
150  // broadcasting operator
151  marray &operator=(const Type &Rhs) {
152  for (std::size_t I = 0; I < NumElements; ++I) {
153  MData[I] = Rhs;
154  }
155  return *this;
156  }
157 
158  // iterator functions
159  iterator begin() { return MData; }
160 
161  const_iterator begin() const { return MData; }
162 
163  iterator end() { return MData + NumElements; }
164 
165  const_iterator end() const { return MData + NumElements; }
166 
167 #ifdef __SYCL_BINOP
168 #error "Undefine __SYCL_BINOP macro"
169 #endif
170 
171 #ifdef __SYCL_BINOP_INTEGRAL
172 #error "Undefine __SYCL_BINOP_INTEGRAL macro"
173 #endif
174 
175 #define __SYCL_BINOP(BINOP, OPASSIGN) \
176  friend marray operator BINOP(const marray &Lhs, const marray &Rhs) { \
177  marray Ret; \
178  for (size_t I = 0; I < NumElements; ++I) { \
179  Ret[I] = Lhs[I] BINOP Rhs[I]; \
180  } \
181  return Ret; \
182  } \
183  template <typename T> \
184  friend typename std::enable_if_t< \
185  std::is_convertible_v<DataT, T> && \
186  (std::is_fundamental_v<T> || \
187  std::is_same_v<typename std::remove_const<T>::type, half>), \
188  marray> \
189  operator BINOP(const marray &Lhs, const T &Rhs) { \
190  return Lhs BINOP marray(static_cast<DataT>(Rhs)); \
191  } \
192  template <typename T> \
193  friend typename std::enable_if_t< \
194  std::is_convertible_v<DataT, T> && \
195  (std::is_fundamental_v<T> || \
196  std::is_same_v<typename std::remove_const<T>::type, half>), \
197  marray> \
198  operator BINOP(const T &Lhs, const marray &Rhs) { \
199  return marray(static_cast<DataT>(Lhs)) BINOP Rhs; \
200  } \
201  friend marray &operator OPASSIGN(marray &Lhs, const marray &Rhs) { \
202  Lhs = Lhs BINOP Rhs; \
203  return Lhs; \
204  } \
205  template <std::size_t Num = NumElements> \
206  friend typename std::enable_if_t<Num != 1, marray &> operator OPASSIGN( \
207  marray &Lhs, const DataT &Rhs) { \
208  Lhs = Lhs BINOP marray(Rhs); \
209  return Lhs; \
210  }
211 
212 #define __SYCL_BINOP_INTEGRAL(BINOP, OPASSIGN) \
213  template <typename T = DataT, \
214  typename = std::enable_if_t<std::is_integral_v<T>, marray>> \
215  friend marray operator BINOP(const marray &Lhs, const marray &Rhs) { \
216  marray Ret; \
217  for (size_t I = 0; I < NumElements; ++I) { \
218  Ret[I] = Lhs[I] BINOP Rhs[I]; \
219  } \
220  return Ret; \
221  } \
222  template <typename T, typename BaseT = DataT> \
223  friend typename std::enable_if_t<std::is_convertible_v<T, DataT> && \
224  std::is_integral_v<T> && \
225  std::is_integral_v<BaseT>, \
226  marray> \
227  operator BINOP(const marray &Lhs, const T &Rhs) { \
228  return Lhs BINOP marray(static_cast<DataT>(Rhs)); \
229  } \
230  template <typename T, typename BaseT = DataT> \
231  friend typename std::enable_if_t<std::is_convertible_v<T, DataT> && \
232  std::is_integral_v<T> && \
233  std::is_integral_v<BaseT>, \
234  marray> \
235  operator BINOP(const T &Lhs, const marray &Rhs) { \
236  return marray(static_cast<DataT>(Lhs)) BINOP Rhs; \
237  } \
238  template <typename T = DataT, \
239  typename = std::enable_if_t<std::is_integral_v<T>, marray>> \
240  friend marray &operator OPASSIGN(marray &Lhs, const marray &Rhs) { \
241  Lhs = Lhs BINOP Rhs; \
242  return Lhs; \
243  } \
244  template <std::size_t Num = NumElements, typename T = DataT> \
245  friend \
246  typename std::enable_if_t<Num != 1 && std::is_integral_v<T>, marray &> \
247  operator OPASSIGN(marray &Lhs, const DataT &Rhs) { \
248  Lhs = Lhs BINOP marray(Rhs); \
249  return Lhs; \
250  }
251 
252  __SYCL_BINOP(+, +=)
253  __SYCL_BINOP(-, -=)
254  __SYCL_BINOP(*, *=)
255  __SYCL_BINOP(/, /=)
256 
257  __SYCL_BINOP_INTEGRAL(%, %=)
258  __SYCL_BINOP_INTEGRAL(|, |=)
259  __SYCL_BINOP_INTEGRAL(&, &=)
260  __SYCL_BINOP_INTEGRAL(^, ^=)
261  __SYCL_BINOP_INTEGRAL(>>, >>=)
262  __SYCL_BINOP_INTEGRAL(<<, <<=)
263 #undef __SYCL_BINOP
264 #undef __SYCL_BINOP_INTEGRAL
265 
266 #ifdef __SYCL_RELLOGOP
267 #error "Undefine __SYCL_RELLOGOP macro"
268 #endif
269 
270 #ifdef __SYCL_RELLOGOP_INTEGRAL
271 #error "Undefine __SYCL_RELLOGOP_INTEGRAL macro"
272 #endif
273 
274 #define __SYCL_RELLOGOP(RELLOGOP) \
275  friend marray<bool, NumElements> operator RELLOGOP(const marray &Lhs, \
276  const marray &Rhs) { \
277  marray<bool, NumElements> Ret; \
278  for (size_t I = 0; I < NumElements; ++I) { \
279  Ret[I] = Lhs[I] RELLOGOP Rhs[I]; \
280  } \
281  return Ret; \
282  } \
283  template <typename T> \
284  friend typename std::enable_if_t<std::is_convertible_v<T, DataT> && \
285  (std::is_fundamental_v<T> || \
286  std::is_same_v<T, half>), \
287  marray<bool, NumElements>> \
288  operator RELLOGOP(const marray &Lhs, const T &Rhs) { \
289  return Lhs RELLOGOP marray(static_cast<const DataT &>(Rhs)); \
290  } \
291  template <typename T> \
292  friend typename std::enable_if_t<std::is_convertible_v<T, DataT> && \
293  (std::is_fundamental_v<T> || \
294  std::is_same_v<T, half>), \
295  marray<bool, NumElements>> \
296  operator RELLOGOP(const T &Lhs, const marray &Rhs) { \
297  return marray(static_cast<const DataT &>(Lhs)) RELLOGOP Rhs; \
298  }
299 
300  __SYCL_RELLOGOP(==)
301  __SYCL_RELLOGOP(!=)
302  __SYCL_RELLOGOP(>)
303  __SYCL_RELLOGOP(<)
304  __SYCL_RELLOGOP(>=)
305  __SYCL_RELLOGOP(<=)
306  __SYCL_RELLOGOP(&&)
307  __SYCL_RELLOGOP(||)
308 
309 #undef __SYCL_RELLOGOP
310 
311 #ifdef __SYCL_UOP
312 #error "Undefine __SYCL_UOP macro"
313 #endif
314 
315 #define __SYCL_UOP(UOP, OPASSIGN) \
316  template <typename T = DataT> \
317  friend std::enable_if_t< \
318  !std::is_same_v<typename std::remove_cv<T>::type, bool>, marray> \
319  &operator UOP(marray & Lhs) { \
320  Lhs OPASSIGN 1; \
321  return Lhs; \
322  } \
323  template <typename T = DataT> \
324  friend std::enable_if_t< \
325  !std::is_same_v<typename std::remove_cv<T>::type, bool>, marray> \
326  operator UOP(marray & Lhs, int) { \
327  marray Ret(Lhs); \
328  Lhs OPASSIGN 1; \
329  return Ret; \
330  } \
331  template <typename T = DataT> \
332  __SYCL_DEPRECATED( \
333  "++ and -- operators are deprecated for marray<bool, ...>") \
334  friend std::enable_if_t< \
335  std::is_same_v<typename std::remove_cv<T>::type, bool>, marray> \
336  &operator UOP(marray & Lhs) { \
337  Lhs OPASSIGN 1; \
338  return Lhs; \
339  } \
340  template <typename T = DataT> \
341  __SYCL_DEPRECATED( \
342  "++ and -- operators are deprecated for marray<bool, ...>") \
343  friend std::enable_if_t< \
344  std::is_same_v<typename std::remove_cv<T>::type, bool>, marray> \
345  operator UOP(marray & Lhs, int) { \
346  marray Ret(Lhs); \
347  Lhs OPASSIGN 1; \
348  return Ret; \
349  }
350 
351  __SYCL_UOP(++, +=)
352  __SYCL_UOP(--, -=)
353 #undef __SYCL_UOP
354 
355  // Available only when: dataT != cl_float && dataT != cl_double
356  // && dataT != cl_half
357  template <typename T = DataT>
358  friend std::enable_if_t<std::is_integral_v<T>, marray>
359  operator~(const marray &Lhs) {
360  marray Ret;
361  for (size_t I = 0; I < NumElements; ++I) {
362  Ret[I] = ~Lhs[I];
363  }
364  return Ret;
365  }
366 
367  friend marray<bool, NumElements> operator!(const marray &Lhs) {
368  marray<bool, NumElements> Ret;
369  for (size_t I = 0; I < NumElements; ++I) {
370  Ret[I] = !Lhs[I];
371  }
372  return Ret;
373  }
374 
375  friend marray operator+(const marray &Lhs) {
376  marray Ret;
377  for (size_t I = 0; I < NumElements; ++I) {
378  Ret[I] = +Lhs[I];
379  }
380  return Ret;
381  }
382 
383  friend marray operator-(const marray &Lhs) {
384  marray Ret;
385  for (size_t I = 0; I < NumElements; ++I) {
386  Ret[I] = -Lhs[I];
387  }
388  return Ret;
389  }
390 };
391 
392 // marray is device copyable if element type is device copyable.
393 template <typename T, std::size_t N>
394 struct is_device_copyable<sycl::marray<T, N>> : is_device_copyable<T> {};
395 
396 #define __SYCL_MAKE_MARRAY_ALIAS(ALIAS, TYPE, N) \
397  using ALIAS##N = sycl::marray<TYPE, N>;
398 
399 #define __SYCL_MAKE_MARRAY_ALIASES_FOR_ARITHMETIC_TYPES(N) \
400  __SYCL_MAKE_MARRAY_ALIAS(mbool, bool, N) \
401  __SYCL_MAKE_MARRAY_ALIAS(mchar, std::int8_t, N) \
402  __SYCL_MAKE_MARRAY_ALIAS(mshort, std::int16_t, N) \
403  __SYCL_MAKE_MARRAY_ALIAS(mint, std::int32_t, N) \
404  __SYCL_MAKE_MARRAY_ALIAS(mlong, std::int64_t, N) \
405  __SYCL_MAKE_MARRAY_ALIAS(mlonglong, std::int64_t, N) \
406  __SYCL_MAKE_MARRAY_ALIAS(mfloat, float, N) \
407  __SYCL_MAKE_MARRAY_ALIAS(mdouble, double, N) \
408  __SYCL_MAKE_MARRAY_ALIAS(mhalf, half, N)
409 
410 // FIXME: schar, longlong and ulonglong aliases are not defined by SYCL 2020
411 // spec, but they are preserved in SYCL 2020 mode, because SYCL-CTS is
412 // still using them.
413 // See KhronosGroup/SYCL-CTS#446 and KhronosGroup/SYCL-Docs#335
414 #define __SYCL_MAKE_MARRAY_ALIASES_FOR_SIGNED_AND_UNSIGNED_TYPES(N) \
415  __SYCL_MAKE_MARRAY_ALIAS(mschar, std::int8_t, N) \
416  __SYCL_MAKE_MARRAY_ALIAS(muchar, std::uint8_t, N) \
417  __SYCL_MAKE_MARRAY_ALIAS(mushort, std::uint16_t, N) \
418  __SYCL_MAKE_MARRAY_ALIAS(muint, std::uint32_t, N) \
419  __SYCL_MAKE_MARRAY_ALIAS(mulong, std::uint64_t, N) \
420  __SYCL_MAKE_MARRAY_ALIAS(mulonglong, std::uint64_t, N)
421 
422 #define __SYCL_MAKE_MARRAY_ALIASES_FOR_MARRAY_LENGTH(N) \
423  __SYCL_MAKE_MARRAY_ALIASES_FOR_ARITHMETIC_TYPES(N) \
424  __SYCL_MAKE_MARRAY_ALIASES_FOR_SIGNED_AND_UNSIGNED_TYPES(N)
425 
426 __SYCL_MAKE_MARRAY_ALIASES_FOR_MARRAY_LENGTH(2)
427 __SYCL_MAKE_MARRAY_ALIASES_FOR_MARRAY_LENGTH(3)
428 __SYCL_MAKE_MARRAY_ALIASES_FOR_MARRAY_LENGTH(4)
429 __SYCL_MAKE_MARRAY_ALIASES_FOR_MARRAY_LENGTH(8)
430 __SYCL_MAKE_MARRAY_ALIASES_FOR_MARRAY_LENGTH(16)
431 
432 #undef __SYCL_MAKE_MARRAY_ALIAS
433 #undef __SYCL_MAKE_MARRAY_ALIASES_FOR_ARITHMETIC_TYPES
434 #undef __SYCL_MAKE_MARRAY_ALIASES_FOR_SIGNED_AND_UNSIGNED_TYPES
435 #undef __SYCL_MAKE_MARRAY_ALIASES_FOR_MARRAY_LENGTH
436 
437 } // namespace _V1
438 } // namespace sycl
sycl::_V1::marray
Provides a cross-platform math array class template that works on SYCL devices as well as in host C++...
Definition: marray.hpp:49
sycl::_V1::marray::operator[]
const_reference operator[](std::size_t index) const
Definition: marray.hpp:146
sycl::_V1::marray::marray
constexpr marray(const ArgTN &...Args)
Definition: marray.hpp:126
sycl::_V1::marray::const_iterator
const Type * const_iterator
Definition: marray.hpp:57
sycl::_V1::marray::size
static constexpr std::size_t size() noexcept
Definition: marray.hpp:141
sycl::_V1::marray::const_reference
const Type & const_reference
Definition: marray.hpp:55
sycl::_V1::marray::operator!
friend marray< bool, NumElements > operator!(const marray &Lhs)
Definition: marray.hpp:367
sycl::_V1::marray::operator~
friend std::enable_if_t< std::is_integral_v< T >, marray > operator~(const marray &Lhs)
Definition: marray.hpp:359
sycl::_V1::marray::operator-
friend marray operator-(const marray &Lhs)
Definition: marray.hpp:383
sycl::_V1::marray::marray
friend class marray
Definition: marray.hpp:102
sycl::_V1::marray::begin
const_iterator begin() const
Definition: marray.hpp:161
sycl::_V1::marray::value_type
Type value_type
Definition: marray.hpp:53
sycl::_V1::marray::operator+
friend marray operator+(const marray &Lhs)
Definition: marray.hpp:375
sycl::_V1::marray::marray
constexpr marray(const marray< Type, NumElements > &Rhs)=default
sycl::_V1::marray::marray
constexpr marray(const Type &Arg)
Definition: marray.hpp:118
sycl::_V1::marray::begin
iterator begin()
Definition: marray.hpp:159
sycl::_V1::marray::operator[]
reference operator[](std::size_t index)
Definition: marray.hpp:144
sycl::_V1::marray::end
const_iterator end() const
Definition: marray.hpp:165
sycl::_V1::marray::iterator
Type * iterator
Definition: marray.hpp:56
sycl::_V1::marray::operator=
marray & operator=(const marray< Type, NumElements > &Rhs)=default
sycl::_V1::marray::end
iterator end()
Definition: marray.hpp:163
sycl::_V1::marray::marray
constexpr marray()
Definition: marray.hpp:116
sycl::_V1::marray::marray
constexpr marray(marray< Type, NumElements > &&Rhs)=default
sycl::_V1::marray::reference
Type & reference
Definition: marray.hpp:54
sycl::_V1::marray::operator=
marray & operator=(const Type &Rhs)
Definition: marray.hpp:151
sycl::_V1::ext::intel::esimd::rgba_channel::A
@ A
half_type.hpp
is_device_copyable.hpp
__SYCL_UOP
#define __SYCL_UOP(UOP, OPASSIGN)
Definition: marray.hpp:315
__SYCL_BINOP_INTEGRAL
#define __SYCL_BINOP_INTEGRAL(BINOP, OPASSIGN)
Definition: marray.hpp:212
__SYCL_MAKE_MARRAY_ALIASES_FOR_MARRAY_LENGTH
#define __SYCL_MAKE_MARRAY_ALIASES_FOR_MARRAY_LENGTH(N)
Definition: marray.hpp:422
__SYCL_RELLOGOP
#define __SYCL_RELLOGOP(RELLOGOP)
Definition: marray.hpp:274
__SYCL_BINOP
#define __SYCL_BINOP(BINOP, OPASSIGN)
Definition: marray.hpp:175
sycl
Definition: access.hpp:18
noexcept
_Abi const simd< _Tp, _Abi > & noexcept
Definition: simd.hpp:1324
sycl::_V1::detail::ArrayCreator
Definition: common.hpp:413
sycl::_V1::detail::GetMArrayArgsSize
Definition: marray.hpp:31
sycl::_V1::is_device_copyable
is_device_copyable is a user specializable class template to indicate that a type T is device copyabl...
Definition: is_device_copyable.hpp:44