fix nvcc compiler

- MallocHost and FreeHost deprecation.
- Change tests to call new hipHost* equivs.
- Add missing StreamSynchronize.


[ROCm/clr commit: 6984f24d3d]

projects/clr/hipamd/include/nvcc_detail/hip_runtime_api.h

@@ -119,7 +119,9 @@ inline static hipError_t hipFree(void* ptr) {
     return hipCUDAErrorTohipError(cudaFree(ptr));
 }
 
-inline static hipError_t hipMallocHost(void** ptr, size_t size) __attribute__((deprecated("use hipHostAlloc instead"))) {
+inline static hipError_t hipMallocHost(void** ptr, size_t size) __attribute__((deprecated("use hipHostAlloc instead")));
+
+inline static hipError_t hipMallocHost(void** ptr, size_t size) {
     return hipCUDAErrorTohipError(cudaMallocHost(ptr, size));
 }
 
@@ -143,7 +145,8 @@ inline static hipError_t hipHostUnregister(void* ptr){
 	return hipCUDAErrorTohipError(cudaHostUnregister(ptr));
 }
 
-inline static hipError_t hipFreeHost(void* ptr) __attribute__((deprecated("use hipHostFree instead")))  {
+inline static hipError_t hipFreeHost(void* ptr) __attribute__((deprecated("use hipHostFree instead")));
+inline static hipError_t hipFreeHost(void* ptr) {
     return hipCUDAErrorTohipError(cudaFreeHost(ptr));
 }
 
@@ -374,6 +377,10 @@ inline static hipError_t hipStreamCreate(hipStream_t *stream)
     return hipCUDAErrorTohipError(cudaStreamCreate(stream));
 }
 
+inline static hipError_t hipStreamSynchronize(hipStream_t stream) 
+{
+    return hipCUDAErrorTohipError(cudaStreamSynchronize(stream));
+}
 
 inline static hipError_t hipStreamDestroy(hipStream_t stream) 
 {

projects/clr/hipamd/tests/src/hipMemcpyAsync.cpp

@@ -159,7 +159,7 @@ void test_pingpong(hipStream_t stream, size_t numElements, int numInflight, int
     }
 
 
-    HIPCHECK(hipFreeHost(A_h));
+    HIPCHECK(hipHostFree(A_h));
     HIPCHECK(hipFree(A_d));
 }
 
@@ -216,8 +216,8 @@ void test_manyInflightCopies(hipStream_t stream, int numElements, int numCopies,
     }
 
 
-    HIPCHECK(hipFreeHost(A_h1));
-    HIPCHECK(hipFreeHost(A_h2));
+    HIPCHECK(hipHostFree(A_h1));
+    HIPCHECK(hipHostFree(A_h2));
     HIPCHECK(hipFree(A_d));
 }
 

projects/clr/hipamd/tests/src/hipMemcpy_simple.cpp

@@ -97,8 +97,8 @@ void simpleTest2(size_t numElements, bool usePinnedHost)
 
     HIPCHECK(hipFree(A_d));
     if (usePinnedHost) {
-        HIPCHECK(hipFreeHost(A_h1));
-        HIPCHECK(hipFreeHost(A_h2));
+        HIPCHECK(hipHostFree(A_h1));
+        HIPCHECK(hipHostFree(A_h2));
     } else {
         free(A_h1);
         free(A_h2);

projects/clr/hipamd/tests/src/hipPointerAttrib.cpp

@@ -178,7 +178,7 @@ void testSimple()
     HIPASSERT((char*)attribs.hostPointer+Nbytes/2 == (char*)attribs2.hostPointer);
 
 
-    hipFreeHost(A_Pinned_h);
+    hipHostFree(A_Pinned_h);
     e = hipPointerGetAttributes(&attribs, A_Pinned_h);
     HIPASSERT(e == hipErrorUnknown); // Just freed the pointer, this should return an error.
     printf("getAttr:%-20s err=%d (%s), neg-test expected\n", "A_d+NBytes", e, hipGetErrorString(e));
@@ -322,7 +322,7 @@ void clusterAllocs(int numAllocs, size_t minSize, size_t maxSize)
         if (ref._attrib.memoryType == hipMemoryTypeDevice) {
             hipFree(ref._pointer);
         } else {
-            hipFreeHost(ref._pointer);
+            hipHostFree(ref._pointer);
         }
 
     }

projects/clr/hipamd/tests/src/test_common.h

@@ -193,13 +193,13 @@ void freeArrays(T *A_d, T *B_d, T *C_d,
 
     if (usePinnedHost) {
         if (A_h) {
-            HIPCHECK (hipFreeHost(A_h));
+            HIPCHECK (hipHostFree(A_h));
         }
         if (B_h) {
-            HIPCHECK (hipFreeHost(B_h));
+            HIPCHECK (hipHostFree(B_h));
         }
         if (C_h) {
-            HIPCHECK (hipFreeHost(C_h));
+            HIPCHECK (hipHostFree(C_h));
         }
     } else {
         if (A_h) {