From b95d5677695c8d0c9bc1c38ae8a13e91ecb38adf Mon Sep 17 00:00:00 2001
From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com>
Date: Wed, 1 Mar 2023 20:35:51 +0530
Subject: [PATCH] SWDEV-380405 - Correct test cases. (#165)

Change-Id: I55f2e2273a19643b14baeebc8a6f19331c20b09d
---
 catch/unit/memory/hipMemcpyPeer.cc      | 14 ++++++++++----
 catch/unit/memory/hipMemcpyPeerAsync.cc | 16 +++++++++++-----
 2 files changed, 21 insertions(+), 9 deletions(-)

diff --git a/catch/unit/memory/hipMemcpyPeer.cc b/catch/unit/memory/hipMemcpyPeer.cc
index 245b3e0b2c..8c91281cf6 100644
--- a/catch/unit/memory/hipMemcpyPeer.cc
+++ b/catch/unit/memory/hipMemcpyPeer.cc
@@ -138,13 +138,19 @@ TEST_CASE("Unit_hipMemcpyPeer_Positive_ZeroSize") {
     const auto element_count = allocation_size / sizeof(*src_alloc.ptr());
     constexpr auto thread_count = 1024;
     const auto block_count = element_count / thread_count + 1;
-    constexpr int set_value = 22;
+    constexpr int set_value_s = 22;
     HIP_CHECK(hipSetDevice(src_device));
-    VectorSet<<<block_count, thread_count, 0>>>(src_alloc.ptr(), set_value, element_count);
+    VectorSet<<<block_count, thread_count, 0>>>(src_alloc.ptr(), set_value_s, element_count);
     HIP_CHECK(hipGetLastError());
 
-    constexpr int expected_value = 21;
-    std::fill_n(src_alloc.host_ptr(), element_count, expected_value);
+    constexpr int expected_value = 20;
+    HIP_CHECK(hipSetDevice(dst_device));
+    VectorSet<<<block_count, thread_count, 0>>>(dst_alloc.ptr(), expected_value, element_count);
+    HIP_CHECK(hipGetLastError());
+    HIP_CHECK(hipSetDevice(src_device));
+
+    constexpr int set_value_h = 21;
+    std::fill_n(result.host_ptr(), element_count, set_value_h);
 
     HIP_CHECK(hipMemcpyPeer(dst_alloc.ptr(), dst_device, src_alloc.ptr(), src_device, 0));
 
diff --git a/catch/unit/memory/hipMemcpyPeerAsync.cc b/catch/unit/memory/hipMemcpyPeerAsync.cc
index 90f9fdb8f4..a84ade3329 100644
--- a/catch/unit/memory/hipMemcpyPeerAsync.cc
+++ b/catch/unit/memory/hipMemcpyPeerAsync.cc
@@ -154,13 +154,19 @@ TEST_CASE("Unit_hipMemcpyPeerAsync_Positive_ZeroSize") {
     const auto element_count = allocation_size / sizeof(*src_alloc.ptr());
     constexpr auto thread_count = 1024;
     const auto block_count = element_count / thread_count + 1;
-    constexpr int set_value = 22;
+    constexpr int set_value_s = 22;
     HIP_CHECK(hipSetDevice(src_device));
-    VectorSet<<<block_count, thread_count, 0, stream>>>(src_alloc.ptr(), set_value, element_count);
+    VectorSet<<<block_count, thread_count, 0, stream>>>(src_alloc.ptr(), set_value_s, element_count);
     HIP_CHECK(hipGetLastError());
 
-    constexpr int expected_value = 21;
-    std::fill_n(src_alloc.host_ptr(), element_count, expected_value);
+    constexpr int expected_value = 20;
+    HIP_CHECK(hipSetDevice(dst_device));
+    VectorSet<<<block_count, thread_count, 0, stream>>>(dst_alloc.ptr(), expected_value, element_count);
+    HIP_CHECK(hipGetLastError());
+    HIP_CHECK(hipSetDevice(src_device));
+
+    constexpr int set_value_h = 21;
+    std::fill_n(result.host_ptr(), element_count, set_value_h);
 
     HIP_CHECK(
         hipMemcpyPeerAsync(dst_alloc.ptr(), dst_device, src_alloc.ptr(), src_device, 0, stream));
@@ -189,7 +195,7 @@ TEST_CASE("Unit_hipMemcpyPeerAsync_Negative_Parameters") {
   const hipStream_t stream = stream_guard.stream();
 
   constexpr auto InvalidStream = [] {
-    StreamGuard sg(Streams::created);
+    const StreamGuard sg(Streams::created);
     return sg.stream();
   };