From 1707760bbf299c4afc0d7da2a84607d280ed596a Mon Sep 17 00:00:00 2001
From: Ben Sander <ben.sander@amd.com>
Date: Thu, 25 Feb 2016 04:21:24 -0600
Subject: [PATCH] Add abstraction for pinned/unpinned, and sync/async mem
 copies selection in tests

[ROCm/hip commit: ecec7e36d90fe281c9f023c2c3c3fb7e5dff4930]
---
 projects/hip/tests/src/test_common.h | 80 +++++++++++++++++++++++++++-
 1 file changed, 78 insertions(+), 2 deletions(-)

diff --git a/projects/hip/tests/src/test_common.h b/projects/hip/tests/src/test_common.h
index 1bf89f1604..e37eec7e86 100644
--- a/projects/hip/tests/src/test_common.h
+++ b/projects/hip/tests/src/test_common.h
@@ -16,6 +16,16 @@
 #define KCYN  "\x1B[36m"
 #define KWHT  "\x1B[37m"
 
+
+
+#ifdef __HIP_PLATFORM_HCC
+#define TYPENAME(T) typeid(T).name()
+#else
+#define TYPENAME(T) "?"
+#endif
+
+
+
 #define passed() \
     printf ("%sPASSED!%s\n",KGRN, KNRM);\
     exit(0);
@@ -82,12 +92,12 @@ vectorADD(hipLaunchParm lp,
             const T *A_d,
             const T *B_d,
             T *C_d,
-            size_t N)
+            size_t NELEM)
 {
     size_t offset = (hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x);
     size_t stride = hipBlockDim_x * hipGridDim_x ;
 
-    for (size_t i=offset; i<N; i+=stride) {
+    for (size_t i=offset; i<NELEM; i+=stride) {
         C_d[i] = A_d[i] + B_d[i];
 	}
 }
@@ -220,4 +230,70 @@ void checkVectorADD(T* A_h, T* B_h, T* result_H, size_t N, bool expectMatch=true
 
 }
 
+
+//---
+struct Pinned {
+	static const bool isPinned = true;
+	static const char *str() { return "Pinned"; };
+
+    static void *Alloc(size_t sizeBytes) 
+	{
+        void *p; 
+        HIPCHECK(hipMallocHost(&p, sizeBytes));
+        return p;
+    };
+};
+
+
+//---
+struct Unpinned 
+{
+	static const bool isPinned = false;
+	static const char *str() { return "Unpinned"; };
+
+    static void *Alloc(size_t sizeBytes) 
+	{
+        void *p  = malloc (sizeBytes);
+        HIPASSERT(p);
+		return p;
+    };
+};
+
+
+
+struct Memcpy
+{
+	static const char *str() { return "Memcpy"; };
+};
+
+struct MemcpyAsync
+{
+	static const char *str() { return "MemcpyAsync"; };
+};
+
+
+template <typename C> struct MemTraits;
+
+
+template<>
+struct MemTraits<Memcpy>
+{
+
+    static void Copy(void *dest, const void *src, size_t sizeBytes, hipMemcpyKind kind,  hipStream_t stream) 
+	{
+		HIPCHECK(hipMemcpy(dest, src, sizeBytes, kind));
+	}
+};
+
+
+template<>
+struct MemTraits<MemcpyAsync>
+{
+
+    static void Copy(void *dest, const void *src, size_t sizeBytes, hipMemcpyKind kind,  hipStream_t stream) 
+	{
+		HIPCHECK(hipMemcpyAsync(dest, src, sizeBytes, kind, stream));
+	}
+};
+
 }; // namespace HipTest