Skip to content

Commit 82cb6d8

Browse files
steffenlarsensteffenlarsen
authored
[SYCL][CUDA] Fix uses of PI descriptors after renaming (#1310)
The CUDA backend cannot currently build due to the renaming of descriptors done in #1239. This PR adjusts the backend to the new descriptor names. Signed-off-by: Steffen Larsen <[email protected]>
1 parent 656e769 commit 82cb6d8

File tree

3 files changed

+33
-31
lines changed

3 files changed

+33
-31
lines changed

‎sycl/plugins/cuda/pi_cuda.cpp

Lines changed: 30 additions & 28 deletions
Original file line numberDiff line numberDiff line change
@@ -713,6 +713,8 @@ pi_result cuda_piContextGetInfo(pi_context context, pi_context_info param_name,
713713
case PI_CONTEXT_INFO_REFERENCE_COUNT:
714714
return getInfo(param_value_size, param_value, param_value_size_ret,
715715
context->get_reference_count());
716+
default:
717+
PI_HANDLE_UNKNOWN_PARAM_NAME(param_name);
716718
}
717719

718720
return PI_OUT_OF_RESOURCES;
@@ -1032,7 +1034,7 @@ pi_result cuda_piDeviceGetInfo(pi_device device, pi_device_info param_name,
10321034
}
10331035
case PI_DEVICE_INFO_LOCAL_MEM_TYPE: {
10341036
return getInfo(param_value_size, param_value, param_value_size_ret,
1035-
PI_LOCAL_MEM_TYPE_LOCAL);
1037+
PI_DEVICE_LOCAL_MEM_TYPE_LOCAL);
10361038
}
10371039
case PI_DEVICE_INFO_LOCAL_MEM_SIZE: {
10381040
// OpenCL's "local memory" maps most closely to CUDA's "shared memory".
@@ -1073,16 +1075,16 @@ pi_result cuda_piDeviceGetInfo(pi_device device, pi_device_info param_name,
10731075
return getInfo(param_value_size, param_value, param_value_size_ret,
10741076
size_t{1000u});
10751077
}
1076-
case PI_DEVICE_INFO_IS_ENDIAN_LITTLE: {
1078+
case PI_DEVICE_INFO_ENDIAN_LITTLE: {
10771079
return getInfo(param_value_size, param_value, param_value_size_ret, true);
10781080
}
1079-
case PI_DEVICE_INFO_IS_AVAILABLE: {
1081+
case PI_DEVICE_INFO_AVAILABLE: {
10801082
return getInfo(param_value_size, param_value, param_value_size_ret, true);
10811083
}
1082-
case PI_DEVICE_INFO_IS_COMPILER_AVAILABLE: {
1084+
case PI_DEVICE_INFO_COMPILER_AVAILABLE: {
10831085
return getInfo(param_value_size, param_value, param_value_size_ret, true);
10841086
}
1085-
case PI_DEVICE_INFO_IS_LINKER_AVAILABLE: {
1087+
case PI_DEVICE_INFO_LINKER_AVAILABLE: {
10861088
return getInfo(param_value_size, param_value, param_value_size_ret, true);
10871089
}
10881090
case PI_DEVICE_INFO_EXECUTION_CAPABILITIES: {
@@ -1658,8 +1660,8 @@ pi_result cuda_piEnqueueMemBufferWrite(pi_queue command_queue, pi_mem buffer,
16581660
event_wait_list, nullptr);
16591661

16601662
if (event) {
1661-
retImplEv = std::unique_ptr<_pi_event>(
1662-
_pi_event::make_native(PI_COMMAND_MEMBUFFER_WRITE, command_queue));
1663+
retImplEv = std::unique_ptr<_pi_event>(_pi_event::make_native(
1664+
PI_COMMAND_TYPE_MEM_BUFFER_WRITE, command_queue));
16631665
retImplEv->start();
16641666
}
16651667

@@ -1703,8 +1705,8 @@ pi_result cuda_piEnqueueMemBufferRead(pi_queue command_queue, pi_mem buffer,
17031705
event_wait_list, nullptr);
17041706

17051707
if (retEvent) {
1706-
retImplEv = std::unique_ptr<_pi_event>(
1707-
_pi_event::make_native(PI_COMMAND_MEMBUFFER_READ, command_queue));
1708+
retImplEv = std::unique_ptr<_pi_event>(_pi_event::make_native(
1709+
PI_COMMAND_TYPE_MEM_BUFFER_READ, command_queue));
17081710
retImplEv->start();
17091711
}
17101712

@@ -1943,8 +1945,8 @@ pi_result cuda_piEnqueueKernelLaunch(
19431945
auto argIndices = kernel->get_arg_indices();
19441946

19451947
if (event) {
1946-
retImplEv = std::unique_ptr<_pi_event>(
1947-
_pi_event::make_native(PI_COMMAND_KERNEL_LAUNCH, command_queue));
1948+
retImplEv = std::unique_ptr<_pi_event>(_pi_event::make_native(
1949+
PI_COMMAND_TYPE_NDRANGE_KERNEL, command_queue));
19481950
retImplEv->start();
19491951
}
19501952

@@ -2247,15 +2249,15 @@ pi_result cuda_piKernelGetGroupInfo(pi_kernel kernel, pi_device device,
22472249
if (kernel != nullptr) {
22482250

22492251
switch (param_name) {
2250-
case PI_KERNEL_GROUP_INFO_SIZE: {
2252+
case PI_KERNEL_GROUP_INFO_WORK_GROUP_SIZE: {
22512253
int max_threads = 0;
22522254
cl::sycl::detail::pi::assertion(cuFuncGetAttribute(&max_threads,
22532255
CU_FUNC_ATTRIBUTE_MAX_THREADS_PER_BLOCK,
22542256
kernel->get()) == CUDA_SUCCESS);
22552257
return getInfo(param_value_size, param_value, param_value_size_ret,
22562258
size_t(max_threads));
22572259
}
2258-
case PI_KERNEL_COMPILE_GROUP_INFO_SIZE: {
2260+
case PI_KERNEL_GROUP_INFO_COMPILE_WORK_GROUP_SIZE: {
22592261
// Returns the work-group size specified in the kernel source or IL.
22602262
// If the work-group size is not specified in the kernel source or IL,
22612263
// (0, 0, 0) is returned.
@@ -2266,7 +2268,7 @@ pi_result cuda_piKernelGetGroupInfo(pi_kernel kernel, pi_device device,
22662268
return getInfoArray(3, param_value_size, param_value,
22672269
param_value_size_ret, group_size);
22682270
}
2269-
case PI_KERNEL_LOCAL_MEM_SIZE: {
2271+
case PI_KERNEL_GROUP_INFO_LOCAL_MEM_SIZE: {
22702272
// OpenCL LOCAL == CUDA SHARED
22712273
int bytes = 0;
22722274
cl::sycl::detail::pi::assertion(cuFuncGetAttribute(&bytes,
@@ -2275,7 +2277,7 @@ pi_result cuda_piKernelGetGroupInfo(pi_kernel kernel, pi_device device,
22752277
return getInfo(param_value_size, param_value, param_value_size_ret,
22762278
pi_uint64(bytes));
22772279
}
2278-
case PI_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE: {
2280+
case PI_KERNEL_GROUP_INFO_PREFERRED_WORK_GROUP_SIZE_MULTIPLE: {
22792281
// Work groups should be multiples of the warp size
22802282
int warpSize = 0;
22812283
cl::sycl::detail::pi::assertion(
@@ -2284,7 +2286,7 @@ pi_result cuda_piKernelGetGroupInfo(pi_kernel kernel, pi_device device,
22842286
return getInfo(param_value_size, param_value, param_value_size_ret,
22852287
static_cast<size_t>(warpSize));
22862288
}
2287-
case PI_KERNEL_PRIVATE_MEM_SIZE: {
2289+
case PI_KERNEL_GROUP_INFO_PRIVATE_MEM_SIZE: {
22882290
// OpenCL PRIVATE == CUDA LOCAL
22892291
int bytes = 0;
22902292
cl::sycl::detail::pi::assertion(
@@ -2373,7 +2375,7 @@ pi_result cuda_piEventGetInfo(pi_event event, pi_event_info param_name,
23732375
assert(event != nullptr);
23742376

23752377
switch (param_name) {
2376-
case PI_EVENT_INFO_QUEUE:
2378+
case PI_EVENT_INFO_COMMAND_QUEUE:
23772379
return getInfo<pi_queue>(param_value_size, param_value,
23782380
param_value_size_ret, event->get_queue());
23792381
case PI_EVENT_INFO_COMMAND_TYPE:
@@ -2554,7 +2556,7 @@ pi_result cuda_piEnqueueEventsWait(pi_queue command_queue,
25542556

25552557
if (event) {
25562558
auto new_event =
2557-
_pi_event::make_native(PI_COMMAND_EVENTS_WAIT, command_queue);
2559+
_pi_event::make_native(PI_COMMAND_TYPE_MARKER, command_queue);
25582560
new_event->start();
25592561
new_event->record();
25602562
*event = new_event;
@@ -2648,8 +2650,8 @@ pi_result cuda_piEnqueueMemBufferReadRect(
26482650
event_wait_list, nullptr);
26492651

26502652
if (retEvent) {
2651-
retImplEv = std::unique_ptr<_pi_event>(
2652-
_pi_event::make_native(PI_COMMAND_MEMBUFFER_READ, command_queue));
2653+
retImplEv = std::unique_ptr<_pi_event>(_pi_event::make_native(
2654+
PI_COMMAND_TYPE_MEM_BUFFER_READ, command_queue));
26532655
retImplEv->start();
26542656
}
26552657

@@ -2699,8 +2701,8 @@ pi_result cuda_piEnqueueMemBufferWriteRect(
26992701
event_wait_list, nullptr);
27002702

27012703
if (retEvent) {
2702-
retImplEv = std::unique_ptr<_pi_event>(
2703-
_pi_event::make_native(PI_COMMAND_MEMBUFFER_WRITE, command_queue));
2704+
retImplEv = std::unique_ptr<_pi_event>(_pi_event::make_native(
2705+
PI_COMMAND_TYPE_MEM_BUFFER_WRITE, command_queue));
27042706
retImplEv->start();
27052707
}
27062708

@@ -2754,8 +2756,8 @@ pi_result cuda_piEnqueueMemBufferCopy(pi_queue command_queue, pi_mem src_buffer,
27542756
result = PI_CHECK_ERROR(cuMemcpyDtoDAsync(dst, src, size, stream));
27552757

27562758
if (event) {
2757-
auto new_event =
2758-
_pi_event::make_native(PI_COMMAND_MEMBUFFER_COPY, command_queue);
2759+
auto new_event = _pi_event::make_native(PI_COMMAND_TYPE_MEM_BUFFER_COPY,
2760+
command_queue);
27592761
new_event->record();
27602762
*event = new_event;
27612763
}
@@ -2792,8 +2794,8 @@ pi_result cuda_piEnqueueMemBufferCopyRect(
27922794
event_wait_list, nullptr);
27932795

27942796
if (event) {
2795-
retImplEv = std::unique_ptr<_pi_event>(
2796-
_pi_event::make_native(PI_COMMAND_MEMBUFFER_COPY, command_queue));
2797+
retImplEv = std::unique_ptr<_pi_event>(_pi_event::make_native(
2798+
PI_COMMAND_TYPE_MEM_BUFFER_COPY, command_queue));
27972799
retImplEv->start();
27982800
}
27992801

@@ -2896,8 +2898,8 @@ pi_result cuda_piEnqueueMemBufferFill(pi_queue command_queue, pi_mem buffer,
28962898
}
28972899

28982900
if (event) {
2899-
auto new_event =
2900-
_pi_event::make_native(PI_COMMAND_MEMBUFFER_FILL, command_queue);
2901+
auto new_event = _pi_event::make_native(PI_COMMAND_TYPE_MEM_BUFFER_FILL,
2902+
command_queue);
29012903
new_event->record();
29022904
*event = new_event;
29032905
}

‎sycl/plugins/cuda/pi_cuda.hpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -273,7 +273,7 @@ class _pi_event {
273273
pi_context get_context() const noexcept { return context_; };
274274

275275
bool is_user_event() const noexcept {
276-
return get_command_type() == PI_COMMAND_USER;
276+
return get_command_type() == PI_COMMAND_TYPE_USER;
277277
}
278278

279279
bool is_native_event() const noexcept { return !is_user_event(); }
@@ -297,7 +297,7 @@ class _pi_event {
297297
// make a user event. CUDA has no concept of user events, so this
298298
// functionality is implemented by the CUDA PI implementation.
299299
static pi_event make_user(pi_context context) {
300-
return new _pi_event(PI_COMMAND_USER, context, nullptr);
300+
return new _pi_event(PI_COMMAND_TYPE_USER, context, nullptr);
301301
}
302302

303303
// construct a native CUDA. This maps closely to the underlying CUDA event.

‎sycl/unittests/pi/cuda/test_events.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -83,7 +83,7 @@ TEST_F(DISABLED_CudaEventTests, PICreateEvent) {
8383

8484
TEST_F(DISABLED_CudaEventTests, piGetInfoNativeEvent) {
8585

86-
auto foo = _pi_event::make_native(PI_COMMAND_KERNEL_LAUNCH, _queue);
86+
auto foo = _pi_event::make_native(PI_COMMAND_TYPE_NDRANGE_KERNEL, _queue);
8787
ASSERT_NE(foo, nullptr);
8888

8989
pi_event_status paramValue = {};

0 commit comments

Comments
 (0)