[Mlir-commits] [mlir] 4c1eaf2 - [mlir] fix the rocm runtime wrapper to account for cuda / rocm api differences

[Tobias Gysi]

Author: Tobias Gysi
Date: 2021-01-20T18:48:32+01:00
New Revision: 4c1eaf26ae70b9f0e441b0f613871d697c4c9a7d

URL: https://github.com/llvm/llvm-project/commit/4c1eaf26ae70b9f0e441b0f613871d697c4c9a7d
DIFF: https://github.com/llvm/llvm-project/commit/4c1eaf26ae70b9f0e441b0f613871d697c4c9a7d.diff

LOG: [mlir] fix the rocm runtime wrapper to account for cuda / rocm api differences

The patch adapts the rocm runtime wrapper due to subtle differences between the cuda and the rocm/hip runtime api.

Reviewed By: csigg

Differential Revision: https://reviews.llvm.org/D95027

Added: 
    

Modified: 
    mlir/tools/mlir-rocm-runner/rocm-runtime-wrappers.cpp

Removed: 
    


################################################################################
diff  --git a/mlir/tools/mlir-rocm-runner/rocm-runtime-wrappers.cpp b/mlir/tools/mlir-rocm-runner/rocm-runtime-wrappers.cpp
index 4f62f204f4a8..028e2e3b55d1 100644
--- a/mlir/tools/mlir-rocm-runner/rocm-runtime-wrappers.cpp
+++ b/mlir/tools/mlir-rocm-runner/rocm-runtime-wrappers.cpp
@@ -36,7 +36,7 @@ static auto InitializeCtx = [] {
   HIP_REPORT_IF_ERROR(hipInit(/*flags=*/0));
   hipDevice_t device;
   HIP_REPORT_IF_ERROR(hipDeviceGet(&device, /*ordinal=*/0));
-  hipContext_t context;
+  hipCtx_t context;
   HIP_REPORT_IF_ERROR(hipCtxCreate(&context, /*flags=*/0, device));
   return 0;
 }();
@@ -110,17 +110,18 @@ extern "C" void mgpuEventRecord(hipEvent_t event, hipStream_t stream) {
 
 extern "C" void *mgpuMemAlloc(uint64_t sizeBytes, hipStream_t /*stream*/) {
   void *ptr;
-  HIP_REPORT_IF_ERROR(hipMemAlloc(&ptr, sizeBytes));
+  HIP_REPORT_IF_ERROR(hipMalloc(&ptr, sizeBytes));
   return ptr;
 }
 
 extern "C" void mgpuMemFree(void *ptr, hipStream_t /*stream*/) {
-  HIP_REPORT_IF_ERROR(hipMemFree(ptr));
+  HIP_REPORT_IF_ERROR(hipFree(ptr));
 }
 
 extern "C" void mgpuMemcpy(void *dst, void *src, uint64_t sizeBytes,
                            hipStream_t stream) {
-  HIP_REPORT_IF_ERROR(hipMemcpyAsync(dst, src, sizeBytes, stream));
+  HIP_REPORT_IF_ERROR(
+      hipMemcpyAsync(dst, src, sizeBytes, hipMemcpyDefault, stream));
 }
 
 /// Helper functions for writing mlir example code