SWDEV-369556 - Unit_hipMemGetAddressRange_Negative cases

- return value for invalid device ptr
- offset greater than size

Change-Id: I8bcf878c3c1c0a29836ea664bfb9048e1114b74e


[ROCm/clr commit: c6a7ce66bb]

projects/clr/hipamd/src/hip_internal.hpp

@@ -547,7 +547,7 @@ extern int ihipGetDevice();
 
 extern hipError_t ihipMalloc(void** ptr, size_t sizeBytes, unsigned int flags);
 extern amd::Memory* getMemoryObject(const void* ptr, size_t& offset, size_t size = 0);
-extern amd::Memory* getMemoryObjectWithOffset(const void* ptr, const size_t size);
+extern amd::Memory* getMemoryObjectWithOffset(const void* ptr, const size_t size = 0);
 extern void getStreamPerThread(hipStream_t& stream);
 extern hipStream_t getPerThreadDefaultStream();
 extern hipError_t ihipUnbindTexture(textureReference* texRef);

projects/clr/hipamd/src/hip_memory.cpp

@@ -68,7 +68,7 @@ amd::Memory* getMemoryObject(const void* ptr, size_t& offset, size_t size) {
 
 // ================================================================================================
 amd::Memory* getMemoryObjectWithOffset(const void* ptr, const size_t size) {
-  size_t offset;
+  size_t offset = 0;
   amd::Memory* memObj = getMemoryObject(ptr, offset);
 
   if (memObj != nullptr) {
@@ -683,11 +683,9 @@ hipError_t hipMemGetAddressRange(hipDeviceptr_t* pbase, size_t* psize, hipDevice
 
   // Since we are using SVM buffer DevicePtr and HostPtr is the same
   void* ptr = dptr;
-  size_t offset = 0;
-  amd::Memory* svmMem = getMemoryObject(ptr, offset);
-
+  amd::Memory* svmMem = getMemoryObjectWithOffset(ptr);
   if (svmMem == nullptr) {
-    HIP_RETURN(hipErrorInvalidDevicePointer);
+    HIP_RETURN(hipErrorNotFound);
   }
 
   *pbase = svmMem->getSvmPtr();