From ff29b47bb814d8ba04636f921d3dd444df1e0aed Mon Sep 17 00:00:00 2001
From: Anusha GodavarthySurya <anusha.godavarthysurya@amd.com>
Date: Wed, 27 Mar 2024 12:19:06 +0000
Subject: [PATCH] SWDEV-452787 - correct hipDrvGraphAddMemcpyNode check

Change-Id: Id58f982edd4f17d675f7a0f61a9b4dea0baebd9b


[ROCm/clr commit: ea4f09e8c071ef02c65e819a56593765a05d0d07]
---
 projects/clr/hipamd/src/hip_graph.cpp        |  2 +-
 projects/clr/hipamd/src/hip_graph_helper.hpp |  2 -
 projects/clr/hipamd/src/hip_memory.cpp       | 47 --------------------
 3 files changed, 1 insertion(+), 50 deletions(-)

diff --git a/projects/clr/hipamd/src/hip_graph.cpp b/projects/clr/hipamd/src/hip_graph.cpp
index 7481b3b7c0..edaec142c0 100644
--- a/projects/clr/hipamd/src/hip_graph.cpp
+++ b/projects/clr/hipamd/src/hip_graph.cpp
@@ -137,7 +137,7 @@ hipError_t ihipDrvGraphAddMemcpyNode(hip::GraphNode** pGraphNode, hip::Graph* gr
       (numDependencies > 0 && pDependencies == nullptr) || pCopyParams == nullptr) {
     return hipErrorInvalidValue;
   }
-  hipError_t status = ihipDrvMemcpy3DParamValidate(pCopyParams);
+  hipError_t status = ihipDrvMemcpy3D_validate(pCopyParams);
   if (status != hipSuccess) {
     return status;
   }
diff --git a/projects/clr/hipamd/src/hip_graph_helper.hpp b/projects/clr/hipamd/src/hip_graph_helper.hpp
index e77954d3d8..c146621185 100644
--- a/projects/clr/hipamd/src/hip_graph_helper.hpp
+++ b/projects/clr/hipamd/src/hip_graph_helper.hpp
@@ -27,8 +27,6 @@ hipError_t ihipMemcpy3D_validate(const hipMemcpy3DParms* p);
 
 hipError_t ihipDrvMemcpy3D_validate(const HIP_MEMCPY3D* pCopy);
 
-hipError_t ihipDrvMemcpy3DParamValidate(const HIP_MEMCPY3D* pCopy);
-
 hipError_t ihipMemcpy_validate(void* dst, const void* src, size_t sizeBytes, hipMemcpyKind kind);
 
 hipError_t ihipMemcpyCommand(amd::Command*& command, void* dst, const void* src, size_t sizeBytes,
diff --git a/projects/clr/hipamd/src/hip_memory.cpp b/projects/clr/hipamd/src/hip_memory.cpp
index c95cdefe93..5c3eaffda1 100644
--- a/projects/clr/hipamd/src/hip_memory.cpp
+++ b/projects/clr/hipamd/src/hip_memory.cpp
@@ -2882,53 +2882,6 @@ hipError_t ihipMemcpy3D_validate(const hipMemcpy3DParms* p) {
   return hipSuccess;
 }
 
-hipError_t ihipDrvMemcpy3DParamValidate(const HIP_MEMCPY3D* p) {
-  // Passing more than one non-zero source or destination will cause hipMemcpy3D() to
-  // return an error.
-  if (p == nullptr || ((p->srcArray != nullptr) && (p->srcHost != nullptr)) ||
-      ((p->dstArray != nullptr) && (p->dstHost != nullptr))) {
-    return hipErrorInvalidValue;
-  }
-  // The struct passed to hipMemcpy3D() must specify one of srcArray or srcPtr and one of dstArray
-  // or dstPtr.
-  if (((p->srcArray == nullptr) && (p->srcHost == nullptr)) ||
-      ((p->dstArray == nullptr) && (p->dstHost == nullptr))) {
-    return hipErrorInvalidValue;
-  }
-
-  // If the source and destination are both arrays, hipMemcpy3D() will return an error if they do
-  // not have the same element size.
-  if (((p->srcArray != nullptr) && (p->dstArray != nullptr)) &&
-      (hip::getElementSize(p->srcArray) != hip::getElementSize(p->dstArray))) {
-    return hipErrorInvalidValue;
-  }
-
-  // dst/src pitch must be less than max pitch
-  auto* deviceHandle = g_devices[hip::getCurrentDevice()->deviceId()]->devices()[0];
-  const auto& info = deviceHandle->info();
-  constexpr auto int32_max = static_cast<uint64_t>(std::numeric_limits<int32_t>::max());
-  auto maxPitch = std::min(info.maxMemAllocSize_, int32_max);
-
-  // negative pitch cases
-  if (p->srcPitch >= maxPitch || p->dstPitch >= maxPitch) {
-    return hipErrorInvalidValue;
-  }
-
-  if (p->dstArray == nullptr && p->srcArray == nullptr) {
-    if ((p->WidthInBytes + p->srcXInBytes > p->srcPitch) ||
-        (p->WidthInBytes + p->dstXInBytes > p->dstPitch)) {
-      return hipErrorInvalidValue;
-    }
-  }
-  if (p->srcMemoryType < hipMemoryTypeHost || p->srcMemoryType > hipMemoryTypeManaged) {
-    return hipErrorInvalidMemcpyDirection;
-  }
-  if (p->dstMemoryType < hipMemoryTypeHost || p->dstMemoryType > hipMemoryTypeManaged) {
-    return hipErrorInvalidMemcpyDirection;
-  }
-  return hipSuccess;
-}
-
 hipError_t ihipDrvMemcpy3D_validate(const HIP_MEMCPY3D* pCopy) {
   hipError_t status;
   if (pCopy->WidthInBytes == 0 || pCopy->Height == 0 || pCopy->Depth == 0) {