Client: don't use hipDeviceSynchronize for blocking CPU to get CPU timer

ROCm · May 3, 2024 · 69d6262 · 69d6262
1 parent fde584b
commit 69d6262
Show file tree

Hide file tree

Showing 2 changed files with 2 additions and 11 deletions.
diff --git a/clients/common/utility.cpp b/clients/common/utility.cpp
@@ -119,9 +119,9 @@ double get_time_us_sync_device(void)
 /*! \brief  CPU Timer(in microsecond): synchronize with given queue/stream and return wall time */
 double get_time_us_sync(hipStream_t stream)
 {
-    if(hipDeviceSynchronize() != hipSuccess)
+    if(hipStreamSynchronize(stream) != hipSuccess)
     {
-        hipblaslt_cerr << "Synchronizing device failed" << std::endl;
+        hipblaslt_cerr << "Synchronizing stream failed" << std::endl;
     }
 
     auto now = std::chrono::steady_clock::now();

diff --git a/clients/include/testing_matmul.hpp b/clients/include/testing_matmul.hpp
@@ -2470,7 +2470,6 @@ void testing_matmul_with_bias(const Arguments& arg)
                 CHECK_HIP_ERROR(hipEventRecord(event_gpu_time_start, stream));
             else
             {
-                CHECK_HIP_ERROR(hipStreamSynchronize(stream));
                 flush_time_used = get_time_us_sync(stream);
             }
             for(int i = 0; i < flush_iter; i++)
@@ -2486,7 +2485,6 @@ void testing_matmul_with_bias(const Arguments& arg)
             }
             else
             {
-                CHECK_HIP_ERROR(hipStreamSynchronize(stream));
                 flush_time_used = get_time_us_sync(stream) - flush_time_used;
             }
             flush_time_used /= flush_iter;
@@ -2513,7 +2511,6 @@ void testing_matmul_with_bias(const Arguments& arg)
                         CHECK_HIP_ERROR(hipEventRecord(event_gpu_time_start, stream));
                     else
                     {
-                        CHECK_HIP_ERROR(hipStreamSynchronize(stream));
                         gpu_time_used = get_time_us_sync(stream);
                     }
 
@@ -2562,7 +2559,6 @@ void testing_matmul_with_bias(const Arguments& arg)
                         CHECK_HIP_ERROR(hipEventRecord(event_gpu_time_start, stream));
                     else
                     {
-                        CHECK_HIP_ERROR(hipStreamSynchronize(stream));
                         gpu_time_used = get_time_us_sync(stream);
                     }
                     for(int i = 0; i < number_hot_calls; i++)
@@ -2608,7 +2604,6 @@ void testing_matmul_with_bias(const Arguments& arg)
                 }
                 else
                 {
-                    CHECK_HIP_ERROR(hipStreamSynchronize(stream));
                     gpu_time_used = get_time_us_sync(stream) - gpu_time_used;
                 }
             }
@@ -2645,7 +2640,6 @@ void testing_matmul_with_bias(const Arguments& arg)
                         CHECK_HIP_ERROR(hipEventRecord(event_gpu_time_start, stream));
                     else
                     {
-                        CHECK_HIP_ERROR(hipStreamSynchronize(stream));
                         gpu_time_used = get_time_us_sync(stream);
                     }
 
@@ -2664,7 +2658,6 @@ void testing_matmul_with_bias(const Arguments& arg)
                     }
                     else
                     {
-                        CHECK_HIP_ERROR(hipStreamSynchronize(stream));
                         gpu_time_used = get_time_us_sync(stream) - gpu_time_used;
                     }
                 }
@@ -2689,7 +2682,6 @@ void testing_matmul_with_bias(const Arguments& arg)
                         CHECK_HIP_ERROR(hipEventRecord(event_gpu_time_start, stream));
                     else
                     {
-                        CHECK_HIP_ERROR(hipStreamSynchronize(stream));
                         gpu_time_used = get_time_us_sync(stream);
                     }
 
@@ -2707,7 +2699,6 @@ void testing_matmul_with_bias(const Arguments& arg)
                     }
                     else
                     {
-                        CHECK_HIP_ERROR(hipStreamSynchronize(stream));
                         gpu_time_used = get_time_us_sync(stream) - gpu_time_used;
                     }
                 }