Skip to content

Commit c7d3c00

Browse files
jchlandajchlanda
authored
[SYCL][NFC] Make UR_CHECK_ERROR a void return macro (#11100)
`UR_CHECK_ERROR` was designed to return `ur_result_t`, however in practice it was guaranteed to only ever return `UR_RESULT_SUCCESS`, as other paths would either terminate, abort or throw. This in turns leads to poor quality/error prone code, as the codebase was littered with: * statements not checking the return value - depending on the compiler generating a warning, * extra check on the return which was only ever going to be true. Some care was required, as the codebase has a habit of accumulating err codes across branches, so depending on the use case the initial value of `ur_result_t Result`s had to be set accordingly (now that `UR_CHECK_ERROR` does not return).
1 parent 3b0a1db commit c7d3c00

File tree

21 files changed

+247
-279
lines changed

21 files changed

+247
-279
lines changed

‎sycl/plugins/unified_runtime/ur/adapters/cuda/common.cpp

Lines changed: 6 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -33,10 +33,10 @@ ur_result_t mapErrorUR(CUresult Result) {
3333
}
3434
}
3535

36-
ur_result_t checkErrorUR(CUresult Result, const char *Function, int Line,
37-
const char *File) {
36+
void checkErrorUR(CUresult Result, const char *Function, int Line,
37+
const char *File) {
3838
if (Result == CUDA_SUCCESS || Result == CUDA_ERROR_DEINITIALIZED) {
39-
return UR_RESULT_SUCCESS;
39+
return;
4040
}
4141

4242
if (std::getenv("SYCL_PI_SUPPRESS_ERROR_MESSAGE") == nullptr &&
@@ -64,10 +64,10 @@ ur_result_t checkErrorUR(CUresult Result, const char *Function, int Line,
6464
throw mapErrorUR(Result);
6565
}
6666

67-
ur_result_t checkErrorUR(ur_result_t Result, const char *Function, int Line,
68-
const char *File) {
67+
void checkErrorUR(ur_result_t Result, const char *Function, int Line,
68+
const char *File) {
6969
if (Result == UR_RESULT_SUCCESS) {
70-
return UR_RESULT_SUCCESS;
70+
return;
7171
}
7272

7373
if (std::getenv("SYCL_PI_SUPPRESS_ERROR_MESSAGE") == nullptr &&

‎sycl/plugins/unified_runtime/ur/adapters/cuda/common.hpp

Lines changed: 4 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -19,11 +19,11 @@ ur_result_t mapErrorUR(CUresult Result);
1919
/// \return UR_RESULT_SUCCESS if \param Result was CUDA_SUCCESS.
2020
/// \throw ur_result_t exception (integer) if input was not success.
2121
///
22-
ur_result_t checkErrorUR(CUresult Result, const char *Function, int Line,
23-
const char *File);
22+
void checkErrorUR(CUresult Result, const char *Function, int Line,
23+
const char *File);
2424

25-
ur_result_t checkErrorUR(ur_result_t Result, const char *Function, int Line,
26-
const char *File);
25+
void checkErrorUR(ur_result_t Result, const char *Function, int Line,
26+
const char *File);
2727

2828
#define UR_CHECK_ERROR(Result) \
2929
checkErrorUR(Result, __func__, __LINE__, __FILE__)

‎sycl/plugins/unified_runtime/ur/adapters/cuda/enqueue.cpp

Lines changed: 42 additions & 47 deletions
Original file line numberDiff line numberDiff line change
@@ -31,7 +31,8 @@ ur_result_t enqueueEventsWait(ur_queue_handle_t CommandQueue, CUstream Stream,
3131
if (Event->getStream() == Stream) {
3232
return UR_RESULT_SUCCESS;
3333
} else {
34-
return UR_CHECK_ERROR(cuStreamWaitEvent(Stream, Event->get(), 0));
34+
UR_CHECK_ERROR(cuStreamWaitEvent(Stream, Event->get(), 0));
35+
return UR_RESULT_SUCCESS;
3536
}
3637
});
3738
return Result;
@@ -193,8 +194,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueEventsWaitWithBarrier(
193194
const ur_event_handle_t *phEventWaitList, ur_event_handle_t *phEvent) {
194195
// This function makes one stream work on the previous work (or work
195196
// represented by input events) and then all future work waits on that stream.
196-
ur_result_t Result;
197-
198197
try {
199198
ScopedContext Active(hQueue->getContext());
200199
uint32_t StreamToken;
@@ -228,23 +227,21 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueEventsWaitWithBarrier(
228227
Event->getComputeStreamToken())) {
229228
return UR_RESULT_SUCCESS;
230229
} else {
231-
return UR_CHECK_ERROR(
230+
UR_CHECK_ERROR(
232231
cuStreamWaitEvent(CuStream, Event->get(), 0));
232+
return UR_RESULT_SUCCESS;
233233
}
234234
});
235235
}
236236

237-
Result = UR_CHECK_ERROR(cuEventRecord(hQueue->BarrierEvent, CuStream));
237+
UR_CHECK_ERROR(cuEventRecord(hQueue->BarrierEvent, CuStream));
238238
for (unsigned int i = 0; i < hQueue->ComputeAppliedBarrier.size(); i++) {
239239
hQueue->ComputeAppliedBarrier[i] = false;
240240
}
241241
for (unsigned int i = 0; i < hQueue->TransferAppliedBarrier.size(); i++) {
242242
hQueue->TransferAppliedBarrier[i] = false;
243243
}
244244
}
245-
if (Result != UR_RESULT_SUCCESS) {
246-
return Result;
247-
}
248245

249246
if (phEvent) {
250247
*phEvent = ur_event_handle_t_::makeNative(
@@ -430,7 +427,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueKernelLaunch(
430427
Device->getMaxChosenLocalMem()));
431428
}
432429

433-
Result = UR_CHECK_ERROR(cuLaunchKernel(
430+
UR_CHECK_ERROR(cuLaunchKernel(
434431
CuFunc, BlocksPerGrid[0], BlocksPerGrid[1], BlocksPerGrid[2],
435432
ThreadsPerBlock[0], ThreadsPerBlock[1], ThreadsPerBlock[2], LocalSize,
436433
CuStream, const_cast<void **>(ArgIndices.data()), nullptr));
@@ -502,7 +499,9 @@ static ur_result_t commonEnqueueMemBufferCopyRect(
502499
params.dstPitch = dst_row_pitch;
503500
params.dstHeight = dst_slice_pitch / dst_row_pitch;
504501

505-
return UR_CHECK_ERROR(cuMemcpy3DAsync(&params, cu_stream));
502+
UR_CHECK_ERROR(cuMemcpy3DAsync(&params, cu_stream));
503+
504+
return UR_RESULT_SUCCESS;
506505
}
507506

508507
UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferReadRect(
@@ -540,7 +539,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferReadRect(
540539
}
541540

542541
if (blockingRead) {
543-
Result = UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
542+
UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
544543
}
545544

546545
if (phEvent) {
@@ -587,7 +586,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferWriteRect(
587586
}
588587

589588
if (blockingWrite) {
590-
Result = UR_CHECK_ERROR(cuStreamSynchronize(cuStream));
589+
UR_CHECK_ERROR(cuStreamSynchronize(cuStream));
591590
}
592591

593592
if (phEvent) {
@@ -614,7 +613,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferCopy(
614613

615614
try {
616615
ScopedContext Active(hQueue->getContext());
617-
ur_result_t Result;
616+
ur_result_t Result = UR_RESULT_SUCCESS;
618617

619618
auto Stream = hQueue->getNextTransferStream();
620619
Result =
@@ -630,7 +629,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferCopy(
630629
auto Src = hBufferSrc->Mem.BufferMem.get() + srcOffset;
631630
auto Dst = hBufferDst->Mem.BufferMem.get() + dstOffset;
632631

633-
Result = UR_CHECK_ERROR(cuMemcpyDtoDAsync(Dst, Src, size, Stream));
632+
UR_CHECK_ERROR(cuMemcpyDtoDAsync(Dst, Src, size, Stream));
634633

635634
if (phEvent) {
636635
UR_CHECK_ERROR(RetImplEvent->record());
@@ -705,10 +704,7 @@ ur_result_t commonMemSetLargePattern(CUstream Stream, uint32_t PatternSize,
705704

706705
// Get 4-byte chunk of the pattern and call cuMemsetD32Async
707706
auto Value = *(static_cast<const uint32_t *>(pPattern));
708-
auto Result = UR_CHECK_ERROR(cuMemsetD32Async(Ptr, Value, Count32, Stream));
709-
if (Result != UR_RESULT_SUCCESS) {
710-
return Result;
711-
}
707+
UR_CHECK_ERROR(cuMemsetD32Async(Ptr, Value, Count32, Stream));
712708
for (auto step = 4u; step < NumberOfSteps; ++step) {
713709
// take 1 byte of the pattern
714710
Value = *(static_cast<const uint8_t *>(pPattern) + step);
@@ -737,8 +733,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferFill(
737733
ScopedContext Active(hQueue->getContext());
738734

739735
auto Stream = hQueue->getNextTransferStream();
740-
ur_result_t Result;
741-
Result =
736+
ur_result_t Result =
742737
enqueueEventsWait(hQueue, Stream, numEventsInWaitList, phEventWaitList);
743738

744739
if (phEvent) {
@@ -755,17 +750,17 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferFill(
755750
switch (patternSize) {
756751
case 1: {
757752
auto Value = *static_cast<const uint8_t *>(pPattern);
758-
Result = UR_CHECK_ERROR(cuMemsetD8Async(DstDevice, Value, N, Stream));
753+
UR_CHECK_ERROR(cuMemsetD8Async(DstDevice, Value, N, Stream));
759754
break;
760755
}
761756
case 2: {
762757
auto Value = *static_cast<const uint16_t *>(pPattern);
763-
Result = UR_CHECK_ERROR(cuMemsetD16Async(DstDevice, Value, N, Stream));
758+
UR_CHECK_ERROR(cuMemsetD16Async(DstDevice, Value, N, Stream));
764759
break;
765760
}
766761
case 4: {
767762
auto Value = *static_cast<const uint32_t *>(pPattern);
768-
Result = UR_CHECK_ERROR(cuMemsetD32Async(DstDevice, Value, N, Stream));
763+
UR_CHECK_ERROR(cuMemsetD32Async(DstDevice, Value, N, Stream));
769764
break;
770765
}
771766
default: {
@@ -843,7 +838,8 @@ static ur_result_t commonEnqueueMemImageNDCopy(
843838
}
844839
CpyDesc.WidthInBytes = Region.width;
845840
CpyDesc.Height = Region.height;
846-
return UR_CHECK_ERROR(cuMemcpy2DAsync(&CpyDesc, CuStream));
841+
UR_CHECK_ERROR(cuMemcpy2DAsync(&CpyDesc, CuStream));
842+
return UR_RESULT_SUCCESS;
847843
}
848844
if (ImgType == UR_MEM_TYPE_IMAGE3D) {
849845
CUDA_MEMCPY3D CpyDesc;
@@ -869,7 +865,8 @@ static ur_result_t commonEnqueueMemImageNDCopy(
869865
CpyDesc.WidthInBytes = Region.width;
870866
CpyDesc.Height = Region.height;
871867
CpyDesc.Depth = Region.depth;
872-
return UR_CHECK_ERROR(cuMemcpy3DAsync(&CpyDesc, CuStream));
868+
UR_CHECK_ERROR(cuMemcpy3DAsync(&CpyDesc, CuStream));
869+
return UR_RESULT_SUCCESS;
873870
}
874871
return UR_RESULT_ERROR_INVALID_VALUE;
875872
}
@@ -896,7 +893,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageRead(
896893
CUarray Array = hImage->Mem.SurfaceMem.getArray();
897894

898895
CUDA_ARRAY_DESCRIPTOR ArrayDesc;
899-
Result = UR_CHECK_ERROR(cuArrayGetDescriptor(&ArrayDesc, Array));
896+
UR_CHECK_ERROR(cuArrayGetDescriptor(&ArrayDesc, Array));
900897

901898
int ElementByteSize = imageElementByteSize(ArrayDesc);
902899

@@ -913,7 +910,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageRead(
913910
UR_CHECK_ERROR(RetImplEvent->start());
914911
}
915912
if (ImgType == UR_MEM_TYPE_IMAGE1D) {
916-
Result = UR_CHECK_ERROR(
913+
UR_CHECK_ERROR(
917914
cuMemcpyAtoHAsync(pDst, Array, ByteOffsetX, BytesToCopy, CuStream));
918915
} else {
919916
ur_rect_region_t AdjustedRegion = {BytesToCopy, region.height,
@@ -923,7 +920,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageRead(
923920
Result = commonEnqueueMemImageNDCopy(
924921
CuStream, ImgType, AdjustedRegion, &Array, CU_MEMORYTYPE_ARRAY,
925922
SrcOffset, pDst, CU_MEMORYTYPE_HOST, ur_rect_offset_t{});
926-
927923
if (Result != UR_RESULT_SUCCESS) {
928924
return Result;
929925
}
@@ -935,7 +931,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageRead(
935931
}
936932

937933
if (blockingRead) {
938-
Result = UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
934+
UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
939935
}
940936
} catch (ur_result_t Err) {
941937
return Err;
@@ -969,7 +965,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageWrite(
969965
CUarray Array = hImage->Mem.SurfaceMem.getArray();
970966

971967
CUDA_ARRAY_DESCRIPTOR ArrayDesc;
972-
Result = UR_CHECK_ERROR(cuArrayGetDescriptor(&ArrayDesc, Array));
968+
UR_CHECK_ERROR(cuArrayGetDescriptor(&ArrayDesc, Array));
973969

974970
int ElementByteSize = imageElementByteSize(ArrayDesc);
975971

@@ -986,7 +982,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageWrite(
986982

987983
ur_mem_type_t ImgType = hImage->Mem.SurfaceMem.getImageType();
988984
if (ImgType == UR_MEM_TYPE_IMAGE1D) {
989-
Result = UR_CHECK_ERROR(
985+
UR_CHECK_ERROR(
990986
cuMemcpyHtoAAsync(Array, ByteOffsetX, pSrc, BytesToCopy, CuStream));
991987
} else {
992988
ur_rect_region_t AdjustedRegion = {BytesToCopy, region.height,
@@ -1041,9 +1037,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageCopy(
10411037
CUarray DstArray = hImageDst->Mem.SurfaceMem.getArray();
10421038

10431039
CUDA_ARRAY_DESCRIPTOR SrcArrayDesc;
1044-
Result = UR_CHECK_ERROR(cuArrayGetDescriptor(&SrcArrayDesc, SrcArray));
1040+
UR_CHECK_ERROR(cuArrayGetDescriptor(&SrcArrayDesc, SrcArray));
10451041
CUDA_ARRAY_DESCRIPTOR DstArrayDesc;
1046-
Result = UR_CHECK_ERROR(cuArrayGetDescriptor(&DstArrayDesc, DstArray));
1042+
UR_CHECK_ERROR(cuArrayGetDescriptor(&DstArrayDesc, DstArray));
10471043

10481044
UR_ASSERT(SrcArrayDesc.Format == DstArrayDesc.Format,
10491045
UR_RESULT_ERROR_INVALID_MEM_OBJECT);
@@ -1069,8 +1065,8 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageCopy(
10691065

10701066
ur_mem_type_t ImgType = hImageSrc->Mem.SurfaceMem.getImageType();
10711067
if (ImgType == UR_MEM_TYPE_IMAGE1D) {
1072-
Result = UR_CHECK_ERROR(cuMemcpyAtoA(DstArray, DstByteOffsetX, SrcArray,
1073-
SrcByteOffsetX, BytesToCopy));
1068+
UR_CHECK_ERROR(cuMemcpyAtoA(DstArray, DstByteOffsetX, SrcArray,
1069+
SrcByteOffsetX, BytesToCopy));
10741070
} else {
10751071
ur_rect_region_t AdjustedRegion = {BytesToCopy, region.height,
10761072
region.depth};
@@ -1080,7 +1076,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageCopy(
10801076
Result = commonEnqueueMemImageNDCopy(
10811077
CuStream, ImgType, AdjustedRegion, &SrcArray, CU_MEMORYTYPE_ARRAY,
10821078
SrcOffset, &DstArray, CU_MEMORYTYPE_ARRAY, DstOffset);
1083-
10841079
if (Result != UR_RESULT_SUCCESS) {
10851080
return Result;
10861081
}
@@ -1282,13 +1277,13 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueUSMMemcpy(
12821277
UR_COMMAND_USM_MEMCPY, hQueue, CuStream));
12831278
UR_CHECK_ERROR(EventPtr->start());
12841279
}
1285-
Result = UR_CHECK_ERROR(
1280+
UR_CHECK_ERROR(
12861281
cuMemcpyAsync((CUdeviceptr)pDst, (CUdeviceptr)pSrc, size, CuStream));
12871282
if (phEvent) {
12881283
UR_CHECK_ERROR(EventPtr->record());
12891284
}
12901285
if (blocking) {
1291-
Result = UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
1286+
UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
12921287
}
12931288
if (phEvent) {
12941289
*phEvent = EventPtr.release();
@@ -1347,7 +1342,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueUSMPrefetch(
13471342
UR_COMMAND_MEM_BUFFER_COPY, hQueue, CuStream));
13481343
UR_CHECK_ERROR(EventPtr->start());
13491344
}
1350-
Result = UR_CHECK_ERROR(
1345+
UR_CHECK_ERROR(
13511346
cuMemPrefetchAsync((CUdeviceptr)pMem, size, Device->get(), CuStream));
13521347
if (phEvent) {
13531348
UR_CHECK_ERROR(EventPtr->record());
@@ -1485,14 +1480,14 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueUSMMemcpy2D(
14851480
CpyDesc.WidthInBytes = width;
14861481
CpyDesc.Height = height;
14871482

1488-
result = UR_CHECK_ERROR(cuMemcpy2DAsync(&CpyDesc, cuStream));
1483+
UR_CHECK_ERROR(cuMemcpy2DAsync(&CpyDesc, cuStream));
14891484

14901485
if (phEvent) {
14911486
UR_CHECK_ERROR(RetImplEvent->record());
14921487
*phEvent = RetImplEvent.release();
14931488
}
14941489
if (blocking) {
1495-
result = UR_CHECK_ERROR(cuStreamSynchronize(cuStream));
1490+
UR_CHECK_ERROR(cuStreamSynchronize(cuStream));
14961491
}
14971492
} catch (ur_result_t err) {
14981493
result = err;
@@ -1608,9 +1603,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueDeviceGlobalVariableWrite(
16081603
try {
16091604
CUdeviceptr DeviceGlobal = 0;
16101605
size_t DeviceGlobalSize = 0;
1611-
Result = UR_CHECK_ERROR(cuModuleGetGlobal(&DeviceGlobal, &DeviceGlobalSize,
1612-
hProgram->get(),
1613-
DeviceGlobalName.c_str()));
1606+
UR_CHECK_ERROR(cuModuleGetGlobal(&DeviceGlobal, &DeviceGlobalSize,
1607+
hProgram->get(),
1608+
DeviceGlobalName.c_str()));
16141609

16151610
if (offset + count > DeviceGlobalSize)
16161611
return UR_RESULT_ERROR_INVALID_VALUE;
@@ -1640,9 +1635,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueDeviceGlobalVariableRead(
16401635
try {
16411636
CUdeviceptr DeviceGlobal = 0;
16421637
size_t DeviceGlobalSize = 0;
1643-
Result = UR_CHECK_ERROR(cuModuleGetGlobal(&DeviceGlobal, &DeviceGlobalSize,
1644-
hProgram->get(),
1645-
DeviceGlobalName.c_str()));
1638+
UR_CHECK_ERROR(cuModuleGetGlobal(&DeviceGlobal, &DeviceGlobalSize,
1639+
hProgram->get(),
1640+
DeviceGlobalName.c_str()));
16461641

16471642
if (offset + count > DeviceGlobalSize)
16481643
return UR_RESULT_ERROR_INVALID_VALUE;

0 commit comments

Comments
 (0)