DPC++ Runtime
Runtime libraries for oneAPI DPC++
item.hpp
Go to the documentation of this file.
1 //==------------ item.hpp --- SYCL iteration item --------------------------==//
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> // for __SYCL_ASSUME_INT
12 #include <sycl/detail/defines_elementary.hpp> // for __SYCL_ALWAYS_INLINE, __SYC...
13 #include <sycl/detail/helpers.hpp> // for Builder
14 #include <sycl/detail/item_base.hpp> // for id, range, ItemBase
15 #include <sycl/exception.hpp> // for make_error_code, errc, exce...
16 #include <sycl/id.hpp> // for id, item
17 #include <sycl/range.hpp> // for range
18 
19 #include <cstddef> // for size_t
20 #include <type_traits> // for enable_if_t, conditional_t
21 
22 namespace sycl {
23 inline namespace _V1 {
24 
25 namespace detail {
26 class Builder;
27 template <typename TransformedArgType, int Dims, typename KernelType>
28 class RoundedRangeKernel;
29 template <typename TransformedArgType, int Dims, typename KernelType>
30 class RoundedRangeKernelWithKH;
31 } // namespace detail
32 
37 template <int Dimensions = 1, bool with_offset = true> class item {
38 public:
39  static constexpr int dimensions = Dimensions;
40 
41 private:
42 #ifndef __SYCL_DISABLE_ITEM_TO_INT_CONV__
43  /* Helper class for conversion operator. Void type is not suitable. User
44  * cannot even try to get address of the operator __private_class(). User
45  * may try to get an address of operator void() and will get the
46  * compile-time error */
47  class __private_class;
48 
49  template <bool B, typename T>
50  using EnableIfT = std::conditional_t<B, T, __private_class>;
51 #endif // __SYCL_DISABLE_ITEM_TO_INT_CONV__
52 public:
53  item() = delete;
54 
55  id<Dimensions> get_id() const { return MImpl.MIndex; }
56 
57  size_t __SYCL_ALWAYS_INLINE get_id(int Dimension) const {
58  size_t Id = MImpl.MIndex[Dimension];
60  return Id;
61  }
62 
63  size_t __SYCL_ALWAYS_INLINE operator[](int Dimension) const {
64  size_t Id = MImpl.MIndex[Dimension];
66  return Id;
67  }
68 
69  range<Dimensions> get_range() const { return MImpl.MExtent; }
70 
71  size_t __SYCL_ALWAYS_INLINE get_range(int Dimension) const {
72  size_t Id = MImpl.MExtent[Dimension];
74  return Id;
75  }
76 #ifndef __SYCL_DISABLE_ITEM_TO_INT_CONV__
77  operator EnableIfT<Dimensions == 1, std::size_t>() const { return get_id(0); }
78 #endif // __SYCL_DISABLE_ITEM_TO_INT_CONV__
79  template <bool has_offset = with_offset>
80  __SYCL2020_DEPRECATED("offsets are deprecated in SYCL2020")
81  std::enable_if_t<has_offset, id<Dimensions>> get_offset() const {
82  return MImpl.MOffset;
83  }
84 
85  template <bool has_offset = with_offset>
86  __SYCL2020_DEPRECATED("offsets are deprecated in SYCL2020")
87  std::enable_if_t<has_offset, size_t> __SYCL_ALWAYS_INLINE
88  get_offset(int Dimension) const {
89  size_t Id = MImpl.MOffset[Dimension];
91  return Id;
92  }
93 
94  template <bool has_offset = with_offset>
95  operator std::enable_if_t<!has_offset, item<Dimensions, true>>() const {
96  return detail::Builder::createItem<Dimensions, true>(
97  MImpl.MExtent, MImpl.MIndex, /*Offset*/ {});
98  }
99 
101  size_t Id = MImpl.get_linear_id();
102  __SYCL_ASSUME_INT(Id);
103  return Id;
104  }
105 
106  item(const item &rhs) = default;
107 
109 
110  item &operator=(const item &rhs) = default;
111 
112  item &operator=(item &&rhs) = default;
113 
114  bool operator==(const item &rhs) const { return rhs.MImpl == MImpl; }
115 
116  bool operator!=(const item &rhs) const { return rhs.MImpl != MImpl; }
117 
118 protected:
119  template <bool has_offset = with_offset>
120  item(std::enable_if_t<has_offset, const range<Dimensions>> &extent,
121  const id<Dimensions> &index, const id<Dimensions> &offset)
122  : MImpl{extent, index, offset} {}
123 
124  template <bool has_offset = with_offset>
125  item(std::enable_if_t<!has_offset, const range<Dimensions>> &extent,
126  const id<Dimensions> &index)
127  : MImpl{extent, index} {}
128 
129  friend class detail::Builder;
130 
131 private:
133 };
134 
135 template <int Dims>
136 __SYCL_DEPRECATED("use sycl::ext::oneapi::experimental::this_item() instead")
137 item<Dims> this_item() {
138 #ifdef __SYCL_DEVICE_ONLY__
139  return detail::Builder::getElement(detail::declptr<item<Dims>>());
140 #else
141  throw sycl::exception(
142  sycl::make_error_code(sycl::errc::feature_not_supported),
143  "Free function calls are not supported on host");
144 #endif
145 }
146 
147 namespace ext::oneapi::experimental {
148 template <int Dims> item<Dims> this_item() {
149 #ifdef __SYCL_DEVICE_ONLY__
150  return sycl::detail::Builder::getElement(sycl::detail::declptr<item<Dims>>());
151 #else
152  throw sycl::exception(
153  sycl::make_error_code(sycl::errc::feature_not_supported),
154  "Free function calls are not supported on host");
155 #endif
156 }
157 } // namespace ext::oneapi::experimental
158 } // namespace _V1
159 } // namespace sycl
A unique identifier of an item in an index space.
Definition: id.hpp:36
Identifies an instance of the function object executing at each point in a range.
Definition: item.hpp:37
item(std::enable_if_t<!has_offset, const range< Dimensions >> &extent, const id< Dimensions > &index)
Definition: item.hpp:125
item(item< Dimensions, with_offset > &&rhs)=default
size_t __SYCL_ALWAYS_INLINE operator[](int Dimension) const
Definition: item.hpp:63
bool operator==(const item &rhs) const
Definition: item.hpp:114
item & operator=(item &&rhs)=default
id< Dimensions > get_id() const
Definition: item.hpp:55
item(std::enable_if_t< has_offset, const range< Dimensions >> &extent, const id< Dimensions > &index, const id< Dimensions > &offset)
Definition: item.hpp:120
std::enable_if_t< has_offset, id< Dimensions > > get_offset() const
Definition: item.hpp:81
size_t __SYCL_ALWAYS_INLINE get_range(int Dimension) const
Definition: item.hpp:71
size_t __SYCL_ALWAYS_INLINE get_id(int Dimension) const
Definition: item.hpp:57
static constexpr int dimensions
Definition: item.hpp:39
range< Dimensions > get_range() const
Definition: item.hpp:69
item(const item &rhs)=default
size_t __SYCL_ALWAYS_INLINE get_linear_id() const
Definition: item.hpp:100
item & operator=(const item &rhs)=default
bool operator!=(const item &rhs) const
Definition: item.hpp:116
#define __SYCL_ASSUME_INT(x)
Definition: defines.hpp:17
#define __SYCL_ALWAYS_INLINE
class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(local_accessor) local_accessor class __SYCL_EBO __SYCL_SPECIAL_CLASS Dimensions
Definition: accessor.hpp:3233
signed char __SYCL2020_DEPRECATED
Definition: aliases.hpp:94
item< Dims > this_item()
Definition: item.hpp:137
__SYCL_DEPRECATED("This is a deprecated argument type for SYCL nan built-in function.") std
std::error_code make_error_code(sycl::errc E) noexcept
Constructs an error code using e and sycl_category()
Definition: exception.cpp:94
Definition: access.hpp:18