Modify HIP_RETURN to print useful details

Change-Id: I23892c2d9a738b0298cdf24106d688a792937c73
2020-06-05 10:45:28 -07:00
@@ -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<hip::Event*>(start);
  hip::Event* eStop  = reinterpret_cast<hip::Event*>(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) {
@@ -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 {
@@ -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) {
@@ -195,8 +195,6 @@ static hipError_t ihipStreamCreate(hipStream_t* stream,

  *stream = reinterpret_cast<hipStream_t>(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<uint32_t> 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);
 }

 // ================================================================================================