From 97fc90c58fdf44d264e61c9c1ea5ab95d68c0dc5 Mon Sep 17 00:00:00 2001 From: Ioannis Assiouras <38722728+iassiour@users.noreply.github.com> Date: Tue, 23 Sep 2025 14:14:51 +0100 Subject: [PATCH] SWDEV-556250 - Added synchronization before validating the result in Unit_hipStreamLegacy* tests (#1062) --- .../catch/unit/stream/hipStreamLegacy_Ext.cc | 13 ++++++++++++- .../unit/stream/hipStreamLegacy_compiler_options.cc | 2 ++ .../catch/unit/stream/hipStreamLegacy_exe.cc | 1 + 3 files changed, 15 insertions(+), 1 deletion(-) diff --git a/projects/hip-tests/catch/unit/stream/hipStreamLegacy_Ext.cc b/projects/hip-tests/catch/unit/stream/hipStreamLegacy_Ext.cc index fcfbabfe3f..abde78deb3 100644 --- a/projects/hip-tests/catch/unit/stream/hipStreamLegacy_Ext.cc +++ b/projects/hip-tests/catch/unit/stream/hipStreamLegacy_Ext.cc @@ -95,6 +95,7 @@ TEST_CASE("Unit_hipStreamLegacy_WithBlockingStream") { HIP_CHECK(hipMemcpyAsync(devArr, hostArrSrc, NBYTES, hipMemcpyHostToDevice, stream)); HIP_CHECK(hipMemcpyAsync(hostArrDst, devArr, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] << " Expected value : 1 \n"); @@ -129,6 +130,7 @@ void launchFunction(hipStream_t stream) { HIP_CHECK_THREAD(hipMemcpyAsync(devArr, hostArrSrc, NBYTES, hipMemcpyHostToDevice, stream)); HIP_CHECK_THREAD(hipMemcpyAsync(hostArrDst, devArr, NBYTES, hipMemcpyDeviceToHost, stream)); + HIP_CHECK_THREAD(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { if (hostArrDst[i] != 5) { @@ -223,6 +225,7 @@ TEST_CASE("Unit_hipStreamLegacy_WithNonBlockingStream") { HIP_CHECK(hipMemcpyAsync(devArr, hostArrSrc, NBYTES, hipMemcpyHostToDevice, stream)); HIP_CHECK(hipMemcpyAsync(hostArrDst, devArr, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] << " Expected value : 10 or 11 \n"); @@ -265,6 +268,7 @@ TEST_CASE("Unit_hipStreamLegacy_WithStreamPerThread") { HIP_CHECK(hipMemcpyAsync(devArr, hostArrSrc, NBYTES, hipMemcpyHostToDevice, hipStreamPerThread)); HIP_CHECK(hipMemcpyAsync(hostArrDst, devArr, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] << " Expected value : 15 \n"); @@ -317,6 +321,7 @@ TEST_CASE("Unit_hipStreamLegacy_MultiDevice") { HIP_CHECK(hipMemcpyAsync(devArr, hostArrSrc, NBYTES, hipMemcpyHostToDevice, hipStreamLegacy)); HIP_CHECK(hipMemcpyAsync(hostArrDst, devArr, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] @@ -377,6 +382,7 @@ TEST_CASE("Unit_hipStreamLegacy_H2H_H2D_D2D_D2H_Default") { HIP_CHECK(hipMemcpyAsync(devArr2, devArr1, NBYTES, hipMemcpyDeviceToDevice, hipStreamLegacy)); HIP_CHECK(hipMemcpyAsync(hostArr3, devArr2, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); HIP_CHECK(hipMemcpyAsync(hostArr4, hostArr3, NBYTES, hipMemcpyDefault, hipStreamLegacy)); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArr4[i] << " Expected value : 30 \n"); @@ -481,6 +487,7 @@ TEST_CASE("Unit_hipStreamLegacy_MultiDeviceMultiOperation") { // Finally copy daat to hostArr4 HIP_CHECK(hipSetDevice(currentDevice)); HIP_CHECK(hipMemcpyAsync(h2Dev0, d2Dev0, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << h2Dev0[i] @@ -561,6 +568,7 @@ TEST_CASE("Unit_hipStreamLegacy_TwoThreadsEachOneDiffOperation") { std::thread D2H_Thread(copyFromDeviceToHost, devArr, hostArrDst); D2H_Thread.join(); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] << " Expected value : 50 \n"); @@ -623,7 +631,8 @@ TEST_CASE("Unit_hipStreamLegacy_TwoDevicesEachOneDiffOperation") { HIP_CHECK(hipSetDevice(1)); HIP_CHECK(hipMemcpyAsync(hostArrDst, devArrDev1, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); - + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); + for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] << " Expected value : 500 or 501 \n"); @@ -704,6 +713,7 @@ TEST_CASE("Unit_hipStreamLegacy_TwoThreadsInTwoDevicesEachOneDiffOperation") { dev0Thread.join(); std::thread dev1Thread(operationsInDev1, devArrDev1, hostArrDst); dev1Thread.join(); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] @@ -765,6 +775,7 @@ TEST_CASE("Unit_hipStreamLegacy_WithKernel") { addOneKernel<<<1, 1, 0, hipStreamLegacy>>>(devArr, N); addOneKernel<<<1, 1, 0, hipStreamLegacy>>>(devArr, N); HIP_CHECK(hipMemcpyAsync(hostArrDst, devArr, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] << " Expected value : 3 \n"); diff --git a/projects/hip-tests/catch/unit/stream/hipStreamLegacy_compiler_options.cc b/projects/hip-tests/catch/unit/stream/hipStreamLegacy_compiler_options.cc index 0f8e0b72e2..7fb023f19e 100644 --- a/projects/hip-tests/catch/unit/stream/hipStreamLegacy_compiler_options.cc +++ b/projects/hip-tests/catch/unit/stream/hipStreamLegacy_compiler_options.cc @@ -64,6 +64,7 @@ TEST_CASE("Unit_hipStreamLegacy_WithSptCompilerOption") { HIP_CHECK(hipMemcpyAsync(devArr, hostArrSrc, NBYTES, hipMemcpyHostToDevice, hipStreamLegacy)); HIP_CHECK(hipMemcpyAsync(hostArrDst, devArr, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] << " Expected value : 1 \n"); @@ -133,6 +134,7 @@ TEST_CASE("Unit_hipStreamLegacy_TwoThreadsDiffOperationWithSptCompOption") { H2D_Thread.join(); std::thread D2H_Thread(copyDeviceToHost, devArr, hostArrDst); D2H_Thread.join(); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { INFO("At index : " << i << " Got value : " << hostArrDst[i] << " Expected value : 50 \n"); diff --git a/projects/hip-tests/catch/unit/stream/hipStreamLegacy_exe.cc b/projects/hip-tests/catch/unit/stream/hipStreamLegacy_exe.cc index d2a7a9996f..d5d7ad7cfb 100644 --- a/projects/hip-tests/catch/unit/stream/hipStreamLegacy_exe.cc +++ b/projects/hip-tests/catch/unit/stream/hipStreamLegacy_exe.cc @@ -79,6 +79,7 @@ int main() { HIP_CHECK(hipMemcpyAsync(devArr1, hostArr2, NBYTES, hipMemcpyHostToDevice, hipStreamLegacy)); HIP_CHECK(hipMemcpyAsync(devArr2, devArr1, NBYTES, hipMemcpyDeviceToDevice, hipStreamLegacy)); HIP_CHECK(hipMemcpyAsync(hostArr3, devArr2, NBYTES, hipMemcpyDeviceToHost, hipStreamLegacy)); + HIP_CHECK(hipStreamSynchronize(hipStreamLegacy)); for (int i = 0; i < N; i++) { if (hostArr3[i] != elementVal) {