Skip to content

Commit 7a2f44b

Browse files
steffenlarsensteffenlarsen
authored
[SYCL] Fix naming and use of deprecated address spaces (#6289)
This commit makes the following changes. * Adds ext_intel_global_host_space in accordance with the sycl_ext_intel_usm_address_spaces extensions. * Deprecates ext_intel_host_device_space as it was misspelled and has been replaced by ext_intel_global_host_space. * Replaces uses of deprecated address spaces with ext_intel_global_device_space and ext_intel_global_host_space Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>
1 parent 9397cbc commit 7a2f44b

File tree

9 files changed

+92
-74
lines changed

9 files changed

+92
-74
lines changed

‎sycl/include/CL/sycl/access/access.hpp

Lines changed: 13 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -48,13 +48,16 @@ enum class address_space : int {
4848
constant_space = 2,
4949
local_space = 3,
5050
ext_intel_global_device_space = 4,
51-
ext_intel_host_device_space = 5,
51+
ext_intel_global_host_space = 5,
5252
global_device_space __SYCL2020_DEPRECATED(
5353
"use 'ext_intel_global_device_space' instead") =
5454
ext_intel_global_device_space,
5555
global_host_space __SYCL2020_DEPRECATED(
56-
"use 'ext_intel_host_device_space' instead") =
57-
ext_intel_host_device_space,
56+
"use 'ext_intel_global_host_space' instead") =
57+
ext_intel_global_host_space,
58+
ext_intel_host_device_space __SYCL2020_DEPRECATED(
59+
"use 'ext_intel_global_host_space' instead") =
60+
ext_intel_global_host_space,
5861
generic_space = 6, // TODO generic_space address space is not supported yet
5962
};
6063

@@ -141,7 +144,7 @@ template <access::target accessTarget> struct TargetToAS {
141144
#ifdef __ENABLE_USM_ADDR_SPACE__
142145
template <> struct TargetToAS<access::target::device> {
143146
constexpr static access::address_space AS =
144-
access::address_space::global_device_space;
147+
access::address_space::ext_intel_global_device_space;
145148
};
146149
#endif // __ENABLE_USM_ADDR_SPACE__
147150

@@ -174,12 +177,14 @@ struct DecoratedType<ElementType, access::address_space::global_space> {
174177
};
175178

176179
template <typename ElementType>
177-
struct DecoratedType<ElementType, access::address_space::global_device_space> {
180+
struct DecoratedType<ElementType,
181+
access::address_space::ext_intel_global_device_space> {
178182
using type = __OPENCL_GLOBAL_DEVICE_AS__ ElementType;
179183
};
180184

181185
template <typename ElementType>
182-
struct DecoratedType<ElementType, access::address_space::global_host_space> {
186+
struct DecoratedType<ElementType,
187+
access::address_space::ext_intel_global_host_space> {
183188
using type = __OPENCL_GLOBAL_HOST_AS__ ElementType;
184189
};
185190

@@ -222,12 +227,12 @@ template <class T> struct remove_AS<__OPENCL_GLOBAL_HOST_AS__ T> {
222227

223228
template <class T> struct deduce_AS<__OPENCL_GLOBAL_DEVICE_AS__ T> {
224229
static const access::address_space value =
225-
access::address_space::global_device_space;
230+
access::address_space::ext_intel_global_device_space;
226231
};
227232

228233
template <class T> struct deduce_AS<__OPENCL_GLOBAL_HOST_AS__ T> {
229234
static const access::address_space value =
230-
access::address_space::global_host_space;
235+
access::address_space::ext_intel_global_host_space;
231236
};
232237
#endif // __ENABLE_USM_ADDR_SPACE__
233238

‎sycl/include/CL/sycl/atomic.hpp

Lines changed: 9 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -48,7 +48,7 @@ template <cl::sycl::access::address_space AS> struct IsValidAtomicAddressSpace {
4848
static constexpr bool value =
4949
(AS == access::address_space::global_space ||
5050
AS == access::address_space::local_space ||
51-
AS == access::address_space::global_device_space);
51+
AS == access::address_space::ext_intel_global_device_space);
5252
};
5353

5454
// Type trait to translate a cl::sycl::access::address_space to
@@ -58,7 +58,8 @@ template <> struct GetSpirvMemoryScope<access::address_space::global_space> {
5858
static constexpr auto scope = __spv::Scope::Device;
5959
};
6060
template <>
61-
struct GetSpirvMemoryScope<access::address_space::global_device_space> {
61+
struct GetSpirvMemoryScope<
62+
access::address_space::ext_intel_global_device_space> {
6263
static constexpr auto scope = __spv::Scope::Device;
6364
};
6465
template <> struct GetSpirvMemoryScope<access::address_space::local_space> {
@@ -177,7 +178,7 @@ class __SYCL2020_DEPRECATED(
177178
"long long, float");
178179
static_assert(detail::IsValidAtomicAddressSpace<addressSpace>::value,
179180
"Invalid SYCL atomic address_space. Valid address spaces are: "
180-
"global_space, local_space, global_device_space");
181+
"global_space, local_space, ext_intel_global_device_space");
181182
static constexpr auto SpirvScope =
182183
detail::GetSpirvMemoryScope<addressSpace>::scope;
183184

@@ -196,20 +197,22 @@ class __SYCL2020_DEPRECATED(
196197
}
197198

198199
#ifdef __ENABLE_USM_ADDR_SPACE__
199-
// Create atomic in global_space with one from global_device_space
200+
// Create atomic in global_space with one from ext_intel_global_device_space
200201
template <access::address_space _Space = addressSpace,
201202
typename = typename detail::enable_if_t<
202203
_Space == addressSpace &&
203204
addressSpace == access::address_space::global_space>>
204-
atomic(const atomic<T, access::address_space::global_device_space> &RHS) {
205+
atomic(const atomic<T, access::address_space::ext_intel_global_device_space>
206+
&RHS) {
205207
Ptr = RHS.Ptr;
206208
}
207209

208210
template <access::address_space _Space = addressSpace,
209211
typename = typename detail::enable_if_t<
210212
_Space == addressSpace &&
211213
addressSpace == access::address_space::global_space>>
212-
atomic(atomic<T, access::address_space::global_device_space> &&RHS) {
214+
atomic(
215+
atomic<T, access::address_space::ext_intel_global_device_space> &&RHS) {
213216
Ptr = RHS.Ptr;
214217
}
215218
#endif // __ENABLE_USM_ADDR_SPACE__

‎sycl/include/CL/sycl/atomic_ref.hpp

Lines changed: 5 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -44,7 +44,7 @@ struct IsValidAtomicRefAddressSpace {
4444
static constexpr bool value =
4545
(AS == access::address_space::global_space ||
4646
AS == access::address_space::local_space ||
47-
AS == access::address_space::global_device_space ||
47+
AS == access::address_space::ext_intel_global_device_space ||
4848
AS == access::address_space::generic_space);
4949
};
5050

@@ -119,10 +119,10 @@ class atomic_ref_base {
119119
"Invalid atomic type. Valid types are int, unsigned int, long, "
120120
"unsigned long, long long, unsigned long long, float, double "
121121
"and pointer types");
122-
static_assert(
123-
detail::IsValidAtomicRefAddressSpace<AddressSpace>::value,
124-
"Invalid atomic address_space. Valid address spaces are: "
125-
"global_space, local_space, global_device_space, generic_space");
122+
static_assert(detail::IsValidAtomicRefAddressSpace<AddressSpace>::value,
123+
"Invalid atomic address_space. Valid address spaces are: "
124+
"global_space, local_space, ext_intel_global_device_space, "
125+
"generic_space");
126126
static_assert(
127127
detail::IsValidDefaultOrder<DefaultOrder>::value,
128128
"Invalid default memory_order for atomics. Valid defaults are: "

‎sycl/include/CL/sycl/detail/generic_type_lists.hpp

Lines changed: 6 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -539,22 +539,22 @@ namespace gvl {
539539
using all_address_space_list = address_space_list<
540540
access::address_space::local_space, access::address_space::global_space,
541541
access::address_space::private_space, access::address_space::constant_space,
542-
access::address_space::global_device_space,
543-
access::address_space::global_host_space>;
542+
access::address_space::ext_intel_global_device_space,
543+
access::address_space::ext_intel_global_host_space>;
544544

545545
using nonconst_address_space_list =
546546
address_space_list<access::address_space::local_space,
547547
access::address_space::global_space,
548548
access::address_space::private_space,
549-
access::address_space::global_device_space,
550-
access::address_space::global_host_space>;
549+
access::address_space::ext_intel_global_device_space,
550+
access::address_space::ext_intel_global_host_space>;
551551

552552
using nonlocal_address_space_list =
553553
address_space_list<access::address_space::global_space,
554554
access::address_space::private_space,
555555
access::address_space::constant_space,
556-
access::address_space::global_device_space,
557-
access::address_space::global_host_space>;
556+
access::address_space::ext_intel_global_device_space,
557+
access::address_space::ext_intel_global_host_space>;
558558
} // namespace gvl
559559
} // namespace detail
560560
} // namespace sycl

‎sycl/include/CL/sycl/multi_ptr.hpp

Lines changed: 42 additions & 35 deletions
Original file line numberDiff line numberDiff line change
@@ -123,15 +123,16 @@ template <typename ElementType, access::address_space Space> class multi_ptr {
123123
return reinterpret_cast<ReturnPtr>(m_Pointer)[index];
124124
}
125125

126-
// Only if Space == global_space || global_device_space || generic_space
127-
template <int dimensions, access::mode Mode,
128-
access::placeholder isPlaceholder, typename PropertyListT,
129-
access::address_space _Space = Space,
130-
typename = typename detail::enable_if_t<
131-
_Space == Space &&
132-
(Space == access::address_space::generic_space ||
133-
Space == access::address_space::global_space ||
134-
Space == access::address_space::global_device_space)>>
126+
// Only if Space is in
127+
// {global_space, ext_intel_global_device_space, generic_space}
128+
template <
129+
int dimensions, access::mode Mode, access::placeholder isPlaceholder,
130+
typename PropertyListT, access::address_space _Space = Space,
131+
typename = typename detail::enable_if_t<
132+
_Space == Space &&
133+
(Space == access::address_space::generic_space ||
134+
Space == access::address_space::global_space ||
135+
Space == access::address_space::ext_intel_global_device_space)>>
135136
multi_ptr(accessor<ElementType, dimensions, Mode, access::target::device,
136137
isPlaceholder, PropertyListT>
137138
Accessor) {
@@ -170,8 +171,9 @@ template <typename ElementType, access::address_space Space> class multi_ptr {
170171
// 2. from multi_ptr<ElementType, Space> to multi_ptr<const ElementType,
171172
// Space>
172173

173-
// Only if Space == global_space || global_device_space || generic_space and
174-
// element type is const
174+
// Only if Space is in
175+
// {global_space, ext_intel_global_device_space, generic_space} and element
176+
// type is const
175177
template <
176178
int dimensions, access::mode Mode, access::placeholder isPlaceholder,
177179
typename PropertyListT, access::address_space _Space = Space,
@@ -180,7 +182,7 @@ template <typename ElementType, access::address_space Space> class multi_ptr {
180182
_Space == Space &&
181183
(Space == access::address_space::generic_space ||
182184
Space == access::address_space::global_space ||
183-
Space == access::address_space::global_device_space) &&
185+
Space == access::address_space::ext_intel_global_device_space) &&
184186
std::is_const<ET>::value && std::is_same<ET, ElementType>::value>>
185187
multi_ptr(accessor<typename detail::remove_const_t<ET>, dimensions, Mode,
186188
access::target::device, isPlaceholder, PropertyListT>
@@ -302,13 +304,14 @@ template <typename ElementType, access::address_space Space> class multi_ptr {
302304

303305
#ifdef __ENABLE_USM_ADDR_SPACE__
304306
// Explicit conversion to global_space
305-
// Only available if Space == address_space::global_device_space ||
306-
// Space == address_space::global_host_space
307-
template <access::address_space _Space = Space,
308-
typename = typename detail::enable_if_t<
309-
_Space == Space &&
310-
(Space == access::address_space::global_device_space ||
311-
Space == access::address_space::global_host_space)>>
307+
// Only available if Space == address_space::ext_intel_global_device_space ||
308+
// Space == address_space::ext_intel_global_host_space
309+
template <
310+
access::address_space _Space = Space,
311+
typename = typename detail::enable_if_t<
312+
_Space == Space &&
313+
(Space == access::address_space::ext_intel_global_device_space ||
314+
Space == access::address_space::ext_intel_global_host_space)>>
312315
explicit
313316
operator multi_ptr<ElementType, access::address_space::global_space>() const {
314317
using global_pointer_t = typename detail::DecoratedType<
@@ -392,14 +395,16 @@ template <access::address_space Space> class multi_ptr<void, Space> {
392395
return *this;
393396
}
394397

395-
// Only if Space == global_space || global_device_space || generic_space
396-
template <typename ElementType, int dimensions, access::mode Mode,
397-
typename PropertyListT, access::address_space _Space = Space,
398-
typename = typename detail::enable_if_t<
399-
_Space == Space &&
400-
(Space == access::address_space::generic_space ||
401-
Space == access::address_space::global_space ||
402-
Space == access::address_space::global_device_space)>>
398+
// Only if Space is in
399+
// {global_space, ext_intel_global_device_space, generic_space}
400+
template <
401+
typename ElementType, int dimensions, access::mode Mode,
402+
typename PropertyListT, access::address_space _Space = Space,
403+
typename = typename detail::enable_if_t<
404+
_Space == Space &&
405+
(Space == access::address_space::generic_space ||
406+
Space == access::address_space::global_space ||
407+
Space == access::address_space::ext_intel_global_device_space)>>
403408
multi_ptr(accessor<ElementType, dimensions, Mode, access::target::device,
404409
access::placeholder::false_t, PropertyListT>
405410
Accessor)
@@ -515,14 +520,16 @@ class multi_ptr<const void, Space> {
515520
return *this;
516521
}
517522

518-
// Only if Space == global_space || global_device_space || generic_space
519-
template <typename ElementType, int dimensions, access::mode Mode,
520-
typename PropertyListT, access::address_space _Space = Space,
521-
typename = typename detail::enable_if_t<
522-
_Space == Space &&
523-
(Space == access::address_space::generic_space ||
524-
Space == access::address_space::global_space ||
525-
Space == access::address_space::global_device_space)>>
523+
// Only if Space is in
524+
// {global_space, ext_intel_global_device_space, generic_space}
525+
template <
526+
typename ElementType, int dimensions, access::mode Mode,
527+
typename PropertyListT, access::address_space _Space = Space,
528+
typename = typename detail::enable_if_t<
529+
_Space == Space &&
530+
(Space == access::address_space::generic_space ||
531+
Space == access::address_space::global_space ||
532+
Space == access::address_space::ext_intel_global_device_space)>>
526533
multi_ptr(accessor<ElementType, dimensions, Mode, access::target::device,
527534
access::placeholder::false_t, PropertyListT>
528535
Accessor)

‎sycl/include/CL/sycl/pointers.hpp

Lines changed: 3 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -25,11 +25,12 @@ using global_ptr = multi_ptr<ElementType, access::address_space::global_space>;
2525

2626
template <typename ElementType>
2727
using device_ptr =
28-
multi_ptr<ElementType, access::address_space::global_device_space>;
28+
multi_ptr<ElementType,
29+
access::address_space::ext_intel_global_device_space>;
2930

3031
template <typename ElementType>
3132
using host_ptr =
32-
multi_ptr<ElementType, access::address_space::global_host_space>;
33+
multi_ptr<ElementType, access::address_space::ext_intel_global_host_space>;
3334

3435
template <typename ElementType>
3536
using local_ptr = multi_ptr<ElementType, access::address_space::local_space>;

‎sycl/include/sycl/ext/intel/experimental/fpga_lsu.hpp

Lines changed: 6 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -158,11 +158,12 @@ template <class... _mem_access_params> class lsu final {
158158
static_assert(_cache_val >= 0, "cache size parameter must be non-negative");
159159

160160
template <access::address_space _space> static void check_space() {
161-
static_assert(_space == access::address_space::global_space ||
162-
_space == access::address_space::global_device_space ||
163-
_space == access::address_space::global_host_space,
164-
"lsu controls are only supported for global_ptr, "
165-
"device_ptr, and host_ptr objects");
161+
static_assert(
162+
_space == access::address_space::global_space ||
163+
_space == access::address_space::ext_intel_global_device_space ||
164+
_space == access::address_space::ext_intel_global_host_space,
165+
"lsu controls are only supported for global_ptr, "
166+
"device_ptr, and host_ptr objects");
166167
}
167168

168169
static void check_load() {

‎sycl/include/sycl/ext/intel/fpga_lsu.hpp

Lines changed: 6 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -98,11 +98,12 @@ template <class... _mem_access_params> class lsu final {
9898
static_assert(_cache_val >= 0, "cache size parameter must be non-negative");
9999

100100
template <access::address_space _space> static void check_space() {
101-
static_assert(_space == access::address_space::global_space ||
102-
_space == access::address_space::global_device_space ||
103-
_space == access::address_space::global_host_space,
104-
"lsu controls are only supported for global_ptr, "
105-
"device_ptr, and host_ptr objects");
101+
static_assert(
102+
_space == access::address_space::global_space ||
103+
_space == access::address_space::ext_intel_global_device_space ||
104+
_space == access::address_space::ext_intel_global_host_space,
105+
"lsu controls are only supported for global_ptr, "
106+
"device_ptr, and host_ptr objects");
106107
}
107108

108109
static void check_load() {

‎sycl/include/sycl/ext/oneapi/atomic_ref.hpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -47,7 +47,7 @@ template <cl::sycl::access::address_space AS>
4747
using IsValidAtomicAddressSpace =
4848
bool_constant<AS == access::address_space::global_space ||
4949
AS == access::address_space::local_space ||
50-
AS == access::address_space::global_device_space>;
50+
AS == access::address_space::ext_intel_global_device_space>;
5151

5252
// DefaultOrder parameter is limited to read-modify-write orders
5353
template <memory_order Order>
@@ -122,7 +122,7 @@ class atomic_ref_base {
122122
"and pointer types");
123123
static_assert(detail::IsValidAtomicAddressSpace<AddressSpace>::value,
124124
"Invalid atomic address_space. Valid address spaces are: "
125-
"global_space, local_space, global_device_space");
125+
"global_space, local_space, ext_intel_global_device_space");
126126
static_assert(
127127
detail::IsValidDefaultOrder<DefaultOrder>::value,
128128
"Invalid default memory_order for atomics. Valid defaults are: "

0 commit comments

Comments
 (0)