Merge branch 'privatestaging' of https://github.com/AMDComputeLibraries/HIP-privatestaging into privatestaging

src/hip_hcc.cpp

@@ -797,7 +797,7 @@ hipError_t ihipDevice_t::getProperties(hipDeviceProp_t* prop)
     // Group memory will not be paged out, so, the physical memory size is the total shared memory size, and also equal to the group region size.
     prop->maxSharedMemoryPerMultiProcessor = prop->totalGlobalMem;
 
-#ifdef USE_ROCR_V2
+#if USE_ROCR_V2
     // Get Max memory clock frequency
     //err = hsa_region_get_info(*am_region, (hsa_region_info_t)HSA_AMD_REGION_INFO_MAX_CLOCK_FREQUENCY, &prop->memoryClockRate);
     DeviceErrorCheck(err);
@@ -1330,12 +1330,10 @@ hipError_t hipDeviceGetAttribute(int* pi, hipDeviceAttribute_t attr, int device)
             *pi = prop->regsPerBlock; break;
         case hipDeviceAttributeClockRate:
             *pi = prop->clockRate; break;
-#ifdef USE_ROCR_V2
         case hipDeviceAttributeMemoryClockRate:
             *pi = prop->memoryClockRate; break;
         case hipDeviceAttributeMemoryBusWidth:
             *pi = prop->memoryBusWidth; break;
-#endif
         case hipDeviceAttributeMultiprocessorCount:
             *pi = prop->multiProcessorCount; break;
         case hipDeviceAttributeComputeMode:
@@ -1798,7 +1796,12 @@ hipError_t hipPointerGetAttributes(hipPointerAttribute_t *attributes, void* ptr)
         attributes->hostPointer   = amPointerInfo._hostPointer;
         attributes->devicePointer = amPointerInfo._devicePointer;
         attributes->isManaged     = 0;
-
+        if(attributes->memoryType == hipMemoryTypeHost){
+            attributes->hostPointer = ptr;
+        }
+        if(attributes->memoryType == hipMemoryTypeDevice){
+            attributes->devicePointer = ptr;
+        }
         attributes->allocationFlags = amPointerInfo._appAllocationFlags;
         attributes->device          = amPointerInfo._appId;
 
@@ -2370,6 +2373,7 @@ hipError_t hipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKind
 #endif
 /**
  * @result #hipSuccess, #hipErrorInvalidDevice, #hipErrorInvalidMemcpyDirection, #hipErrorInvalidValue
+ * @warning on HCC hipMemcpyAsync does not support overlapped H2D and D2H copies.
  */
 //---
 hipError_t hipMemcpyAsync(void* dst, const void* src, size_t sizeBytes, hipMemcpyKind kind, hipStream_t stream)