Skip to content

[SYCL][NFC] Make UR_CHECK_ERROR a void return macro #11100

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 5 commits into from
Sep 22, 2023
Merged

[SYCL][NFC] Make UR_CHECK_ERROR a void return macro #11100

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
5 commits
Select commit Hold shift + click to select a range
b233b62
[SYCL][NFC] Make UR_CHECK_ERROR a void return function
jchlanda Sep 7, 2023
b0fa260
Explicit throw
jchlanda Sep 18, 2023
db018ed
Simplify error checking
jchlanda Sep 18, 2023
b164aa9
Merge fix
jchlanda Sep 18, 2023
f010abe
Update CUDA's signature
jchlanda Sep 18, 2023
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
12 changes: 6 additions & 6 deletions sycl/plugins/unified_runtime/ur/adapters/cuda/common.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -33,10 +33,10 @@ ur_result_t mapErrorUR(CUresult Result) {
}
}

ur_result_t checkErrorUR(CUresult Result, const char *Function, int Line,
const char *File) {
void checkErrorUR(CUresult Result, const char *Function, int Line,
const char *File) {
if (Result == CUDA_SUCCESS || Result == CUDA_ERROR_DEINITIALIZED) {
return UR_RESULT_SUCCESS;
return;
}

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

ur_result_t checkErrorUR(ur_result_t Result, const char *Function, int Line,
const char *File) {
void checkErrorUR(ur_result_t Result, const char *Function, int Line,
const char *File) {
if (Result == UR_RESULT_SUCCESS) {
return UR_RESULT_SUCCESS;
return;
}

if (std::getenv("SYCL_PI_SUPPRESS_ERROR_MESSAGE") == nullptr &&
Expand Down
8 changes: 4 additions & 4 deletions sycl/plugins/unified_runtime/ur/adapters/cuda/common.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -19,11 +19,11 @@ ur_result_t mapErrorUR(CUresult Result);
/// \return UR_RESULT_SUCCESS if \param Result was CUDA_SUCCESS.
/// \throw ur_result_t exception (integer) if input was not success.
///
ur_result_t checkErrorUR(CUresult Result, const char *Function, int Line,
const char *File);
void checkErrorUR(CUresult Result, const char *Function, int Line,
const char *File);

ur_result_t checkErrorUR(ur_result_t Result, const char *Function, int Line,
const char *File);
void checkErrorUR(ur_result_t Result, const char *Function, int Line,
const char *File);

#define UR_CHECK_ERROR(Result) \
checkErrorUR(Result, __func__, __LINE__, __FILE__)
Expand Down
89 changes: 42 additions & 47 deletions sycl/plugins/unified_runtime/ur/adapters/cuda/enqueue.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -31,7 +31,8 @@ ur_result_t enqueueEventsWait(ur_queue_handle_t CommandQueue, CUstream Stream,
if (Event->getStream() == Stream) {
return UR_RESULT_SUCCESS;
} else {
return UR_CHECK_ERROR(cuStreamWaitEvent(Stream, Event->get(), 0));
UR_CHECK_ERROR(cuStreamWaitEvent(Stream, Event->get(), 0));
return UR_RESULT_SUCCESS;
}
});
return Result;
Expand Down Expand Up @@ -193,8 +194,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueEventsWaitWithBarrier(
const ur_event_handle_t *phEventWaitList, ur_event_handle_t *phEvent) {
// This function makes one stream work on the previous work (or work
// represented by input events) and then all future work waits on that stream.
ur_result_t Result;

try {
ScopedContext Active(hQueue->getContext());
uint32_t StreamToken;
Expand Down Expand Up @@ -228,23 +227,21 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueEventsWaitWithBarrier(
Event->getComputeStreamToken())) {
return UR_RESULT_SUCCESS;
} else {
return UR_CHECK_ERROR(
UR_CHECK_ERROR(
cuStreamWaitEvent(CuStream, Event->get(), 0));
return UR_RESULT_SUCCESS;
}
});
}

Result = UR_CHECK_ERROR(cuEventRecord(hQueue->BarrierEvent, CuStream));
UR_CHECK_ERROR(cuEventRecord(hQueue->BarrierEvent, CuStream));
for (unsigned int i = 0; i < hQueue->ComputeAppliedBarrier.size(); i++) {
hQueue->ComputeAppliedBarrier[i] = false;
}
for (unsigned int i = 0; i < hQueue->TransferAppliedBarrier.size(); i++) {
hQueue->TransferAppliedBarrier[i] = false;
}
}
if (Result != UR_RESULT_SUCCESS) {
return Result;
}

if (phEvent) {
*phEvent = ur_event_handle_t_::makeNative(
Expand Down Expand Up @@ -430,7 +427,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueKernelLaunch(
Device->getMaxChosenLocalMem()));
}

Result = UR_CHECK_ERROR(cuLaunchKernel(
UR_CHECK_ERROR(cuLaunchKernel(
CuFunc, BlocksPerGrid[0], BlocksPerGrid[1], BlocksPerGrid[2],
ThreadsPerBlock[0], ThreadsPerBlock[1], ThreadsPerBlock[2], LocalSize,
CuStream, const_cast<void **>(ArgIndices.data()), nullptr));
Expand Down Expand Up @@ -502,7 +499,9 @@ static ur_result_t commonEnqueueMemBufferCopyRect(
params.dstPitch = dst_row_pitch;
params.dstHeight = dst_slice_pitch / dst_row_pitch;

return UR_CHECK_ERROR(cuMemcpy3DAsync(&params, cu_stream));
UR_CHECK_ERROR(cuMemcpy3DAsync(&params, cu_stream));

return UR_RESULT_SUCCESS;
}

UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferReadRect(
Expand Down Expand Up @@ -540,7 +539,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferReadRect(
}

if (blockingRead) {
Result = UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
}

if (phEvent) {
Expand Down Expand Up @@ -587,7 +586,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferWriteRect(
}

if (blockingWrite) {
Result = UR_CHECK_ERROR(cuStreamSynchronize(cuStream));
UR_CHECK_ERROR(cuStreamSynchronize(cuStream));
}

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

try {
ScopedContext Active(hQueue->getContext());
ur_result_t Result;
ur_result_t Result = UR_RESULT_SUCCESS;

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

Result = UR_CHECK_ERROR(cuMemcpyDtoDAsync(Dst, Src, size, Stream));
UR_CHECK_ERROR(cuMemcpyDtoDAsync(Dst, Src, size, Stream));

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

// Get 4-byte chunk of the pattern and call cuMemsetD32Async
auto Value = *(static_cast<const uint32_t *>(pPattern));
auto Result = UR_CHECK_ERROR(cuMemsetD32Async(Ptr, Value, Count32, Stream));
if (Result != UR_RESULT_SUCCESS) {
return Result;
}
UR_CHECK_ERROR(cuMemsetD32Async(Ptr, Value, Count32, Stream));
for (auto step = 4u; step < NumberOfSteps; ++step) {
// take 1 byte of the pattern
Value = *(static_cast<const uint8_t *>(pPattern) + step);
Expand Down Expand Up @@ -737,8 +733,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferFill(
ScopedContext Active(hQueue->getContext());

auto Stream = hQueue->getNextTransferStream();
ur_result_t Result;
Result =
ur_result_t Result =
enqueueEventsWait(hQueue, Stream, numEventsInWaitList, phEventWaitList);

if (phEvent) {
Expand All @@ -755,17 +750,17 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemBufferFill(
switch (patternSize) {
case 1: {
auto Value = *static_cast<const uint8_t *>(pPattern);
Result = UR_CHECK_ERROR(cuMemsetD8Async(DstDevice, Value, N, Stream));
UR_CHECK_ERROR(cuMemsetD8Async(DstDevice, Value, N, Stream));
break;
}
case 2: {
auto Value = *static_cast<const uint16_t *>(pPattern);
Result = UR_CHECK_ERROR(cuMemsetD16Async(DstDevice, Value, N, Stream));
UR_CHECK_ERROR(cuMemsetD16Async(DstDevice, Value, N, Stream));
break;
}
case 4: {
auto Value = *static_cast<const uint32_t *>(pPattern);
Result = UR_CHECK_ERROR(cuMemsetD32Async(DstDevice, Value, N, Stream));
UR_CHECK_ERROR(cuMemsetD32Async(DstDevice, Value, N, Stream));
break;
}
default: {
Expand Down Expand Up @@ -843,7 +838,8 @@ static ur_result_t commonEnqueueMemImageNDCopy(
}
CpyDesc.WidthInBytes = Region.width;
CpyDesc.Height = Region.height;
return UR_CHECK_ERROR(cuMemcpy2DAsync(&CpyDesc, CuStream));
UR_CHECK_ERROR(cuMemcpy2DAsync(&CpyDesc, CuStream));
return UR_RESULT_SUCCESS;
}
if (ImgType == UR_MEM_TYPE_IMAGE3D) {
CUDA_MEMCPY3D CpyDesc;
Expand All @@ -869,7 +865,8 @@ static ur_result_t commonEnqueueMemImageNDCopy(
CpyDesc.WidthInBytes = Region.width;
CpyDesc.Height = Region.height;
CpyDesc.Depth = Region.depth;
return UR_CHECK_ERROR(cuMemcpy3DAsync(&CpyDesc, CuStream));
UR_CHECK_ERROR(cuMemcpy3DAsync(&CpyDesc, CuStream));
return UR_RESULT_SUCCESS;
}
return UR_RESULT_ERROR_INVALID_VALUE;
}
Expand All @@ -896,7 +893,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageRead(
CUarray Array = hImage->Mem.SurfaceMem.getArray();

CUDA_ARRAY_DESCRIPTOR ArrayDesc;
Result = UR_CHECK_ERROR(cuArrayGetDescriptor(&ArrayDesc, Array));
UR_CHECK_ERROR(cuArrayGetDescriptor(&ArrayDesc, Array));

int ElementByteSize = imageElementByteSize(ArrayDesc);

Expand All @@ -913,7 +910,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageRead(
UR_CHECK_ERROR(RetImplEvent->start());
}
if (ImgType == UR_MEM_TYPE_IMAGE1D) {
Result = UR_CHECK_ERROR(
UR_CHECK_ERROR(
cuMemcpyAtoHAsync(pDst, Array, ByteOffsetX, BytesToCopy, CuStream));
} else {
ur_rect_region_t AdjustedRegion = {BytesToCopy, region.height,
Expand All @@ -923,7 +920,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageRead(
Result = commonEnqueueMemImageNDCopy(
CuStream, ImgType, AdjustedRegion, &Array, CU_MEMORYTYPE_ARRAY,
SrcOffset, pDst, CU_MEMORYTYPE_HOST, ur_rect_offset_t{});

if (Result != UR_RESULT_SUCCESS) {
return Result;
}
Expand All @@ -935,7 +931,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageRead(
}

if (blockingRead) {
Result = UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
}
} catch (ur_result_t Err) {
return Err;
Expand Down Expand Up @@ -969,7 +965,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageWrite(
CUarray Array = hImage->Mem.SurfaceMem.getArray();

CUDA_ARRAY_DESCRIPTOR ArrayDesc;
Result = UR_CHECK_ERROR(cuArrayGetDescriptor(&ArrayDesc, Array));
UR_CHECK_ERROR(cuArrayGetDescriptor(&ArrayDesc, Array));

int ElementByteSize = imageElementByteSize(ArrayDesc);

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

ur_mem_type_t ImgType = hImage->Mem.SurfaceMem.getImageType();
if (ImgType == UR_MEM_TYPE_IMAGE1D) {
Result = UR_CHECK_ERROR(
UR_CHECK_ERROR(
cuMemcpyHtoAAsync(Array, ByteOffsetX, pSrc, BytesToCopy, CuStream));
} else {
ur_rect_region_t AdjustedRegion = {BytesToCopy, region.height,
Expand Down Expand Up @@ -1041,9 +1037,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageCopy(
CUarray DstArray = hImageDst->Mem.SurfaceMem.getArray();

CUDA_ARRAY_DESCRIPTOR SrcArrayDesc;
Result = UR_CHECK_ERROR(cuArrayGetDescriptor(&SrcArrayDesc, SrcArray));
UR_CHECK_ERROR(cuArrayGetDescriptor(&SrcArrayDesc, SrcArray));
CUDA_ARRAY_DESCRIPTOR DstArrayDesc;
Result = UR_CHECK_ERROR(cuArrayGetDescriptor(&DstArrayDesc, DstArray));
UR_CHECK_ERROR(cuArrayGetDescriptor(&DstArrayDesc, DstArray));

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

ur_mem_type_t ImgType = hImageSrc->Mem.SurfaceMem.getImageType();
if (ImgType == UR_MEM_TYPE_IMAGE1D) {
Result = UR_CHECK_ERROR(cuMemcpyAtoA(DstArray, DstByteOffsetX, SrcArray,
SrcByteOffsetX, BytesToCopy));
UR_CHECK_ERROR(cuMemcpyAtoA(DstArray, DstByteOffsetX, SrcArray,
SrcByteOffsetX, BytesToCopy));
} else {
ur_rect_region_t AdjustedRegion = {BytesToCopy, region.height,
region.depth};
Expand All @@ -1080,7 +1076,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueMemImageCopy(
Result = commonEnqueueMemImageNDCopy(
CuStream, ImgType, AdjustedRegion, &SrcArray, CU_MEMORYTYPE_ARRAY,
SrcOffset, &DstArray, CU_MEMORYTYPE_ARRAY, DstOffset);

if (Result != UR_RESULT_SUCCESS) {
return Result;
}
Expand Down Expand Up @@ -1282,13 +1277,13 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueUSMMemcpy(
UR_COMMAND_USM_MEMCPY, hQueue, CuStream));
UR_CHECK_ERROR(EventPtr->start());
}
Result = UR_CHECK_ERROR(
UR_CHECK_ERROR(
cuMemcpyAsync((CUdeviceptr)pDst, (CUdeviceptr)pSrc, size, CuStream));
if (phEvent) {
UR_CHECK_ERROR(EventPtr->record());
}
if (blocking) {
Result = UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
UR_CHECK_ERROR(cuStreamSynchronize(CuStream));
}
if (phEvent) {
*phEvent = EventPtr.release();
Expand Down Expand Up @@ -1347,7 +1342,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueUSMPrefetch(
UR_COMMAND_MEM_BUFFER_COPY, hQueue, CuStream));
UR_CHECK_ERROR(EventPtr->start());
}
Result = UR_CHECK_ERROR(
UR_CHECK_ERROR(
cuMemPrefetchAsync((CUdeviceptr)pMem, size, Device->get(), CuStream));
if (phEvent) {
UR_CHECK_ERROR(EventPtr->record());
Expand Down Expand Up @@ -1485,14 +1480,14 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueUSMMemcpy2D(
CpyDesc.WidthInBytes = width;
CpyDesc.Height = height;

result = UR_CHECK_ERROR(cuMemcpy2DAsync(&CpyDesc, cuStream));
UR_CHECK_ERROR(cuMemcpy2DAsync(&CpyDesc, cuStream));

if (phEvent) {
UR_CHECK_ERROR(RetImplEvent->record());
*phEvent = RetImplEvent.release();
}
if (blocking) {
result = UR_CHECK_ERROR(cuStreamSynchronize(cuStream));
UR_CHECK_ERROR(cuStreamSynchronize(cuStream));
}
} catch (ur_result_t err) {
result = err;
Expand Down Expand Up @@ -1608,9 +1603,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueDeviceGlobalVariableWrite(
try {
CUdeviceptr DeviceGlobal = 0;
size_t DeviceGlobalSize = 0;
Result = UR_CHECK_ERROR(cuModuleGetGlobal(&DeviceGlobal, &DeviceGlobalSize,
hProgram->get(),
DeviceGlobalName.c_str()));
UR_CHECK_ERROR(cuModuleGetGlobal(&DeviceGlobal, &DeviceGlobalSize,
hProgram->get(),
DeviceGlobalName.c_str()));

if (offset + count > DeviceGlobalSize)
return UR_RESULT_ERROR_INVALID_VALUE;
Expand Down Expand Up @@ -1640,9 +1635,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urEnqueueDeviceGlobalVariableRead(
try {
CUdeviceptr DeviceGlobal = 0;
size_t DeviceGlobalSize = 0;
Result = UR_CHECK_ERROR(cuModuleGetGlobal(&DeviceGlobal, &DeviceGlobalSize,
hProgram->get(),
DeviceGlobalName.c_str()));
UR_CHECK_ERROR(cuModuleGetGlobal(&DeviceGlobal, &DeviceGlobalSize,
hProgram->get(),
DeviceGlobalName.c_str()));

if (offset + count > DeviceGlobalSize)
return UR_RESULT_ERROR_INVALID_VALUE;
Expand Down
Loading