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