diff --git a/projects/hip-tests/catch/unit/memory/hipMemcpyDtoD.cc b/projects/hip-tests/catch/unit/memory/hipMemcpyDtoD.cc index ff5eb2c90c..6eb6c5342f 100644 --- a/projects/hip-tests/catch/unit/memory/hipMemcpyDtoD.cc +++ b/projects/hip-tests/catch/unit/memory/hipMemcpyDtoD.cc @@ -50,48 +50,50 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpyDtoD_Basic", "", if (numDevices > 1) { int canAccessPeer = 0; HIP_CHECK(hipDeviceCanAccessPeer(&canAccessPeer, 0, 1)); + HIP_CHECK(hipSetDevice(0)); if (canAccessPeer) { - HIP_CHECK(hipSetDevice(0)); - HipTest::initArrays(&A_d, &B_d, &C_d, &A_h, &B_h, &C_h, - NUM_ELM, false); - HIP_CHECK(hipSetDevice(1)); - HIP_CHECK(hipMalloc(&X_d, Nbytes)); - HIP_CHECK(hipMalloc(&Y_d, Nbytes)); - HIP_CHECK(hipMalloc(&Z_d, Nbytes)); - - HIP_CHECK(hipSetDevice(0)); - HIP_CHECK(hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice)); - HIP_CHECK(hipMemcpy(B_d, B_h, Nbytes, hipMemcpyHostToDevice)); - hipLaunchKernelGGL(HipTest::vectorADD, dim3(1), - dim3(1), 0, 0, - static_cast(A_d), - static_cast(B_d), C_d, NUM_ELM); - HIP_CHECK(hipGetLastError()); - HIP_CHECK(hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost)); - HIP_CHECK(hipDeviceSynchronize()); - HipTest::checkVectorADD(A_h, B_h, C_h, NUM_ELM); - - HIP_CHECK(hipSetDevice(1)); - HIP_CHECK(hipMemcpyDtoD((hipDeviceptr_t)X_d, (hipDeviceptr_t)A_d, - Nbytes)); - HIP_CHECK(hipMemcpyDtoD((hipDeviceptr_t)Y_d, (hipDeviceptr_t)B_d, - Nbytes)); - - hipLaunchKernelGGL(HipTest::vectorADD, dim3(1), - dim3(1), 0, 0, - static_cast(X_d), - static_cast(Y_d), Z_d, NUM_ELM); - HIP_CHECK(hipGetLastError()); - HIP_CHECK(hipMemcpyDtoH(C_h, (hipDeviceptr_t)Z_d, Nbytes)); - HIP_CHECK(hipDeviceSynchronize()); - HipTest::checkVectorADD(A_h, B_h, C_h, NUM_ELM); - - HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); - HIP_CHECK(hipFree(X_d)); - HIP_CHECK(hipFree(Y_d)); - HIP_CHECK(hipFree(Z_d)); - } else { - SUCCEED("Machine does not seem to have P2P Capabilities"); + HIP_CHECK(hipDeviceEnablePeerAccess(1, 0)); } + else { + INFO("Machine does not have P2P Capabilities"); + } + HipTest::initArrays(&A_d, &B_d, &C_d, &A_h, &B_h, &C_h, + NUM_ELM, false); + HIP_CHECK(hipSetDevice(1)); + HIP_CHECK(hipMalloc(&X_d, Nbytes)); + HIP_CHECK(hipMalloc(&Y_d, Nbytes)); + HIP_CHECK(hipMalloc(&Z_d, Nbytes)); + + HIP_CHECK(hipSetDevice(0)); + HIP_CHECK(hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(B_d, B_h, Nbytes, hipMemcpyHostToDevice)); + hipLaunchKernelGGL(HipTest::vectorADD, dim3(1), + dim3(1), 0, 0, + static_cast(A_d), + static_cast(B_d), C_d, NUM_ELM); + HIP_CHECK(hipGetLastError()); + HIP_CHECK(hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost)); + HIP_CHECK(hipDeviceSynchronize()); + HipTest::checkVectorADD(A_h, B_h, C_h, NUM_ELM); + + HIP_CHECK(hipSetDevice(1)); + HIP_CHECK(hipMemcpyDtoD((hipDeviceptr_t)X_d, (hipDeviceptr_t)A_d, + Nbytes)); + HIP_CHECK(hipMemcpyDtoD((hipDeviceptr_t)Y_d, (hipDeviceptr_t)B_d, + Nbytes)); + + hipLaunchKernelGGL(HipTest::vectorADD, dim3(1), + dim3(1), 0, 0, + static_cast(X_d), + static_cast(Y_d), Z_d, NUM_ELM); + HIP_CHECK(hipGetLastError()); + HIP_CHECK(hipMemcpyDtoH(C_h, (hipDeviceptr_t)Z_d, Nbytes)); + HIP_CHECK(hipDeviceSynchronize()); + HipTest::checkVectorADD(A_h, B_h, C_h, NUM_ELM); + + HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); + HIP_CHECK(hipFree(X_d)); + HIP_CHECK(hipFree(Y_d)); + HIP_CHECK(hipFree(Z_d)); } } diff --git a/projects/hip-tests/catch/unit/memory/hipMemcpyDtoDAsync.cc b/projects/hip-tests/catch/unit/memory/hipMemcpyDtoDAsync.cc index 1cd8581d4b..312e6f8339 100644 --- a/projects/hip-tests/catch/unit/memory/hipMemcpyDtoDAsync.cc +++ b/projects/hip-tests/catch/unit/memory/hipMemcpyDtoDAsync.cc @@ -52,52 +52,54 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpyDtoDAsync_Basic", "", if (numDevices > 1) { int canAccessPeer = 0; HIP_CHECK(hipDeviceCanAccessPeer(&canAccessPeer, 0, 1)); + HIP_CHECK(hipSetDevice(0)); if (canAccessPeer) { - HIP_CHECK(hipSetDevice(0)); - HipTest::initArrays(&A_d, &B_d, &C_d, &A_h, &B_h, &C_h, - NUM_ELM, false); - HIP_CHECK(hipSetDevice(1)); - HIP_CHECK(hipMalloc(&X_d, Nbytes)); - HIP_CHECK(hipMalloc(&Y_d, Nbytes)); - HIP_CHECK(hipMalloc(&Z_d, Nbytes)); - - HIP_CHECK(hipSetDevice(0)); - HIP_CHECK(hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice)); - HIP_CHECK(hipMemcpy(B_d, B_h, Nbytes, hipMemcpyHostToDevice)); - hipLaunchKernelGGL(HipTest::vectorADD, dim3(1), - dim3(1), 0, 0, - static_cast(A_d), - static_cast(B_d), C_d, NUM_ELM); - HIP_CHECK(hipGetLastError()); - HIP_CHECK(hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost)); - HIP_CHECK(hipDeviceSynchronize()); - HipTest::checkVectorADD(A_h, B_h, C_h, NUM_ELM); - - HIP_CHECK(hipSetDevice(1)); - HIP_CHECK(hipStreamCreate(&stream)); - HIP_CHECK(hipMemcpyDtoDAsync((hipDeviceptr_t)X_d, (hipDeviceptr_t)A_d, - Nbytes, stream)); - HIP_CHECK(hipMemcpyDtoDAsync((hipDeviceptr_t)Y_d, (hipDeviceptr_t)B_d, - Nbytes, stream)); - HIP_CHECK(hipStreamSynchronize(stream)); - - hipLaunchKernelGGL(HipTest::vectorADD, dim3(1), - dim3(1), 0, 0, - static_cast(X_d), - static_cast(Y_d), Z_d, NUM_ELM); - HIP_CHECK(hipGetLastError()); - HIP_CHECK(hipMemcpyDtoHAsync(C_h, (hipDeviceptr_t)Z_d, Nbytes, stream)); - HIP_CHECK(hipStreamSynchronize(stream)); - HIP_CHECK(hipDeviceSynchronize()); - HipTest::checkVectorADD(A_h, B_h, C_h, NUM_ELM); - - HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); - HIP_CHECK(hipFree(X_d)); - HIP_CHECK(hipFree(Y_d)); - HIP_CHECK(hipFree(Z_d)); - } else { - SUCCEED("Machine does not seem to have P2P Capabilities"); + HIP_CHECK(hipDeviceEnablePeerAccess(1, 0)); } + else { + INFO("Machine does not have P2P Capabilities"); + } + HipTest::initArrays(&A_d, &B_d, &C_d, &A_h, &B_h, &C_h, + NUM_ELM, false); + HIP_CHECK(hipSetDevice(1)); + HIP_CHECK(hipMalloc(&X_d, Nbytes)); + HIP_CHECK(hipMalloc(&Y_d, Nbytes)); + HIP_CHECK(hipMalloc(&Z_d, Nbytes)); + + HIP_CHECK(hipSetDevice(0)); + HIP_CHECK(hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(B_d, B_h, Nbytes, hipMemcpyHostToDevice)); + hipLaunchKernelGGL(HipTest::vectorADD, dim3(1), + dim3(1), 0, 0, + static_cast(A_d), + static_cast(B_d), C_d, NUM_ELM); + HIP_CHECK(hipGetLastError()); + HIP_CHECK(hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost)); + HIP_CHECK(hipDeviceSynchronize()); + HipTest::checkVectorADD(A_h, B_h, C_h, NUM_ELM); + + HIP_CHECK(hipSetDevice(1)); + HIP_CHECK(hipStreamCreate(&stream)); + HIP_CHECK(hipMemcpyDtoDAsync((hipDeviceptr_t)X_d, (hipDeviceptr_t)A_d, + Nbytes, stream)); + HIP_CHECK(hipMemcpyDtoDAsync((hipDeviceptr_t)Y_d, (hipDeviceptr_t)B_d, + Nbytes, stream)); + HIP_CHECK(hipStreamSynchronize(stream)); + + hipLaunchKernelGGL(HipTest::vectorADD, dim3(1), + dim3(1), 0, 0, + static_cast(X_d), + static_cast(Y_d), Z_d, NUM_ELM); + HIP_CHECK(hipGetLastError()); + HIP_CHECK(hipMemcpyDtoHAsync(C_h, (hipDeviceptr_t)Z_d, Nbytes, stream)); + HIP_CHECK(hipStreamSynchronize(stream)); + HIP_CHECK(hipDeviceSynchronize()); + HipTest::checkVectorADD(A_h, B_h, C_h, NUM_ELM); + + HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); + HIP_CHECK(hipFree(X_d)); + HIP_CHECK(hipFree(Y_d)); + HIP_CHECK(hipFree(Z_d)); } }