Basic support for memory allocation on GPU using CUDA unified memory. Partially addressing #284. IOR support completed.

Author: JulianKunkel

configure.ac

@@ -73,6 +73,26 @@ AS_IF([test "$ac_cv_header_gpfs_h" = "yes" -o "$ac_cv_header_gpfs_fcntl_h" = "ye
     ])
 ])
 
+# Check for CUDA
+AC_ARG_WITH([cuda],
+        [AS_HELP_STRING([--with-cuda],
+                [support configurable CUDA @<:@default=check@:>@])],
+        [], [with_cuda=check])
+
+AS_IF([test "x$with_cuda" != xno], [
+        AC_CHECK_HEADERS([cuda_runtime.h], [AC_DEFINE([HAVE_CUDA], [], [CUDA GPU API found])], [
+                if test "x$with_cuda" != xcheck; then
+                        AC_MSG_FAILURE([--with-cuda was given, <cuda_runtime.h> not found])
+                fi
+        ])
+AS_IF([test "$ac_cv_header_cuda_runtime_h" = "yes"], [
+        AC_SEARCH_LIBS([cudaMalloc], [cudart], [],
+        [AC_MSG_ERROR([Library containing cudaMalloc symbol not found])])
+    ])
+])
+AM_CONDITIONAL([USE_CUDA], [test x$with_cuda = xyes])
+
+
 # Check for system capabilities
 AC_SYS_LARGEFILE
 

src/Makefile.am

@@ -41,6 +41,10 @@ extraLDFLAGS  += -L/opt/hadoop-2.2.0/lib/native
 extraLDADD    += -lhdfs
 endif
 
+if USE_CUDA
+extraLDADD    += -lcudart
+endif
+
 if USE_HDF5_AIORI
 extraSOURCES  += aiori-HDF5.c
 extraLDADD    += -lhdf5 -lz

src/ior.c

@@ -472,44 +472,6 @@ static int CountErrors(IOR_param_t * test, int access, int errors)
         return (allErrors);
 }
 
-/*
- * Allocate a page-aligned (required by O_DIRECT) buffer.
- */
-static void *aligned_buffer_alloc(size_t size)
-{
-        size_t pageMask;
-        char *buf, *tmp;
-        char *aligned;
-
-#ifdef HAVE_SYSCONF
-        long pageSize = sysconf(_SC_PAGESIZE);
-#else
-        size_t pageSize = getpagesize();
-#endif
-
-        pageMask = pageSize - 1;
-        buf = malloc(size + pageSize + sizeof(void *));
-        if (buf == NULL)
-                ERR("out of memory");
-        /* find the alinged buffer */
-        tmp = buf + sizeof(char *);
-        aligned = tmp + pageSize - ((size_t) tmp & pageMask);
-        /* write a pointer to the original malloc()ed buffer into the bytes
-           preceding "aligned", so that the aligned buffer can later be free()ed */
-        tmp = aligned - sizeof(void *);
-        *(void **)tmp = buf;
-
-        return (void *)aligned;
-}
-
-/*
- * Free a buffer allocated by aligned_buffer_alloc().
- */
-static void aligned_buffer_free(void *buf)
-{
-        free(*(void **)((char *)buf - sizeof(char *)));
-}
-
 void AllocResults(IOR_test_t *test)
 {
   int reps;
@@ -1018,7 +980,7 @@ static void InitTests(IOR_test_t *tests)
 static void XferBuffersSetup(IOR_io_buffers* ioBuffers, IOR_param_t* test,
                              int pretendRank)
 {
-        ioBuffers->buffer = aligned_buffer_alloc(test->transferSize);
+        ioBuffers->buffer = aligned_buffer_alloc(test->transferSize, test->gpuMemoryFlags);
 }
 
 /*
@@ -1027,7 +989,7 @@ static void XferBuffersSetup(IOR_io_buffers* ioBuffers, IOR_param_t* test,
 static void XferBuffersFree(IOR_io_buffers* ioBuffers, IOR_param_t* test)
 
 {
-        aligned_buffer_free(ioBuffers->buffer);
+        aligned_buffer_free(ioBuffers->buffer, test->gpuMemoryFlags);
 }
 
 
@@ -1829,7 +1791,7 @@ static IOR_offset_t WriteOrRead(IOR_param_t *test, IOR_results_t *results,
 
         void * randomPrefillBuffer = NULL;
         if(test->randomPrefillBlocksize && (access == WRITE || access == WRITECHECK)){
-          randomPrefillBuffer = aligned_buffer_alloc(test->randomPrefillBlocksize);
+          randomPrefillBuffer = aligned_buffer_alloc(test->randomPrefillBlocksize, test->gpuMemoryFlags);
           // store invalid data into the buffer
           memset(randomPrefillBuffer, -1, test->randomPrefillBlocksize);
         }
@@ -1951,7 +1913,7 @@ static IOR_offset_t WriteOrRead(IOR_param_t *test, IOR_results_t *results,
                 backend->fsync(fd, test->backend_options);       /*fsync after all accesses */
         }
         if(randomPrefillBuffer){
-          aligned_buffer_free(randomPrefillBuffer);
+          aligned_buffer_free(randomPrefillBuffer, test->gpuMemoryFlags);
         }
 
         return (dataMoved);

src/ior.h

@@ -101,7 +101,8 @@ typedef struct
     MPI_Comm     testComm;           /* Current MPI communicator */
     MPI_Comm     mpi_comm_world;           /* The global MPI communicator */
     int dryRun;                      /* do not perform any I/Os just run evtl. inputs print dummy output */
-  int dualMount;                   /* dual mount points */
+    int dualMount;                   /* dual mount points */
+    int gpuMemoryFlags;              /* use the GPU to store the data */
     int numTasks;                    /* number of tasks for test */
     int numNodes;                    /* number of nodes for test */
     int numTasksOnNode0;             /* number of tasks on node 0 (usually all the same, but don't have to be, use with caution) */

src/mdtest.c

@@ -666,10 +666,7 @@ void mdtest_read(int random, int dirs, const long dir_iter, char *path) {
 
     /* allocate read buffer */
     if (o.read_bytes > 0) {
-        int alloc_res = posix_memalign((void**)&read_buffer, sysconf(_SC_PAGESIZE), o.read_bytes);
-        if (alloc_res) {
-            FAIL("out of memory");
-        }
+        read_buffer = aligned_buffer_alloc(o.read_bytes, 0);
         memset(read_buffer, -1, o.read_bytes);
     }
 
@@ -764,7 +761,7 @@ void mdtest_read(int random, int dirs, const long dir_iter, char *path) {
         o.backend->close (aiori_fh, o.backend_options);
     }
     if(o.read_bytes){
-      free(read_buffer);
+      aligned_buffer_free(read_buffer, 0);
     }
 }
 
@@ -2301,10 +2298,7 @@ mdtest_results_t * mdtest_run(int argc, char **argv, MPI_Comm world_com, FILE *
 
     /* allocate and initialize write buffer with # */
     if (o.write_bytes > 0) {
-        int alloc_res = posix_memalign((void**)& o.write_buffer, sysconf(_SC_PAGESIZE), o.write_bytes);
-        if (alloc_res) {
-            FAIL("out of memory");
-        }
+        o.write_buffer = aligned_buffer_alloc(o.write_bytes, 0);
         generate_memory_pattern(o.write_buffer, o.write_bytes);
     }
 
@@ -2445,7 +2439,7 @@ mdtest_results_t * mdtest_run(int argc, char **argv, MPI_Comm world_com, FILE *
     }
 
     if (o.write_bytes > 0) {
-      free(o.write_buffer);
+      aligned_buffer_free(o.write_buffer, 0);
     }
 
     return o.summary_table;

src/parse_options.c

@@ -123,6 +123,8 @@ void DecodeDirective(char *line, IOR_param_t *params, options_all_t * module_opt
                 params->testFileName  = strdup(value);
         } else if (strcasecmp(option, "dualmount") == 0){
                 params->dualMount = atoi(value);
+        } else if (strcasecmp(option, "allocateBufferOnGPU") == 0) {
+                params->gpuMemoryFlags = atoi(value);                
         } else if (strcasecmp(option, "deadlineforstonewalling") == 0) {
                 params->deadlineForStonewalling = atoi(value);
         } else if (strcasecmp(option, "stoneWallingWearOut") == 0) {
@@ -398,6 +400,7 @@ option_help * createGlobalOptions(IOR_param_t * params){
     {.help="  -O stoneWallingWearOut=1           -- once the stonewalling timeout is over, all process finish to access the amount of data", .arg = OPTION_OPTIONAL_ARGUMENT},
     {.help="  -O stoneWallingWearOutIterations=N -- stop after processing this number of iterations, needed for reading data back written with stoneWallingWearOut", .arg = OPTION_OPTIONAL_ARGUMENT},
     {.help="  -O stoneWallingStatusFile=FILE     -- this file keeps the number of iterations from stonewalling during write and allows to use them for read", .arg = OPTION_OPTIONAL_ARGUMENT},
+    {.help="  -O allocateBufferOnGPU=1                     -- allocate I/O buffers on the GPU", .arg = OPTION_OPTIONAL_ARGUMENT},
     {'e', NULL,        "fsync -- perform a fsync() operation at the end of each read/write phase", OPTION_FLAG, 'd', & params->fsync},
     {'E', NULL,        "useExistingTestFile -- do not remove test file before write access", OPTION_FLAG, 'd', & params->useExistingTestFile},
     {'f', NULL,        "scriptFile -- test script name", OPTION_OPTIONAL_ARGUMENT, 's', & params->testscripts},

src/utilities.c

@@ -37,6 +37,10 @@
 #include <sys/types.h>
 #include <time.h>
 
+#ifdef HAVE_CUDA
+#include <cuda_runtime.h>
+#endif
+
 #ifndef _WIN32
 #  include <regex.h>
 #  ifdef __sun                    /* SunOS does not support statfs(), instead uses statvfs() */
@@ -864,3 +868,63 @@ unsigned long GetProcessorAndCore(int *chip, int *core){
 	return 1;
 }
 #endif
+
+
+
+/*
+ * Allocate a page-aligned (required by O_DIRECT) buffer.
+ */
+void *aligned_buffer_alloc(size_t size, int gpu)
+{
+  size_t pageMask;
+  char *buf, *tmp;
+  char *aligned;
+
+  if(gpu){
+#ifdef HAVE_CUDA
+    // use unified memory here to allow drop-in-replacement
+    if (cudaMallocManaged((void**) & buf, size, cudaMemAttachGlobal) != cudaSuccess){
+      ERR("Cannot allocate buffer on GPU");
+    }
+    return buf;
+#else
+    ERR("No CUDA supported, cannot allocate on the GPU");
+#endif
+  }
+
+#ifdef HAVE_SYSCONF
+  long pageSize = sysconf(_SC_PAGESIZE);
+#else
+  size_t pageSize = getpagesize();
+#endif
+
+  pageMask = pageSize - 1;
+  buf = safeMalloc(size + pageSize + sizeof(void *));
+  /* find the alinged buffer */
+  tmp = buf + sizeof(char *);
+  aligned = tmp + pageSize - ((size_t) tmp & pageMask);
+  /* write a pointer to the original malloc()ed buffer into the bytes
+     preceding "aligned", so that the aligned buffer can later be free()ed */
+  tmp = aligned - sizeof(void *);
+  *(void **)tmp = buf;
+
+  return (void *)aligned;
+}
+
+/*
+ * Free a buffer allocated by aligned_buffer_alloc().
+ */
+void aligned_buffer_free(void *buf, int gpu)
+{
+  if(gpu){
+#ifdef HAVE_CUDA
+    if (cudaFree(buf) != cudaSuccess){
+      WARN("Cannot free buffer on GPU");
+    }
+    return;
+#else
+    ERR("No CUDA supported, cannot free on the GPU");
+#endif
+  }
+  free(*(void **)((char *)buf - sizeof(char *)));
+}

src/utilities.h

@@ -58,5 +58,6 @@ void init_clock(MPI_Comm com);
 double GetTimeStamp(void);
 char * PrintTimestamp(); // TODO remove this function
 unsigned long GetProcessorAndCore(int *chip, int *core);
-
+void *aligned_buffer_alloc(size_t size, int gpu);
+void aligned_buffer_free(void *buf, int gpu);
 #endif  /* !_UTILITIES_H */