DPC++ Runtime
Runtime libraries for oneAPI DPC++
common.hpp
Go to the documentation of this file.
1 //==---------- common.hpp ----- Common declarations ------------------------==//
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/detail/defines.hpp>
13 #include <sycl/detail/export.hpp>
14 #include <sycl/detail/pi.hpp>
16 
17 #include <cstdint>
18 #include <string>
19 
20 // Default signature enables the passing of user code location information to
21 // public methods as a default argument. If the end-user wants to disable the
22 // code location information, they must compile the code with
23 // -DDISABLE_SYCL_INSTRUMENTATION_METADATA flag
24 namespace sycl {
26 namespace detail {
27 
28 // The check for output iterator is commented out as it blocks set_final_data
29 // with void * argument to be used.
30 // TODO: Align these checks with the SYCL specification when the behaviour
31 // with void * is clarified.
32 template <typename DataT>
33 using EnableIfOutputPointerT = std::enable_if_t<
34  /*is_output_iterator<DataT>::value &&*/ std::is_pointer<DataT>::value>;
35 
36 template <typename DataT>
37 using EnableIfOutputIteratorT = std::enable_if_t<
38  /*is_output_iterator<DataT>::value &&*/ !std::is_pointer<DataT>::value>;
39 
40 #if !defined(NDEBUG) && (_MSC_VER > 1929 || __has_builtin(__builtin_FILE))
41 #define __CODELOC_FILE_NAME __builtin_FILE()
42 #else
43 #define __CODELOC_FILE_NAME nullptr
44 #endif
45 
46 #if _MSC_VER > 1929 || __has_builtin(__builtin_FUNCTION)
47 #define __CODELOC_FUNCTION __builtin_FUNCTION()
48 #else
49 #define __CODELOC_FUNCTION nullptr
50 #endif
51 
52 #if _MSC_VER > 1929 || __has_builtin(__builtin_LINE)
53 #define __CODELOC_LINE __builtin_LINE()
54 #else
55 #define __CODELOC_LINE 0
56 #endif
57 
58 #if _MSC_VER > 1929 || __has_builtin(__builtin_COLUMN)
59 #define __CODELOC_COLUMN __builtin_COLUMN()
60 #else
61 #define __CODELOC_COLUMN 0
62 #endif
63 
64 // Data structure that captures the user code location information using the
65 // builtin capabilities of the compiler
66 struct code_location {
67  static constexpr code_location
68  current(const char *fileName = __CODELOC_FILE_NAME,
69  const char *funcName = __CODELOC_FUNCTION,
70  unsigned long lineNo = __CODELOC_LINE,
71  unsigned long columnNo = __CODELOC_COLUMN) noexcept {
72  return code_location(fileName, funcName, lineNo, columnNo);
73  }
74 
75 #undef __CODELOC_FILE_NAME
76 #undef __CODELOC_FUNCTION
77 #undef __CODELOC_LINE
78 #undef __CODELOC_COLUMN
79 
80  constexpr code_location(const char *file, const char *func, int line,
81  int col) noexcept
82  : MFileName(file), MFunctionName(func), MLineNo(line), MColumnNo(col) {}
83 
84  constexpr code_location() noexcept
85  : MFileName(nullptr), MFunctionName(nullptr), MLineNo(0), MColumnNo(0) {}
86 
87  constexpr unsigned long lineNumber() const noexcept { return MLineNo; }
88  constexpr unsigned long columnNumber() const noexcept { return MColumnNo; }
89  constexpr const char *fileName() const noexcept { return MFileName; }
90  constexpr const char *functionName() const noexcept { return MFunctionName; }
91 
92 private:
93  const char *MFileName;
94  const char *MFunctionName;
95  unsigned long MLineNo;
96  unsigned long MColumnNo;
97 };
98 
99 // The C++ FE may instrument user calls with code location metadata.
100 // If it does then that will appear as an extra last argument.
101 // Having _TWO_ mid-param #ifdefs makes the functions very difficult to read.
102 // Here we simplify the &CodeLoc declaration to be _CODELOCPARAM(&CodeLoc) and
103 // _CODELOCARG(&CodeLoc).
104 
105 #ifndef DISABLE_SYCL_INSTRUMENTATION_METADATA
106 #define _CODELOCONLYPARAM(a) \
107  const detail::code_location a = detail::code_location::current()
108 #define _CODELOCPARAM(a) \
109  , const detail::code_location a = detail::code_location::current()
110 #define _CODELOCPARAMDEF(a) , const detail::code_location a
111 
112 #define _CODELOCARG(a)
113 #define _CODELOCFW(a) , a
114 #else
115 #define _CODELOCONLYPARAM(a)
116 #define _CODELOCPARAM(a)
117 
118 #define _CODELOCARG(a) const detail::code_location a = {}
119 #define _CODELOCFW(a)
120 #endif
121 
122 } // namespace detail
123 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
124 } // namespace sycl
125 
126 namespace sycl {
128 namespace detail {
129 
130 __SYCL_EXPORT const char *stringifyErrorCode(pi_int32 error);
131 
132 static inline std::string codeToString(pi_int32 code) {
133  return std::string(std::to_string(code) + " (" + stringifyErrorCode(code) +
134  ")");
135 }
136 
137 } // namespace detail
138 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
139 } // namespace sycl
140 
141 #ifdef __SYCL_DEVICE_ONLY__
142 // TODO remove this when 'assert' is supported in device code
143 #define __SYCL_ASSERT(x)
144 #else
145 #define __SYCL_ASSERT(x) assert(x)
146 #endif // #ifdef __SYCL_DEVICE_ONLY__
147 
148 #define __SYCL_PI_ERROR_REPORT \
149  "Native API failed. " /*__FILE__*/ \
150  /* TODO: replace __FILE__ to report only relative path*/ \
151  /* ":" __SYCL_STRINGIFY(__LINE__) ": " */ \
152  "Native API returns: "
153 
154 #ifndef __SYCL_SUPPRESS_PI_ERROR_REPORT
156 // TODO: rename all names with direct use of OCL/OPENCL to be backend agnostic.
157 #define __SYCL_REPORT_PI_ERR_TO_STREAM(expr) \
158  { \
159  auto code = expr; \
160  if (code != PI_SUCCESS) { \
161  std::cerr << __SYCL_PI_ERROR_REPORT << sycl::detail::codeToString(code) \
162  << std::endl; \
163  } \
164  }
165 #endif
166 
167 #ifndef SYCL_SUPPRESS_EXCEPTIONS
168 #include <sycl/exception.hpp>
169 // SYCL 1.2.1 exceptions
170 #define __SYCL_REPORT_PI_ERR_TO_EXC(expr, exc, str) \
171  { \
172  auto code = expr; \
173  if (code != PI_SUCCESS) { \
174  std::string err_str = \
175  str ? "\n" + std::string(str) + "\n" : std::string{}; \
176  throw exc(__SYCL_PI_ERROR_REPORT + sycl::detail::codeToString(code) + \
177  err_str, \
178  code); \
179  } \
180  }
181 #define __SYCL_REPORT_PI_ERR_TO_EXC_THROW(code, exc, str) \
182  __SYCL_REPORT_PI_ERR_TO_EXC(code, exc, str)
183 #define __SYCL_REPORT_PI_ERR_TO_EXC_BASE(code) \
184  __SYCL_REPORT_PI_ERR_TO_EXC(code, sycl::runtime_error, nullptr)
185 #else
186 #define __SYCL_REPORT_PI_ERR_TO_EXC_BASE(code) \
187  __SYCL_REPORT_PI_ERR_TO_STREAM(code)
188 #endif
189 // SYCL 2020 exceptions
190 #define __SYCL_REPORT_ERR_TO_EXC_VIA_ERRC(expr, errc) \
191  { \
192  auto code = expr; \
193  if (code != PI_SUCCESS) { \
194  throw sycl::exception(sycl::make_error_code(errc), \
195  __SYCL_PI_ERROR_REPORT + \
196  sycl::detail::codeToString(code)); \
197  } \
198  }
199 #define __SYCL_REPORT_ERR_TO_EXC_THROW_VIA_ERRC(code, errc) \
200  __SYCL_REPORT_ERR_TO_EXC_VIA_ERRC(code, errc)
201 
202 #ifdef __SYCL_SUPPRESS_PI_ERROR_REPORT
203 // SYCL 1.2.1 exceptions
204 #define __SYCL_CHECK_OCL_CODE(X) (void)(X)
205 #define __SYCL_CHECK_OCL_CODE_THROW(X, EXC, STR) \
206  { \
207  (void)(X); \
208  (void)(STR); \
209  }
210 #define __SYCL_CHECK_OCL_CODE_NO_EXC(X) (void)(X)
211 // SYCL 2020 exceptions
212 #define __SYCL_CHECK_CODE_THROW_VIA_ERRC(X, ERRC) (void)(X)
213 #else
214 // SYCL 1.2.1 exceptions
215 #define __SYCL_CHECK_OCL_CODE(X) __SYCL_REPORT_PI_ERR_TO_EXC_BASE(X)
216 #define __SYCL_CHECK_OCL_CODE_THROW(X, EXC, STR) \
217  __SYCL_REPORT_PI_ERR_TO_EXC_THROW(X, EXC, STR)
218 #define __SYCL_CHECK_OCL_CODE_NO_EXC(X) __SYCL_REPORT_PI_ERR_TO_STREAM(X)
219 // SYCL 2020 exceptions
220 #define __SYCL_CHECK_CODE_THROW_VIA_ERRC(X, ERRC) \
221  __SYCL_REPORT_ERR_TO_EXC_THROW_VIA_ERRC(X, ERRC)
222 #endif
223 
224 // Helper for enabling empty-base optimizations on MSVC.
225 // TODO: Remove this when MSVC has this optimization enabled by default.
226 #ifdef _MSC_VER
227 #define __SYCL_EBO __declspec(empty_bases)
228 #else
229 #define __SYCL_EBO
230 #endif
231 
232 namespace sycl {
234 namespace detail {
235 
236 // Helper function for extracting implementation from SYCL's interface objects.
237 // Note! This function relies on the fact that all SYCL interface classes
238 // contain "impl" field that points to implementation object. "impl" field
239 // should be accessible from this function.
240 //
241 // Note that due to a bug in MSVC compilers (including MSVC2019 v19.20), it
242 // may not recognize the usage of this function in friend member declarations
243 // if the template parameter name there is not equal to the name used here,
244 // i.e. 'Obj'. For example, using 'Obj' here and 'T' in such declaration
245 // would trigger that error in MSVC:
246 // template <class T>
247 // friend decltype(T::impl) detail::getSyclObjImpl(const T &SyclObject);
248 template <class Obj> decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject) {
249  assert(SyclObject.impl && "every constructor should create an impl");
250  return SyclObject.impl;
251 }
252 
253 // Returns the raw pointer to the impl object of given face object. The caller
254 // must make sure the returned pointer is not captured in a field or otherwise
255 // stored - i.e. must live only as on-stack value.
256 template <class T>
257 typename std::add_pointer_t<typename decltype(T::impl)::element_type>
258 getRawSyclObjImpl(const T &SyclObject) {
259  return SyclObject.impl.get();
260 }
261 
262 // Helper function for creation SYCL interface objects from implementations.
263 // Note! This function relies on the fact that all SYCL interface classes
264 // contain "impl" field that points to implementation object. "impl" field
265 // should be accessible from this function.
266 template <class T> T createSyclObjFromImpl(decltype(T::impl) ImplObj) {
267  return T(ImplObj);
268 }
269 
270 // Produces N-dimensional object of type T whose all components are initialized
271 // to given integer value.
272 template <int N, template <int> class T> struct InitializedVal {
273  template <int Val> static T<N> get();
274 };
275 
276 // Specialization for a one-dimensional type.
277 template <template <int> class T> struct InitializedVal<1, T> {
278  template <int Val> static T<1> get() { return T<1>{Val}; }
279 };
280 
281 // Specialization for a two-dimensional type.
282 template <template <int> class T> struct InitializedVal<2, T> {
283  template <int Val> static T<2> get() { return T<2>{Val, Val}; }
284 };
285 
286 // Specialization for a three-dimensional type.
287 template <template <int> class T> struct InitializedVal<3, T> {
288  template <int Val> static T<3> get() { return T<3>{Val, Val, Val}; }
289 };
290 
292 template <int NDims, int Dim, template <int> class LoopBoundTy, typename FuncTy,
293  template <int> class LoopIndexTy>
295  NDLoopIterateImpl(const LoopIndexTy<NDims> &LowerBound,
296  const LoopBoundTy<NDims> &Stride,
297  const LoopBoundTy<NDims> &UpperBound, FuncTy f,
298  LoopIndexTy<NDims> &Index) {
299  constexpr size_t AdjIdx = NDims - 1 - Dim;
300  for (Index[AdjIdx] = LowerBound[AdjIdx]; Index[AdjIdx] < UpperBound[AdjIdx];
301  Index[AdjIdx] += Stride[AdjIdx]) {
302 
303  NDLoopIterateImpl<NDims, Dim - 1, LoopBoundTy, FuncTy, LoopIndexTy>{
304  LowerBound, Stride, UpperBound, f, Index};
305  }
306  }
307 };
308 
309 // Specialization for Dim=0 to terminate recursion
310 template <int NDims, template <int> class LoopBoundTy, typename FuncTy,
311  template <int> class LoopIndexTy>
312 struct NDLoopIterateImpl<NDims, 0, LoopBoundTy, FuncTy, LoopIndexTy> {
313  NDLoopIterateImpl(const LoopIndexTy<NDims> &LowerBound,
314  const LoopBoundTy<NDims> &Stride,
315  const LoopBoundTy<NDims> &UpperBound, FuncTy f,
316  LoopIndexTy<NDims> &Index) {
317 
318  constexpr size_t AdjIdx = NDims - 1;
319  for (Index[AdjIdx] = LowerBound[AdjIdx]; Index[AdjIdx] < UpperBound[AdjIdx];
320  Index[AdjIdx] += Stride[AdjIdx]) {
321 
322  f(Index);
323  }
324  }
325 };
326 
333 template <int NDims> struct NDLoop {
337  template <template <int> class LoopBoundTy, typename FuncTy,
338  template <int> class LoopIndexTy = LoopBoundTy>
339  static __SYCL_ALWAYS_INLINE void iterate(const LoopBoundTy<NDims> &UpperBound,
340  FuncTy f) {
341  const LoopIndexTy<NDims> LowerBound =
343  const LoopBoundTy<NDims> Stride =
345  LoopIndexTy<NDims> Index =
347 
348  NDLoopIterateImpl<NDims, NDims - 1, LoopBoundTy, FuncTy, LoopIndexTy>{
349  LowerBound, Stride, UpperBound, f, Index};
350  }
351 
355  template <template <int> class LoopBoundTy, typename FuncTy,
356  template <int> class LoopIndexTy = LoopBoundTy>
357  static __SYCL_ALWAYS_INLINE void iterate(const LoopIndexTy<NDims> &LowerBound,
358  const LoopBoundTy<NDims> &Stride,
359  const LoopBoundTy<NDims> &UpperBound,
360  FuncTy f) {
361  LoopIndexTy<NDims> Index =
363  NDLoopIterateImpl<NDims, NDims - 1, LoopBoundTy, FuncTy, LoopIndexTy>{
364  LowerBound, Stride, UpperBound, f, Index};
365  }
366 };
367 
368 constexpr size_t getNextPowerOfTwoHelper(size_t Var, size_t Offset) {
369  return Offset != 64
370  ? getNextPowerOfTwoHelper(Var | (Var >> Offset), Offset * 2)
371  : Var;
372 }
373 
374 // Returns the smallest power of two not less than Var
375 constexpr size_t getNextPowerOfTwo(size_t Var) {
376  return getNextPowerOfTwoHelper(Var - 1, 1) + 1;
377 }
378 
379 // Returns linear index by given index and range
380 template <int Dims, template <int> class T, template <int> class U>
381 size_t getLinearIndex(const T<Dims> &Index, const U<Dims> &Range) {
382  size_t LinearIndex = 0;
383  for (int I = 0; I < Dims; ++I)
384  LinearIndex = LinearIndex * Range[I] + Index[I];
385  return LinearIndex;
386 }
387 
388 // Kernel set ID, used to group kernels (represented by OSModule & kernel name
389 // pairs) into disjoint sets based on the kernel distribution among device
390 // images.
391 using KernelSetId = size_t;
392 // Kernel set ID for kernels contained within the SPIR-V file specified via
393 // environment.
394 constexpr KernelSetId SpvFileKSId = 0;
396 
397 template <typename T> struct InlineVariableHelper {
398  static constexpr T value{};
399 };
400 
401 template <typename T> constexpr T InlineVariableHelper<T>::value;
402 
403 // The function extends or truncates number of dimensions of objects of id
404 // or ranges classes. When extending the new values are filled with
405 // DefaultValue, truncation just removes extra values.
406 template <int NewDim, int DefaultValue, template <int> class T, int OldDim>
407 static T<NewDim> convertToArrayOfN(T<OldDim> OldObj) {
409  const int CopyDims = NewDim > OldDim ? OldDim : NewDim;
410  for (int I = 0; I < CopyDims; ++I)
411  NewObj[I] = OldObj[I];
412  for (int I = CopyDims; I < NewDim; ++I)
413  NewObj[I] = DefaultValue;
414  return NewObj;
415 }
416 
417 } // namespace detail
418 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
419 } // namespace sycl
#define __CODELOC_COLUMN
Definition: common.hpp:61
#define __CODELOC_FILE_NAME
Definition: common.hpp:43
#define __CODELOC_LINE
Definition: common.hpp:55
#define __CODELOC_FUNCTION
Definition: common.hpp:49
#define __SYCL_INLINE_VER_NAMESPACE(X)
#define __SYCL_ALWAYS_INLINE
ESIMD_NODEBUG ESIMD_INLINE std::enable_if_t< __ESIMD_DNS::is_fp_or_dword_type< T >::value &&std::is_floating_point< T >::value, sycl::ext::intel::esimd::simd< T, SZ > > line(sycl::ext::intel::esimd::simd< T, 4 > src0, sycl::ext::intel::esimd::simd< T, SZ > src1, Sat sat={})
Linear equation.
Definition: math.hpp:921
static T< NewDim > convertToArrayOfN(T< OldDim > OldObj)
Definition: common.hpp:407
std::enable_if_t< std::is_pointer< DataT >::value > EnableIfOutputPointerT
Definition: common.hpp:34
std::add_pointer_t< typename decltype(T::impl)::element_type > getRawSyclObjImpl(const T &SyclObject)
Definition: common.hpp:258
constexpr KernelSetId SpvFileKSId
Definition: common.hpp:394
size_t getLinearIndex(const T< Dims > &Index, const U< Dims > &Range)
Definition: common.hpp:381
constexpr KernelSetId LastKSId
Definition: common.hpp:395
const char * stringifyErrorCode(pi_int32 error)
Definition: common.cpp:16
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:248
std::enable_if_t< !std::is_pointer< DataT >::value > EnableIfOutputIteratorT
Definition: common.hpp:38
T createSyclObjFromImpl(decltype(T::impl) ImplObj)
Definition: common.hpp:266
constexpr size_t getNextPowerOfTwo(size_t Var)
Definition: common.hpp:375
size_t KernelSetId
Definition: common.hpp:391
static std::string codeToString(pi_int32 code)
Definition: common.hpp:132
constexpr size_t getNextPowerOfTwoHelper(size_t Var, size_t Offset)
Definition: common.hpp:368
---— Error handling, matching OpenCL plugin semantics.
Definition: access.hpp:14
int32_t pi_int32
Definition: pi.h:106
C++ wrapper of extern "C" PI interfaces.
NDLoopIterateImpl(const LoopIndexTy< NDims > &LowerBound, const LoopBoundTy< NDims > &Stride, const LoopBoundTy< NDims > &UpperBound, FuncTy f, LoopIndexTy< NDims > &Index)
Definition: common.hpp:313
Helper class for the NDLoop.
Definition: common.hpp:294
NDLoopIterateImpl(const LoopIndexTy< NDims > &LowerBound, const LoopBoundTy< NDims > &Stride, const LoopBoundTy< NDims > &UpperBound, FuncTy f, LoopIndexTy< NDims > &Index)
Definition: common.hpp:295
Generates an NDims-dimensional perfect loop nest.
Definition: common.hpp:333
static __SYCL_ALWAYS_INLINE void iterate(const LoopBoundTy< NDims > &UpperBound, FuncTy f)
Generates ND loop nest with {0,..0} .
Definition: common.hpp:339
static __SYCL_ALWAYS_INLINE void iterate(const LoopIndexTy< NDims > &LowerBound, const LoopBoundTy< NDims > &Stride, const LoopBoundTy< NDims > &UpperBound, FuncTy f)
Generates ND loop nest with LowerBound .
Definition: common.hpp:357
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
constexpr code_location(const char *file, const char *func, int line, int col) noexcept
Definition: common.hpp:80
constexpr code_location() noexcept
Definition: common.hpp:84
constexpr unsigned long columnNumber() const noexcept
Definition: common.hpp:88
constexpr const char * fileName() const noexcept
Definition: common.hpp:89
constexpr const char * functionName() const noexcept
Definition: common.hpp:90
constexpr unsigned long lineNumber() const noexcept
Definition: common.hpp:87