From fbba37070cb70526b1653ae5306b66a377fa0e9d Mon Sep 17 00:00:00 2001 From: Saleel Kudchadker Date: Fri, 5 Jun 2020 10:45:28 -0700 Subject: [PATCH] Modify HIP_RETURN to print useful details Change-Id: I23892c2d9a738b0298cdf24106d688a792937c73 --- hipamd/rocclr/hip_event.cpp | 6 +++--- hipamd/rocclr/hip_internal.hpp | 16 ++++++++++++---- hipamd/rocclr/hip_memory.cpp | 20 ++++++++++---------- hipamd/rocclr/hip_stream.cpp | 10 ++++------ 4 files changed, 29 insertions(+), 23 deletions(-) diff --git a/hipamd/rocclr/hip_event.cpp b/hipamd/rocclr/hip_event.cpp index 584a133613..7fd84c0fb7 100644 --- a/hipamd/rocclr/hip_event.cpp +++ b/hipamd/rocclr/hip_event.cpp @@ -192,13 +192,13 @@ hipError_t ihipEventQuery(hipEvent_t event) { hipError_t hipEventCreateWithFlags(hipEvent_t* event, unsigned flags) { HIP_INIT_API(hipEventCreateWithFlags, event, flags); - HIP_RETURN(ihipEventCreateWithFlags(event, flags)); + HIP_RETURN(ihipEventCreateWithFlags(event, flags), *event); } hipError_t hipEventCreate(hipEvent_t* event) { HIP_INIT_API(hipEventCreate, event); - HIP_RETURN(ihipEventCreateWithFlags(event, 0)); + HIP_RETURN(ihipEventCreateWithFlags(event, 0), *event); } hipError_t hipEventDestroy(hipEvent_t event) { @@ -227,7 +227,7 @@ hipError_t hipEventElapsedTime(float *ms, hipEvent_t start, hipEvent_t stop) { hip::Event* eStart = reinterpret_cast(start); hip::Event* eStop = reinterpret_cast(stop); - HIP_RETURN(eStart->elapsedTime(*eStop, *ms)); + HIP_RETURN(eStart->elapsedTime(*eStop, *ms), "Elapsed Time = ", *ms); } hipError_t hipEventRecord(hipEvent_t event, hipStream_t stream) { diff --git a/hipamd/rocclr/hip_internal.hpp b/hipamd/rocclr/hip_internal.hpp index 264b68ecc2..c42bda9fec 100755 --- a/hipamd/rocclr/hip_internal.hpp +++ b/hipamd/rocclr/hip_internal.hpp @@ -57,9 +57,17 @@ typedef struct ihipIpcMemHandle_st { hip::g_device = g_devices[0]; \ } +#define HIP_API_PRINT(...) \ + ClPrint(amd::LOG_INFO, amd::LOG_API, "%-5d: [%zx] %s ( %s )", getpid(), std::this_thread::get_id(), \ + __func__, ToString( __VA_ARGS__ ).c_str()); + +#define HIP_ERROR_PRINT(err, ...) \ + ClPrint(amd::LOG_INFO, amd::LOG_API, "%-5d: [%zx] %s: Returned %s : %s", getpid(), std::this_thread::get_id(), \ + __func__, hipGetErrorName(err), ToString( __VA_ARGS__ ).c_str()); + // This macro should be called at the beginning of every HIP API. #define HIP_INIT_API(cid, ...) \ - ClPrint(amd::LOG_INFO, amd::LOG_API, "%-5d: [%zx] %s ( %s )", getpid(), std::this_thread::get_id(), __func__, ToString( __VA_ARGS__ ).c_str()); \ + HIP_API_PRINT(__VA_ARGS__) \ amd::Thread* thread = amd::Thread::current(); \ if (!VDI_CHECK_THREAD(thread)) { \ HIP_RETURN(hipErrorOutOfMemory); \ @@ -67,9 +75,9 @@ typedef struct ihipIpcMemHandle_st { HIP_INIT() \ HIP_CB_SPAWNER_OBJECT(cid); -#define HIP_RETURN(ret) \ - hip::g_lastError = ret; \ - ClPrint(amd::LOG_INFO, amd::LOG_API, "%-5d: [%zx] %s: Returned %s", getpid(), std::this_thread::get_id(), __func__, hipGetErrorName(hip::g_lastError)); \ +#define HIP_RETURN(ret, ...) \ + hip::g_lastError = ret; \ + HIP_ERROR_PRINT(hip::g_lastError, __VA_ARGS__) \ return hip::g_lastError; namespace hc { diff --git a/hipamd/rocclr/hip_memory.cpp b/hipamd/rocclr/hip_memory.cpp index 290b98cd45..9af33d9c74 100755 --- a/hipamd/rocclr/hip_memory.cpp +++ b/hipamd/rocclr/hip_memory.cpp @@ -100,7 +100,7 @@ hipError_t ihipMalloc(void** ptr, size_t sizeBytes, unsigned int flags) if (*ptr == nullptr) { return hipErrorOutOfMemory; } - ClPrint(amd::LOG_INFO, amd::LOG_API, "%-5d: [%zx] ihipMalloc ptr=0x%zx", getpid(),std::this_thread::get_id(), *ptr); + return hipSuccess; } @@ -207,13 +207,13 @@ hipError_t hipExtMallocWithFlags(void** ptr, size_t sizeBytes, unsigned int flag HIP_RETURN(hipErrorInvalidValue); } - HIP_RETURN(ihipMalloc(ptr, sizeBytes, (flags & hipDeviceMallocFinegrained)? CL_MEM_SVM_ATOMICS: 0)); + HIP_RETURN(ihipMalloc(ptr, sizeBytes, (flags & hipDeviceMallocFinegrained)? CL_MEM_SVM_ATOMICS: 0), *ptr); } hipError_t hipMalloc(void** ptr, size_t sizeBytes) { HIP_INIT_API(hipMalloc, ptr, sizeBytes); - HIP_RETURN(ihipMalloc(ptr, sizeBytes, 0)); + HIP_RETURN(ihipMalloc(ptr, sizeBytes, 0), *ptr); } hipError_t hipHostMalloc(void** ptr, size_t sizeBytes, unsigned int flags) { @@ -241,7 +241,7 @@ hipError_t hipHostMalloc(void** ptr, size_t sizeBytes, unsigned int flags) { ihipFlags |= CL_MEM_SVM_ATOMICS; } - HIP_RETURN(ihipMalloc(ptr, sizeBytes, ihipFlags)); + HIP_RETURN(ihipMalloc(ptr, sizeBytes, ihipFlags), *ptr); } hipError_t hipMallocManaged(void** devPtr, size_t size, @@ -252,7 +252,7 @@ hipError_t hipMallocManaged(void** devPtr, size_t size, HIP_RETURN(hipErrorInvalidValue); } - HIP_RETURN(ihipMalloc(devPtr, size, CL_MEM_SVM_FINE_GRAIN_BUFFER)); + HIP_RETURN(ihipMalloc(devPtr, size, CL_MEM_SVM_FINE_GRAIN_BUFFER), *devPtr); } hipError_t hipFree(void* ptr) { @@ -399,7 +399,7 @@ hipError_t hipMallocPitch(void** ptr, size_t* pitch, size_t width, size_t height HIP_INIT_API(hipMallocPitch, ptr, pitch, width, height); const cl_image_format image_format = { CL_R, CL_UNSIGNED_INT8 }; - HIP_RETURN(ihipMallocPitch(ptr, pitch, width, height, 1, CL_MEM_OBJECT_IMAGE2D, &image_format)); + HIP_RETURN(ihipMallocPitch(ptr, pitch, width, height, 1, CL_MEM_OBJECT_IMAGE2D, &image_format), *ptr); } hipError_t hipMalloc3D(hipPitchedPtr* pitchedDevPtr, hipExtent extent) { @@ -422,7 +422,7 @@ hipError_t hipMalloc3D(hipPitchedPtr* pitchedDevPtr, hipExtent extent) { pitchedDevPtr->ysize = extent.height; } - HIP_RETURN(status); + HIP_RETURN(status, *pitchedDevPtr); } amd::Image* ihipImageCreate(const cl_channel_order channelOrder, @@ -698,7 +698,7 @@ hipError_t hipHostRegister(void* hostPtr, size_t sizeBytes, unsigned int flags) amd::MemObjMap::AddMemObj(hostPtr, mem); HIP_RETURN(hipSuccess); } else { - HIP_RETURN(ihipMalloc(&hostPtr, sizeBytes, flags)); + HIP_RETURN(ihipMalloc(&hostPtr, sizeBytes, flags), hostPtr); } } @@ -733,7 +733,7 @@ hipError_t hipHostUnregister(void* hostPtr) { // Deprecated function: hipError_t hipHostAlloc(void** ptr, size_t sizeBytes, unsigned int flags) { - HIP_RETURN(ihipMalloc(ptr, sizeBytes, flags)); + HIP_RETURN(ihipMalloc(ptr, sizeBytes, flags), *ptr); }; @@ -2303,7 +2303,7 @@ hipError_t hipMallocHost(void** ptr, HIP_RETURN(hipErrorInvalidValue); } - HIP_RETURN(ihipMalloc(ptr, size, CL_MEM_SVM_FINE_GRAIN_BUFFER)); + HIP_RETURN(ihipMalloc(ptr, size, CL_MEM_SVM_FINE_GRAIN_BUFFER), *ptr); } hipError_t hipFreeHost(void *ptr) { diff --git a/hipamd/rocclr/hip_stream.cpp b/hipamd/rocclr/hip_stream.cpp index 59e043385e..e2ab7b18cd 100755 --- a/hipamd/rocclr/hip_stream.cpp +++ b/hipamd/rocclr/hip_stream.cpp @@ -195,8 +195,6 @@ static hipError_t ihipStreamCreate(hipStream_t* stream, *stream = reinterpret_cast(hStream); - ClPrint(amd::LOG_INFO, amd::LOG_API, "ihipStreamCreate: %zx", hStream); - return hipSuccess; } @@ -204,14 +202,14 @@ static hipError_t ihipStreamCreate(hipStream_t* stream, hipError_t hipStreamCreateWithFlags(hipStream_t *stream, unsigned int flags) { HIP_INIT_API(hipStreamCreateWithFlags, stream, flags); - HIP_RETURN(ihipStreamCreate(stream, flags, hip::Stream::Priority::Normal)); + HIP_RETURN(ihipStreamCreate(stream, flags, hip::Stream::Priority::Normal), *stream); } // ================================================================================================ hipError_t hipStreamCreate(hipStream_t *stream) { HIP_INIT_API(hipStreamCreate, stream); - HIP_RETURN(ihipStreamCreate(stream, hipStreamDefault, hip::Stream::Priority::Normal)); + HIP_RETURN(ihipStreamCreate(stream, hipStreamDefault, hip::Stream::Priority::Normal), *stream); } // ================================================================================================ @@ -227,7 +225,7 @@ hipError_t hipStreamCreateWithPriority(hipStream_t* stream, unsigned int flags, streamPriority = hip::Stream::Priority::Normal; } - HIP_RETURN(ihipStreamCreate(stream, flags, streamPriority)); + HIP_RETURN(ihipStreamCreate(stream, flags, streamPriority), *stream); } // ================================================================================================ @@ -354,7 +352,7 @@ hipError_t hipExtStreamCreateWithCUMask(hipStream_t* stream, uint32_t cuMaskSize const std::vector cuMaskv(cuMask, cuMask + cuMaskSize); - HIP_RETURN(ihipStreamCreate(stream, hipStreamDefault, hip::Stream::Priority::Normal, cuMaskv)); + HIP_RETURN(ihipStreamCreate(stream, hipStreamDefault, hip::Stream::Priority::Normal, cuMaskv), *stream); } // ================================================================================================