DPC++ Runtime
Runtime libraries for oneAPI Data Parallel C++
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 <CL/sycl/detail/cl.h>
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
25 namespace sycl {
26 namespace detail {
27 
28 #if !defined(NDEBUG) && (_MSC_VER > 1929 || __has_builtin(__builtin_FILE))
29 #define __CODELOC_FILE_NAME __builtin_FILE()
30 #else
31 #define __CODELOC_FILE_NAME nullptr
32 #endif
33 
34 #if _MSC_VER > 1929 || __has_builtin(__builtin_FUNCTION)
35 #define __CODELOC_FUNCTION __builtin_FUNCTION()
36 #else
37 #define __CODELOC_FUNCTION nullptr
38 #endif
39 
40 #if _MSC_VER > 1929 || __has_builtin(__builtin_LINE)
41 #define __CODELOC_LINE __builtin_LINE()
42 #else
43 #define __CODELOC_LINE 0
44 #endif
45 
46 #if _MSC_VER > 1929 || __has_builtin(__builtin_COLUMN)
47 #define __CODELOC_COLUMN __builtin_COLUMN()
48 #else
49 #define __CODELOC_COLUMN 0
50 #endif
51 
52 // Data structure that captures the user code location information using the
53 // builtin capabilities of the compiler
54 struct code_location {
55  static constexpr code_location
56  current(const char *fileName = __CODELOC_FILE_NAME,
57  const char *funcName = __CODELOC_FUNCTION,
58  unsigned long lineNo = __CODELOC_LINE,
59  unsigned long columnNo = __CODELOC_COLUMN) noexcept {
60  return code_location(fileName, funcName, lineNo, columnNo);
61  }
62 
63 #undef __CODELOC_FILE_NAME
64 #undef __CODELOC_FUNCTION
65 #undef __CODELOC_LINE
66 #undef __CODELOC_COLUMN
67 
68  constexpr code_location(const char *file, const char *func, int line,
69  int col) noexcept
70  : MFileName(file), MFunctionName(func), MLineNo(line), MColumnNo(col) {}
71 
72  constexpr code_location() noexcept
73  : MFileName(nullptr), MFunctionName(nullptr), MLineNo(0), MColumnNo(0) {}
74 
75  constexpr unsigned long lineNumber() const noexcept { return MLineNo; }
76  constexpr unsigned long columnNumber() const noexcept { return MColumnNo; }
77  constexpr const char *fileName() const noexcept { return MFileName; }
78  constexpr const char *functionName() const noexcept { return MFunctionName; }
79 
80 private:
81  const char *MFileName;
82  const char *MFunctionName;
83  unsigned long MLineNo;
84  unsigned long MColumnNo;
85 };
86 } // namespace detail
87 } // namespace sycl
88 } // __SYCL_INLINE_NAMESPACE(cl)
89 
91 namespace sycl {
92 namespace detail {
93 
94 __SYCL_EXPORT const char *stringifyErrorCode(cl_int error);
95 
96 static inline std::string codeToString(cl_int code) {
97  return std::string(std::to_string(code) + " (" + stringifyErrorCode(code) +
98  ")");
99 }
100 
101 } // namespace detail
102 } // namespace sycl
103 } // __SYCL_INLINE_NAMESPACE(cl)
104 
105 #ifdef __SYCL_DEVICE_ONLY__
106 // TODO remove this when 'assert' is supported in device code
107 #define __SYCL_ASSERT(x)
108 #else
109 #define __SYCL_ASSERT(x) assert(x)
110 #endif // #ifdef __SYCL_DEVICE_ONLY__
111 
112 #define __SYCL_OCL_ERROR_REPORT \
113  "Native API failed. " /*__FILE__*/ \
114  /* TODO: replace __FILE__ to report only relative path*/ \
115  /* ":" __SYCL_STRINGIFY(__LINE__) ": " */ \
116  "Native API returns: "
117 
118 #ifndef __SYCL_SUPPRESS_OCL_ERROR_REPORT
119 #include <iostream>
120 // TODO: rename all names with direct use of OCL/OPENCL to be backend agnostic.
121 #define __SYCL_REPORT_OCL_ERR_TO_STREAM(expr) \
122  { \
123  auto code = expr; \
124  if (code != CL_SUCCESS) { \
125  std::cerr << __SYCL_OCL_ERROR_REPORT \
126  << cl::sycl::detail::codeToString(code) << std::endl; \
127  } \
128  }
129 #endif
130 
131 #ifndef SYCL_SUPPRESS_EXCEPTIONS
132 #include <CL/sycl/exception.hpp>
133 // SYCL 1.2.1 exceptions
134 #define __SYCL_REPORT_OCL_ERR_TO_EXC(expr, exc) \
135  { \
136  auto code = expr; \
137  if (code != CL_SUCCESS) { \
138  throw exc(__SYCL_OCL_ERROR_REPORT + \
139  cl::sycl::detail::codeToString(code), \
140  code); \
141  } \
142  }
143 #define __SYCL_REPORT_OCL_ERR_TO_EXC_THROW(code, exc) \
144  __SYCL_REPORT_OCL_ERR_TO_EXC(code, exc)
145 #define __SYCL_REPORT_OCL_ERR_TO_EXC_BASE(code) \
146  __SYCL_REPORT_OCL_ERR_TO_EXC(code, cl::sycl::runtime_error)
147 #else
148 #define __SYCL_REPORT_OCL_ERR_TO_EXC_BASE(code) \
149  __SYCL_REPORT_OCL_ERR_TO_STREAM(code)
150 #endif
151 // SYCL 2020 exceptions
152 #define __SYCL_REPORT_ERR_TO_EXC_VIA_ERRC(expr, errc) \
153  { \
154  auto code = expr; \
155  if (code != CL_SUCCESS) { \
156  throw sycl::exception(sycl::make_error_code(errc), \
157  __SYCL_OCL_ERROR_REPORT + \
158  cl::sycl::detail::codeToString(code)); \
159  } \
160  }
161 #define __SYCL_REPORT_ERR_TO_EXC_THROW_VIA_ERRC(code, errc) \
162  __SYCL_REPORT_ERR_TO_EXC_VIA_ERRC(code, errc)
163 
164 #ifdef __SYCL_SUPPRESS_OCL_ERROR_REPORT
165 // SYCL 1.2.1 exceptions
166 #define __SYCL_CHECK_OCL_CODE(X) (void)(X)
167 #define __SYCL_CHECK_OCL_CODE_THROW(X, EXC) (void)(X)
168 #define __SYCL_CHECK_OCL_CODE_NO_EXC(X) (void)(X)
169 // SYCL 2020 exceptions
170 #define __SYCL_CHECK_CODE_THROW_VIA_ERRC(X, ERRC) (void)(X)
171 #else
172 // SYCL 1.2.1 exceptions
173 #define __SYCL_CHECK_OCL_CODE(X) __SYCL_REPORT_OCL_ERR_TO_EXC_BASE(X)
174 #define __SYCL_CHECK_OCL_CODE_THROW(X, EXC) \
175  __SYCL_REPORT_OCL_ERR_TO_EXC_THROW(X, EXC)
176 #define __SYCL_CHECK_OCL_CODE_NO_EXC(X) __SYCL_REPORT_OCL_ERR_TO_STREAM(X)
177 // SYCL 2020 exceptions
178 #define __SYCL_CHECK_CODE_THROW_VIA_ERRC(X, ERRC) \
179  __SYCL_REPORT_ERR_TO_EXC_THROW_VIA_ERRC(X, ERRC)
180 #endif
181 
183 namespace sycl {
184 namespace detail {
185 
186 // Helper function for extracting implementation from SYCL's interface objects.
187 // Note! This function relies on the fact that all SYCL interface classes
188 // contain "impl" field that points to implementation object. "impl" field
189 // should be accessible from this function.
190 //
191 // Note that due to a bug in MSVC compilers (including MSVC2019 v19.20), it
192 // may not recognize the usage of this function in friend member declarations
193 // if the template parameter name there is not equal to the name used here,
194 // i.e. 'Obj'. For example, using 'Obj' here and 'T' in such declaration
195 // would trigger that error in MSVC:
196 // template <class T>
197 // friend decltype(T::impl) detail::getSyclObjImpl(const T &SyclObject);
198 template <class Obj> decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject) {
199  return SyclObject.impl;
200 }
201 
202 // Returns the raw pointer to the impl object of given face object. The caller
203 // must make sure the returned pointer is not captured in a field or otherwise
204 // stored - i.e. must live only as on-stack value.
205 template <class T>
206 typename detail::add_pointer_t<typename decltype(T::impl)::element_type>
207 getRawSyclObjImpl(const T &SyclObject) {
208  return SyclObject.impl.get();
209 }
210 
211 // Helper function for creation SYCL interface objects from implementations.
212 // Note! This function relies on the fact that all SYCL interface classes
213 // contain "impl" field that points to implementation object. "impl" field
214 // should be accessible from this function.
215 template <class T> T createSyclObjFromImpl(decltype(T::impl) ImplObj) {
216  return T(ImplObj);
217 }
218 
219 // Produces N-dimensional object of type T whose all components are initialized
220 // to given integer value.
221 template <int N, template <int> class T> struct InitializedVal {
222  template <int Val> static T<N> get();
223 };
224 
225 // Specialization for a one-dimensional type.
226 template <template <int> class T> struct InitializedVal<1, T> {
227  template <int Val> static T<1> get() { return T<1>{Val}; }
228 };
229 
230 // Specialization for a two-dimensional type.
231 template <template <int> class T> struct InitializedVal<2, T> {
232  template <int Val> static T<2> get() { return T<2>{Val, Val}; }
233 };
234 
235 // Specialization for a three-dimensional type.
236 template <template <int> class T> struct InitializedVal<3, T> {
237  template <int Val> static T<3> get() { return T<3>{Val, Val, Val}; }
238 };
239 
241 template <int NDIMS, int DIM, template <int> class LoopBoundTy, typename FuncTy,
242  template <int> class LoopIndexTy>
244  NDLoopIterateImpl(const LoopIndexTy<NDIMS> &LowerBound,
245  const LoopBoundTy<NDIMS> &Stride,
246  const LoopBoundTy<NDIMS> &UpperBound, FuncTy f,
247  LoopIndexTy<NDIMS> &Index) {
248  constexpr size_t AdjIdx = NDIMS - 1 - DIM;
249  for (Index[AdjIdx] = LowerBound[AdjIdx]; Index[AdjIdx] < UpperBound[AdjIdx];
250  Index[AdjIdx] += Stride[AdjIdx]) {
251 
252  NDLoopIterateImpl<NDIMS, DIM - 1, LoopBoundTy, FuncTy, LoopIndexTy>{
253  LowerBound, Stride, UpperBound, f, Index};
254  }
255  }
256 };
257 
258 // Specialization for DIM=0 to terminate recursion
259 template <int NDIMS, template <int> class LoopBoundTy, typename FuncTy,
260  template <int> class LoopIndexTy>
261 struct NDLoopIterateImpl<NDIMS, 0, LoopBoundTy, FuncTy, LoopIndexTy> {
262  NDLoopIterateImpl(const LoopIndexTy<NDIMS> &LowerBound,
263  const LoopBoundTy<NDIMS> &Stride,
264  const LoopBoundTy<NDIMS> &UpperBound, FuncTy f,
265  LoopIndexTy<NDIMS> &Index) {
266 
267  constexpr size_t AdjIdx = NDIMS - 1;
268  for (Index[AdjIdx] = LowerBound[AdjIdx]; Index[AdjIdx] < UpperBound[AdjIdx];
269  Index[AdjIdx] += Stride[AdjIdx]) {
270 
271  f(Index);
272  }
273  }
274 };
275 
282 template <int NDIMS> struct NDLoop {
286  template <template <int> class LoopBoundTy, typename FuncTy,
287  template <int> class LoopIndexTy = LoopBoundTy>
288  static __SYCL_ALWAYS_INLINE void iterate(const LoopBoundTy<NDIMS> &UpperBound,
289  FuncTy f) {
290  const LoopIndexTy<NDIMS> LowerBound =
292  const LoopBoundTy<NDIMS> Stride =
294  LoopIndexTy<NDIMS> Index =
296 
297  NDLoopIterateImpl<NDIMS, NDIMS - 1, LoopBoundTy, FuncTy, LoopIndexTy>{
298  LowerBound, Stride, UpperBound, f, Index};
299  }
300 
304  template <template <int> class LoopBoundTy, typename FuncTy,
305  template <int> class LoopIndexTy = LoopBoundTy>
306  static __SYCL_ALWAYS_INLINE void iterate(const LoopIndexTy<NDIMS> &LowerBound,
307  const LoopBoundTy<NDIMS> &Stride,
308  const LoopBoundTy<NDIMS> &UpperBound,
309  FuncTy f) {
310  LoopIndexTy<NDIMS> Index =
312  NDLoopIterateImpl<NDIMS, NDIMS - 1, LoopBoundTy, FuncTy, LoopIndexTy>{
313  LowerBound, Stride, UpperBound, f, Index};
314  }
315 };
316 
317 constexpr size_t getNextPowerOfTwoHelper(size_t Var, size_t Offset) {
318  return Offset != 64
319  ? getNextPowerOfTwoHelper(Var | (Var >> Offset), Offset * 2)
320  : Var;
321 }
322 
323 // Returns the smallest power of two not less than Var
324 constexpr size_t getNextPowerOfTwo(size_t Var) {
325  return getNextPowerOfTwoHelper(Var - 1, 1) + 1;
326 }
327 
328 // Returns linear index by given index and range
329 template <int Dims, template <int> class T, template <int> class U>
330 size_t getLinearIndex(const T<Dims> &Index, const U<Dims> &Range) {
331  size_t LinearIndex = 0;
332  for (int I = 0; I < Dims; ++I)
333  LinearIndex = LinearIndex * Range[I] + Index[I];
334  return LinearIndex;
335 }
336 
337 // Kernel set ID, used to group kernels (represented by OSModule & kernel name
338 // pairs) into disjoint sets based on the kernel distribution among device
339 // images.
340 using KernelSetId = size_t;
341 // Kernel set ID for kernels contained within the SPIR-V file specified via
342 // environment.
343 constexpr KernelSetId SpvFileKSId = 0;
345 
346 template <typename T> struct InlineVariableHelper {
347  static constexpr T value{};
348 };
349 
350 template <typename T> constexpr T InlineVariableHelper<T>::value;
351 } // namespace detail
352 } // namespace sycl
353 } // __SYCL_INLINE_NAMESPACE(cl)
__CODELOC_COLUMN
#define __CODELOC_COLUMN
Definition: common.hpp:49
cl::sycl::detail::getRawSyclObjImpl
detail::add_pointer_t< typename decltype(T::impl)::element_type > getRawSyclObjImpl(const T &SyclObject)
Definition: common.hpp:207
cl::sycl::detail::getLinearIndex
size_t getLinearIndex(const T< Dims > &Index, const U< Dims > &Range)
Definition: common.hpp:330
cl::sycl::detail::get< 0 >
Definition: tuple.hpp:75
cl::sycl::detail::getNextPowerOfTwoHelper
constexpr size_t getNextPowerOfTwoHelper(size_t Var, size_t Offset)
Definition: common.hpp:317
T
__CODELOC_FUNCTION
#define __CODELOC_FUNCTION
Definition: common.hpp:37
cl::sycl::detail::SpvFileKSId
constexpr KernelSetId SpvFileKSId
Definition: common.hpp:343
defines_elementary.hpp
cl::sycl::detail::code_location::code_location
constexpr code_location(const char *file, const char *func, int line, int col) noexcept
Definition: common.hpp:68
cl::sycl::detail::createSyclObjFromImpl
T createSyclObjFromImpl(decltype(T::impl) ImplObj)
Definition: common.hpp:215
cl::sycl::detail::code_location::fileName
constexpr const char * fileName() const noexcept
Definition: common.hpp:77
cl::sycl::detail::InitializedVal
Definition: common.hpp:221
cl::sycl::detail::LastKSId
constexpr KernelSetId LastKSId
Definition: common.hpp:344
detail
Definition: pi_opencl.cpp:86
cl::sycl::detail::NDLoopIterateImpl
Helper class for the NDLoop.
Definition: common.hpp:243
__CODELOC_LINE
#define __CODELOC_LINE
Definition: common.hpp:43
cl::sycl::detail::InlineVariableHelper
Definition: common.hpp:346
cl::sycl::detail::KernelSetId
size_t KernelSetId
Definition: common.hpp:340
cl::sycl::ext::intel::experimental::esimd::line
ESIMD_NODEBUG ESIMD_INLINE std::enable_if_t< detail::is_fp_or_dword_type< T >::value &&std::is_floating_point< T >::value, simd< T, SZ > > line(simd< T, 4 > src0, simd< T, SZ > src1, int flag=saturation_off)
FIXME: linear equation.
Definition: math.hpp:1148
cl::sycl::detail::code_location::columnNumber
constexpr unsigned long columnNumber() const noexcept
Definition: common.hpp:76
cl::sycl::detail::code_location::current
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:56
cl::sycl::detail::code_location
Definition: common.hpp:54
cl::sycl::detail::InitializedVal< 1, T >::get
static T< 1 > get()
Definition: common.hpp:227
cl::sycl::detail::get
Definition: tuple.hpp:59
cl::sycl::detail::InitializedVal< 2, T >::get
static T< 2 > get()
Definition: common.hpp:232
stl_type_traits.hpp
cl::sycl::detail::NDLoop
Generates an NDIMS-dimensional perfect loop nest.
Definition: common.hpp:282
export.hpp
__SYCL_ALWAYS_INLINE
#define __SYCL_ALWAYS_INLINE
Definition: defines_elementary.hpp:29
cl.h
defines.hpp
cl
We provide new interfaces for matrix muliply in this patch:
Definition: access.hpp:13
cl::sycl::detail::stringifyErrorCode
const char * stringifyErrorCode(cl_int error)
Definition: common.cpp:16
cl::sycl::detail::NDLoopIterateImpl::NDLoopIterateImpl
NDLoopIterateImpl(const LoopIndexTy< NDIMS > &LowerBound, const LoopBoundTy< NDIMS > &Stride, const LoopBoundTy< NDIMS > &UpperBound, FuncTy f, LoopIndexTy< NDIMS > &Index)
Definition: common.hpp:244
cl::sycl::detail::getNextPowerOfTwo
constexpr size_t getNextPowerOfTwo(size_t Var)
Definition: common.hpp:324
std::get
constexpr tuple_element< I, tuple< Types... > >::type & get(cl::sycl::detail::tuple< Types... > &Arg) noexcept
Definition: tuple.hpp:199
cl::sycl::detail::add_pointer_t
typename std::add_pointer< T >::type add_pointer_t
Definition: stl_type_traits.hpp:37
cl::sycl::detail::getSyclObjImpl
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:198
cl::sycl::cl_int
std::int32_t cl_int
Definition: aliases.hpp:82
exception.hpp
cl::sycl::detail::code_location::functionName
constexpr const char * functionName() const noexcept
Definition: common.hpp:78
cl::sycl::detail::NDLoopIterateImpl< NDIMS, 0, LoopBoundTy, FuncTy, LoopIndexTy >::NDLoopIterateImpl
NDLoopIterateImpl(const LoopIndexTy< NDIMS > &LowerBound, const LoopBoundTy< NDIMS > &Stride, const LoopBoundTy< NDIMS > &UpperBound, FuncTy f, LoopIndexTy< NDIMS > &Index)
Definition: common.hpp:262
cl::sycl::detail::codeToString
static std::string codeToString(cl_int code)
Definition: common.hpp:96
__CODELOC_FILE_NAME
#define __CODELOC_FILE_NAME
Definition: common.hpp:31
cl::sycl::detail::NDLoop::iterate
static __SYCL_ALWAYS_INLINE void iterate(const LoopBoundTy< NDIMS > &UpperBound, FuncTy f)
Generates ND loop nest with {0,..0} .
Definition: common.hpp:288
cl::sycl::detail::InitializedVal< 3, T >::get
static T< 3 > get()
Definition: common.hpp:237
cl::sycl::detail::code_location::lineNumber
constexpr unsigned long lineNumber() const noexcept
Definition: common.hpp:75
cl::sycl::detail::code_location::code_location
constexpr code_location() noexcept
Definition: common.hpp:72
cl::sycl::detail::NDLoop::iterate
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:306
__SYCL_INLINE_NAMESPACE
#define __SYCL_INLINE_NAMESPACE(X)
Definition: defines_elementary.hpp:12