53 "use 'ext_intel_global_device_space' instead") =
56 "use 'ext_intel_global_host_space' instead") =
59 "use 'ext_intel_global_host_space' instead") =
78 #if __cplusplus >= 201703L
81 inline constexpr mode_tag_t<access_mode::read_write> read_write{};
82 inline constexpr mode_tag_t<access_mode::write> write_only{};
83 inline constexpr mode_target_tag_t<access_mode::read, target::constant_buffer>
91 sycl::detail::InlineVariableHelper<mode_tag_t<access_mode::read>>::value;
92 constexpr
const auto &
read_write = sycl::detail::InlineVariableHelper<
93 mode_tag_t<access_mode::read_write>>::value;
94 constexpr
const auto &write_only =
95 sycl::detail::InlineVariableHelper<mode_tag_t<access_mode::write>>::value;
96 constexpr
const auto &read_constant = sycl::detail::InlineVariableHelper<
97 mode_target_tag_t<access_mode::read, target::constant_buffer>>::value;
106 return T == access::target::host_buffer || T == access::target::host_image;
111 m == access::mode::read_write || m == access::mode::atomic;
115 return m != access::mode::read;
118 #ifdef __SYCL_DEVICE_ONLY__
119 #define __OPENCL_GLOBAL_AS__ __attribute__((opencl_global))
120 #ifdef __ENABLE_USM_ADDR_SPACE__
121 #define __OPENCL_GLOBAL_DEVICE_AS__ __attribute__((opencl_global_device))
122 #define __OPENCL_GLOBAL_HOST_AS__ __attribute__((opencl_global_host))
124 #define __OPENCL_GLOBAL_DEVICE_AS__ __attribute__((opencl_global))
125 #define __OPENCL_GLOBAL_HOST_AS__ __attribute__((opencl_global))
126 #endif // __ENABLE_USM_ADDR_SPACE__
127 #define __OPENCL_LOCAL_AS__ __attribute__((opencl_local))
128 #define __OPENCL_CONSTANT_AS__ __attribute__((opencl_constant))
129 #define __OPENCL_PRIVATE_AS__ __attribute__((opencl_private))
131 #define __OPENCL_GLOBAL_AS__
132 #define __OPENCL_GLOBAL_DEVICE_AS__
133 #define __OPENCL_GLOBAL_HOST_AS__
134 #define __OPENCL_LOCAL_AS__
135 #define __OPENCL_CONSTANT_AS__
136 #define __OPENCL_PRIVATE_AS__
141 access::address_space::global_space;
144 #ifdef __ENABLE_USM_ADDR_SPACE__
147 access::address_space::ext_intel_global_device_space;
149 #endif // __ENABLE_USM_ADDR_SPACE__
153 access::address_space::local_space;
156 template <>
struct TargetToAS<access::target::constant_buffer> {
158 access::address_space::constant_space;
161 template <
typename ElementType, access::address_space addressSpace>
164 template <
typename ElementType>
169 template <
typename ElementType>
174 template <
typename ElementType>
179 template <
typename ElementType>
181 access::address_space::ext_intel_global_device_space> {
185 template <
typename ElementType>
187 access::address_space::ext_intel_global_host_space> {
191 template <
typename ElementType>
198 #if defined(RESTRICT_WRITE_ACCESS_TO_CONSTANT_PTR)
205 template <
typename ElementType>
211 #ifdef __SYCL_DEVICE_ONLY__
212 template <
class T>
struct deduce_AS {
214 "Only types with address space attributes are supported");
219 #ifdef __ENABLE_USM_ADDR_SPACE__
230 access::address_space::ext_intel_global_device_space;
235 access::address_space::ext_intel_global_host_space;
237 #endif // __ENABLE_USM_ADDR_SPACE__
251 access::address_space::global_space;
256 access::address_space::private_space;
265 access::address_space::constant_space;
269 #undef __OPENCL_GLOBAL_AS__
270 #undef __OPENCL_GLOBAL_DEVICE_AS__
271 #undef __OPENCL_GLOBAL_HOST_AS__
272 #undef __OPENCL_LOCAL_AS__
273 #undef __OPENCL_CONSTANT_AS__
274 #undef __OPENCL_PRIVATE_AS__