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>
12 #include <sycl/detail/helpers.hpp>
15 #include <sycl/id.hpp>
16 #include <sycl/range.hpp>
17 
18 #include <cstddef>
19 
20 namespace sycl {
22 namespace detail {
23 class Builder;
24 template <typename TransformedArgType, int Dims, typename KernelType>
25 class RoundedRangeKernel;
26 template <typename TransformedArgType, int Dims, typename KernelType>
27 class RoundedRangeKernelWithKH;
28 
29 namespace reduction {
30 template <int Dims>
32 } // namespace reduction
33 } // namespace detail
34 
39 template <int dimensions = 1, bool with_offset = true> class item {
40 #ifndef __SYCL_DISABLE_ITEM_TO_INT_CONV__
41  /* Helper class for conversion operator. Void type is not suitable. User
42  * cannot even try to get address of the operator __private_class(). User
43  * may try to get an address of operator void() and will get the
44  * compile-time error */
45  class __private_class;
46 
47  template <bool B, typename T>
49 #endif // __SYCL_DISABLE_ITEM_TO_INT_CONV__
50 public:
51  item() = delete;
52 
53  id<dimensions> get_id() const { return MImpl.MIndex; }
54 
55  size_t __SYCL_ALWAYS_INLINE get_id(int dimension) const {
56  size_t Id = MImpl.MIndex[dimension];
58  return Id;
59  }
60 
61  size_t __SYCL_ALWAYS_INLINE operator[](int dimension) const {
62  size_t Id = MImpl.MIndex[dimension];
64  return Id;
65  }
66 
67  range<dimensions> get_range() const { return MImpl.MExtent; }
68 
69  size_t __SYCL_ALWAYS_INLINE get_range(int dimension) const {
70  size_t Id = MImpl.MExtent[dimension];
72  return Id;
73  }
74 #ifndef __SYCL_DISABLE_ITEM_TO_INT_CONV__
75  operator EnableIfT<dimensions == 1, std::size_t>() const { return get_id(0); }
76 #endif // __SYCL_DISABLE_ITEM_TO_INT_CONV__
77  template <bool has_offset = with_offset>
78  __SYCL2020_DEPRECATED("offsets are deprecated in SYCL2020")
79  detail::enable_if_t<has_offset, id<dimensions>> get_offset() const {
80  return MImpl.MOffset;
81  }
82 
83  template <bool has_offset = with_offset>
84  __SYCL2020_DEPRECATED("offsets are deprecated in SYCL2020")
85  detail::enable_if_t<has_offset, size_t> __SYCL_ALWAYS_INLINE
86  get_offset(int dimension) const {
87  size_t Id = MImpl.MOffset[dimension];
89  return Id;
90  }
91 
92  template <bool has_offset = with_offset>
94  return detail::Builder::createItem<dimensions, true>(
95  MImpl.MExtent, MImpl.MIndex, /*Offset*/ {});
96  }
97 
99  size_t Id = MImpl.get_linear_id();
100  __SYCL_ASSUME_INT(Id);
101  return Id;
102  }
103 
104  item(const item &rhs) = default;
105 
107 
108  item &operator=(const item &rhs) = default;
109 
110  item &operator=(item &&rhs) = default;
111 
112  bool operator==(const item &rhs) const { return rhs.MImpl == MImpl; }
113 
114  bool operator!=(const item &rhs) const { return rhs.MImpl != MImpl; }
115 
116 protected:
117  template <bool has_offset = with_offset>
118  item(detail::enable_if_t<has_offset, const range<dimensions>> &extent,
119  const id<dimensions> &index, const id<dimensions> &offset)
120  : MImpl{extent, index, offset} {}
121 
122  template <bool has_offset = with_offset>
123  item(detail::enable_if_t<!has_offset, const range<dimensions>> &extent,
124  const id<dimensions> &index)
125  : MImpl{extent, index} {}
126 
127  friend class detail::Builder;
128 
129 private:
130  // Friend to get access to private method set_allowed_range().
131  template <typename, int, typename> friend class detail::RoundedRangeKernel;
132  template <typename, int, typename>
134  void set_allowed_range(const range<dimensions> rnwi) { MImpl.MExtent = rnwi; }
135 
136  template <int Dims>
137  friend item<Dims, false>
139 
141 };
142 
143 template <int Dims>
144 __SYCL_DEPRECATED("use sycl::ext::oneapi::experimental::this_item() instead")
145 item<Dims> this_item() {
146 #ifdef __SYCL_DEVICE_ONLY__
147  return detail::Builder::getElement(detail::declptr<item<Dims>>());
148 #else
149  throw sycl::exception(
150  sycl::make_error_code(sycl::errc::feature_not_supported),
151  "Free function calls are not supported on host device");
152 #endif
153 }
154 
155 namespace ext {
156 namespace oneapi {
157 namespace experimental {
158 template <int Dims> item<Dims> this_item() {
159 #ifdef __SYCL_DEVICE_ONLY__
160  return sycl::detail::Builder::getElement(sycl::detail::declptr<item<Dims>>());
161 #else
162  throw sycl::exception(
163  sycl::make_error_code(sycl::errc::feature_not_supported),
164  "Free function calls are not supported on host device");
165 #endif
166 }
167 } // namespace experimental
168 } // namespace oneapi
169 } // namespace ext
170 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
171 } // namespace sycl
Identifies an instance of the function object executing at each point in a range.
Definition: item.hpp:39
size_t __SYCL_ALWAYS_INLINE get_id(int dimension) const
Definition: item.hpp:55
item(detail::enable_if_t< has_offset, const range< dimensions >> &extent, const id< dimensions > &index, const id< dimensions > &offset)
Definition: item.hpp:118
size_t __SYCL_ALWAYS_INLINE get_linear_id() const
Definition: item.hpp:98
size_t __SYCL_ALWAYS_INLINE get_range(int dimension) const
Definition: item.hpp:69
bool operator!=(const item &rhs) const
Definition: item.hpp:114
item & operator=(const item &rhs)=default
size_t __SYCL_ALWAYS_INLINE operator[](int dimension) const
Definition: item.hpp:61
range< dimensions > get_range() const
Definition: item.hpp:67
item(const item &rhs)=default
item & operator=(item &&rhs)=default
id< dimensions > get_id() const
Definition: item.hpp:53
bool operator==(const item &rhs) const
Definition: item.hpp:112
item(detail::enable_if_t<!has_offset, const range< dimensions >> &extent, const id< dimensions > &index)
Definition: item.hpp:123
item(item< dimensions, with_offset > &&rhs)=default
#define __SYCL_ASSUME_INT(x)
Definition: defines.hpp:18
#define __SYCL_INLINE_VER_NAMESPACE(X)
#define __SYCL_DEPRECATED(message)
#define __SYCL2020_DEPRECATED(message)
#define __SYCL_ALWAYS_INLINE
item< Dims, false > getDelinearizedItem(range< Dims > Range, id< Dims > Id)
typename std::conditional< B, T, F >::type conditional_t
typename std::enable_if< B, T >::type enable_if_t
std::error_code make_error_code(sycl::errc E) noexcept
Constructs an error code using e and sycl_category()
Definition: exception.cpp:91
auto reduction(buffer< T, 1, AllocatorT > Var, handler &CGH, BinaryOperation, const property_list &PropList={})
Constructs a reduction object using the given buffer Var, handler CGH, reduction operation Combiner,...
Definition: reduction.hpp:2420
---— Error handling, matching OpenCL plugin semantics.
Definition: access.hpp:14