From db9d9faaeb98fb3682ad132203d337fc5e6cff6c Mon Sep 17 00:00:00 2001
From: Ioannis Assiouras <Ioannis.Assiouras@amd.com>
Date: Tue, 21 Mar 2023 11:27:49 +0000
Subject: [PATCH] SWDEV-389611 - Changed hipMallocAsync to return success for
 zero size allocation to match hipMalloc

Change-Id: Ia2cbfaf2df9be37b62f23b2bb5b619205752ba84
---
 hipamd/src/hip_mempool.cpp | 12 ++++++++++--
 1 file changed, 10 insertions(+), 2 deletions(-)

diff --git a/hipamd/src/hip_mempool.cpp b/hipamd/src/hip_mempool.cpp
index f798f8c813..cdbe1929b3 100644
--- a/hipamd/src/hip_mempool.cpp
+++ b/hipamd/src/hip_mempool.cpp
@@ -67,9 +67,13 @@ hipError_t hipDeviceGetMemPool(hipMemPool_t* mem_pool, int device) {
 // ================================================================================================
 hipError_t hipMallocAsync(void** dev_ptr, size_t size, hipStream_t stream) {
   HIP_INIT_API(hipMallocAsync, dev_ptr, size, stream);
-  if ((dev_ptr == nullptr) || (size == 0) || (!hip::isValid(stream))) {
+  if ((dev_ptr == nullptr) || (!hip::isValid(stream))) {
     HIP_RETURN(hipErrorInvalidValue);
   }
+  if (size == 0) {
+    *dev_ptr = nullptr;
+    HIP_RETURN(hipSuccess);
+  }
   auto hip_stream = (stream == nullptr) ? hip::getCurrentDevice()->NullStream() :
     reinterpret_cast<hip::Stream*>(stream);
   auto device = hip_stream->GetDevice();
@@ -235,9 +239,13 @@ hipError_t hipMallocFromPoolAsync(
     hipMemPool_t mem_pool,
     hipStream_t stream) {
   HIP_INIT_API(hipMallocFromPoolAsync, dev_ptr, size, mem_pool, stream);
-  if ((dev_ptr == nullptr) || (size == 0) || (mem_pool == nullptr) || (!hip::isValid(stream))) {
+  if ((dev_ptr == nullptr) || (mem_pool == nullptr) || (!hip::isValid(stream))) {
     HIP_RETURN(hipErrorInvalidValue);
   }
+  if (size == 0) {
+    *dev_ptr = nullptr;
+    HIP_RETURN(hipSuccess);
+  }
   STREAM_CAPTURE(hipMallocAsync, stream, mem_pool, size, dev_ptr);
 
   auto mpool = reinterpret_cast<hip::MemoryPool*>(mem_pool);