17 #include <type_traits>
20 inline namespace _V1 {
23 namespace experimental {
41 static_assert(count < 2,
"Duplicate cache_level L1 specification");
44 static_assert(count < 2,
"Duplicate cache_level L2 specification");
47 static_assert(count < 2,
"Duplicate cache_level L3 specification");
50 static_assert(count < 2,
"Duplicate cache_level L4 specification");
56 static constexpr
const auto mode = M;
57 static constexpr
const int countL1 = ((Ls == cache_level::L1 ? 1 : 0) + ...);
58 static constexpr
const int countL2 = ((Ls == cache_level::L2 ? 1 : 0) + ...);
59 static constexpr
const int countL3 = ((Ls == cache_level::L3 ? 1 : 0) + ...);
60 static constexpr
const int countL4 = ((Ls == cache_level::L4 ? 1 : 0) + ...);
61 static constexpr
const uint32_t
levels = ((1 <<
static_cast<int>(Ls)) | ...);
67 detail::checkLevel2<countL2>(), detail::checkLevel3<countL3>(),
68 detail::checkLevel4<countL4>(), levels << static_cast<int>(M) * 4);
71 template <
typename PropertyT,
typename... Ts>
77 oneapi::experimental::detail::PropKind::CacheControlReadHint> {
78 template <
typename... Cs>
84 oneapi::experimental::detail::PropKind::CacheControlReadAssertion> {
85 template <
typename... Cs>
91 oneapi::experimental::detail::PropKind::CacheControlWrite> {
92 template <
typename... Cs>
96 template <
typename... Cs>
99 template <
typename... Cs>
102 template <
typename... Cs>
109 namespace experimental {
111 template <
typename T,
typename PropertyListT>
class annotated_ptr;
113 template <
typename T,
typename PropertyListT>
117 template <
typename T,
typename PropertyListT>
121 template <
typename T,
typename PropertyListT>
143 static constexpr
int countL(
int levels,
int mask) {
144 return levels & mask ? 1 : 0;
147 template <
int countL1,
int countL2,
int countL3,
int countL4>
149 static_assert(countL1 < 2,
"Conflicting cache_mode at L1");
150 static_assert(countL2 < 2,
"Conflicting cache_mode at L2");
151 static_assert(countL3 < 2,
"Conflicting cache_mode at L3");
152 static_assert(countL4 < 2,
"Conflicting cache_mode at L4");
159 M == cache_mode::uncached || M == cache_mode::cached ||
160 M == cache_mode::streaming,
161 "read_hint must specify cache_mode uncached, cached or streaming");
168 "read_assertion must specify cache_mode invalidate or constant");
173 static_assert(M == cache_mode::uncached || M == cache_mode::write_through ||
174 M == cache_mode::write_back || M == cache_mode::streaming,
175 "write_hint must specify cache_mode uncached, write_through, "
176 "write_back or streaming");
180 template <
typename... Cs>
182 static constexpr
const char *
name =
"sycl-cache-read-hint";
184 ((checkReadHint<Cs::mode>() + ...),
189 ((Cs::encoding) | ...));
192 template <
typename... Cs>
194 intel::experimental::read_assertion_key::value_t<Cs...>> {
195 static constexpr
const char *
name =
"sycl-cache-read-assertion";
197 ((checkReadAssertion<Cs::mode>() + ...),
202 ((Cs::encoding) | ...));
205 template <
typename... Cs>
207 static constexpr
const char *
name =
"sycl-cache-write-hint";
209 ((checkWriteHint<Cs::mode>() + ...),
214 ((Cs::encoding) | ...));
219 template <
typename T,
typename... Cs>
221 : std::bool_constant<std::is_pointer<T>::value> {};
223 template <
typename T,
typename... Cs>
225 T, intel::experimental::read_assertion_key::value_t<Cs...>>
226 : std::bool_constant<std::is_pointer<T>::value> {};
228 template <
typename T,
typename... Cs>
230 : std::bool_constant<std::is_pointer<T>::value> {};
static constexpr void checkLevel4()
static constexpr void checkLevel2()
static constexpr void checkLevel3()
static constexpr void checkLevel1()
sycl::ext::oneapi::experimental::cache_level cache_level
constexpr read_hint_key::value_t< Cs... > read_hint
constexpr write_hint_key::value_t< Cs... > write_hint
constexpr read_assertion_key::value_t< Cs... > read_assertion
static constexpr int checkReadHint()
sycl::ext::intel::experimental::cache_mode cache_mode
static constexpr int L2BIT
static constexpr int L4BIT
static constexpr int checkReadAssertion()
static constexpr int countL(int levels, int mask)
static constexpr int L1BIT
static constexpr int checkWriteHint()
static constexpr void checkUnique()
static constexpr int L3BIT
friend class annotated_ptr
static constexpr const int countL3
static constexpr const int countL4
static constexpr const uint32_t levels
static constexpr const int countL2
static constexpr const int countL1
static constexpr const auto mode
static constexpr const uint32_t encoding