Allow updating application parameters for testing purposes

cudampilib/apps/app/appkernel.cu

@@ -23,7 +23,7 @@ __global__ void appkernel(unsigned char *devPtr) {
   devPtr[my_index] = devPtr[my_index] / 2;
 }
 
-extern "C" void launchkernelinstream(void *devPtr, cudaStream_t stream) {
+extern "C" void launchkernelinstream(void *devPtr, int batchSize, cudaStream_t stream) {
 
   dim3 blocksingrid(2);
   dim3 threadsinblock(1024);
@@ -35,4 +35,4 @@ extern "C" void launchkernelinstream(void *devPtr, cudaStream_t stream) {
   }
 }
 
-extern "C" void launchkernel(void *devPtr) { launchkernelinstream(devPtr, 0); }
+extern "C" void launchkernel(void *devPtr, int batchSize) { launchkernelinstream(devPtr, batchSize, 0); }

cudampilib/apps/collatz/app-streams-collatz.c

@@ -23,12 +23,14 @@ LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OU
 #define ENABLE_OUTPUT_LOGS
 #include "utility.h"
 
-long long VECTORSIZE = COLLATZ_VECTORSIZE;
+struct __cudampi__arguments_type __cudampi__arguments;
+
+long long VECTORSIZE;
 
 double *vectora;
 double *vectorc;
 
-int batchsize = COLLATZ_BATCH_SIZE;
+int batchsize;
 
 long long globalcounter = 0;
 
@@ -44,19 +46,11 @@ int main(int argc, char **argv)
 
   __cudampi__initializeMPI(argc, argv);
 
-  int alldevicescount = 0;
-
-  if (argc > 1) 
-  {
-    streamcount = atoi(argv[1]);
-  }
+  streamcount = __cudampi__arguments.number_of_streams;
+  batchsize = __cudampi__arguments.batch_size;
+  VECTORSIZE = __cudampi__arguments.problem_size;
 
-  if (argc > 2) 
-  {
-    powerlimit = atof(argv[2]);
-    log_message(LOG_INFO, "\nSetting power limit=%f\n", powerlimit);
-    __cudampi__setglobalpowerlimit(powerlimit);
-  }
+  int alldevicescount = 0;
 
   __cudampi__getDeviceCount(&alldevicescount);
 

cudampilib/apps/collatz/appkernelcollatz.cu

@@ -57,17 +57,17 @@ __global__ void appkernel(void *devPtr)
   devPtrc[my_index] = counter;
 }
 
-extern "C" void launchkernelinstream(void *devPtr, cudaStream_t stream) 
+extern "C" void launchkernelinstream(void *devPtr, int batchSize, cudaStream_t stream) 
 {
   dim3 blocksingrid(COLLATZ_BLOCKS_IN_GRID);
-  dim3 threadsinblock(COLLATZ_THREADS_IN_BLOCK);
+  dim3 threadsinblock(batchSize / COLLATZ_BLOCKS_IN_GRID);
 
-  log_message(LOG_DEBUG, "Launichng GPU Kernel with %i blocks in grid and %i threads in block.", COLLATZ_BLOCKS_IN_GRID, COLLATZ_THREADS_IN_BLOCK);
+  log_message(LOG_DEBUG, "Launichng GPU Kernel with %i blocks in grid and %i threads in block.", COLLATZ_BLOCKS_IN_GRID, batchSize / COLLATZ_BLOCKS_IN_GRID);
   appkernel<<<blocksingrid, threadsinblock, 0, stream>>>(devPtr);
 
   if (cudaSuccess != cudaGetLastError()) {
     log_message(LOG_ERROR, "Error during kernel launch in stream");
   }
 }
 
-extern "C" void launchkernel(void *devPtr) { launchkernelinstream(devPtr, 0); }
+extern "C" void launchkernel(void *devPtr, int batchSize) { launchkernelinstream(devPtr, batchSize, 0); }

cudampilib/apps/collatz/collatz_defines.h

@@ -1,9 +1,6 @@
 #ifndef COLLATZ_DEFINES_H
 #define COLLATZ_DEFINES_H
 
-#define COLLATZ_VECTORSIZE 200000000
-#define COLLATZ_BATCH_SIZE 50000
 #define COLLATZ_BLOCKS_IN_GRID 100
-#define COLLATZ_THREADS_IN_BLOCK (COLLATZ_BATCH_SIZE / COLLATZ_BLOCKS_IN_GRID)
 
 #endif // COLLATZ_DEFINES_H

cudampilib/apps/collatz/cpukernelcollatz.c

@@ -56,9 +56,9 @@ void appkernel(void *devPtr, int num_elements, int num_threads)
   }
 }
 
-extern void launchcpukernel(void *devPtr, int num_threads) 
+extern void launchcpukernel(void *devPtr, int batchSize, int num_threads) 
 {
-  int num_elements = COLLATZ_BATCH_SIZE;
+  int num_elements = batchSize;
   log_message(LOG_DEBUG, "Launichng CPU Kernel with %i elements and %i threads.", num_elements, num_threads);
   appkernel(devPtr, num_elements, num_threads);
 }

cudampilib/apps/patternsearch/app-streams-patternsearch.c

@@ -23,12 +23,14 @@ LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OU
 #define ENABLE_OUTPUT_LOGS
 #include "utility.h"
 
-long long VECTORSIZE = PATTERNSEARCH_VECTORSIZE;
+struct __cudampi__arguments_type __cudampi__arguments;
+
+long long VECTORSIZE;
 
 char *vectora;
 char *vectorc;
 
-int batchsize = PATTERNSEARCH_BATCH_SIZE;
+int batchsize;
 
 long long globalcounter = 0;
 
@@ -45,19 +47,11 @@ int main(int argc, char **argv)
 
   __cudampi__initializeMPI(argc, argv);
 
-  int alldevicescount = 0;
-
-  if (argc > 1) 
-  {
-    streamcount = atoi(argv[1]);
-  }
+  streamcount = __cudampi__arguments.number_of_streams;
+  batchsize = __cudampi__arguments.batch_size;
+  VECTORSIZE = __cudampi__arguments.problem_size;
 
-  if (argc > 2) 
-  {
-    powerlimit = atof(argv[2]);
-    log_message(LOG_ERROR, "\nSetting power limit=%f\n", powerlimit);
-    __cudampi__setglobalpowerlimit(powerlimit);
-  }
+  int alldevicescount = 0;
 
   __cudampi__getDeviceCount(&alldevicescount);
 

cudampilib/apps/patternsearch/appkernelpatternsearch.cu

@@ -45,17 +45,17 @@ __global__ void appkernel(void *devPtr) {
   }
 }
 
-extern "C" void launchkernelinstream(void *devPtr, cudaStream_t stream) {
+extern "C" void launchkernelinstream(void *devPtr, int batchSize, cudaStream_t stream) {
 
   dim3 blocksingrid(PATTERNSEARCH_BLOCKS_IN_GRID);
-  dim3 threadsinblock(PATTERNSEARCH_THREADS_IN_BLOCK);
+  dim3 threadsinblock(batchSize / PATTERNSEARCH_BLOCKS_IN_GRID);
 
-  log_message(LOG_DEBUG, "Launichng GPU Kernel with %i blocks in grid and %i threads in block.", PATTERNSEARCH_BLOCKS_IN_GRID, PATTERNSEARCH_THREADS_IN_BLOCK);
+  log_message(LOG_DEBUG, "Launichng GPU Kernel with %i blocks in grid and %i threads in block.", PATTERNSEARCH_BLOCKS_IN_GRID, batchSize / PATTERNSEARCH_BLOCKS_IN_GRID);
   appkernel<<<blocksingrid, threadsinblock, 0, stream>>>(devPtr);
 
   if (cudaSuccess != cudaGetLastError()) {
     log_message(LOG_ERROR, "Error during kernel launch in stream");
   }
 }
 
-extern "C" void launchkernel(void *devPtr) { launchkernelinstream(devPtr, 0); }
+extern "C" void launchkernel(void *devPtr, int batchSize) { launchkernelinstream(devPtr, batchSize, 0); }

cudampilib/apps/patternsearch/cpukernelpatternsearch.c

@@ -49,9 +49,9 @@ void appkernel(void *devPtr, int num_elements, int num_threads)
   }
 }
 
-extern void launchcpukernel(void *devPtr,int num_threads) 
+extern void launchcpukernel(void *devPtr, int batchSize, int num_threads) 
 {
-  int num_elements = PATTERNSEARCH_BATCH_SIZE;
+  int num_elements = batchSize;
   log_message(LOG_DEBUG, "Launichng CPU Kernel with %i elements and %i threads.", num_elements, num_threads);
   appkernel(devPtr, num_elements, num_threads);
 }

cudampilib/apps/patternsearch/patternsearch_defines.h

@@ -1,10 +1,7 @@
 #ifndef PATTERNSEARCH_DEFINES_H
 #define PATTERNSEARCH_DEFINES_H
 
-#define PATTERNSEARCH_VECTORSIZE 400000000
-#define PATTERNSEARCH_BATCH_SIZE 50000
 #define PATTERNSEARCH_BLOCKS_IN_GRID 100
-#define PATTERNSEARCH_THREADS_IN_BLOCK (PATTERNSEARCH_BATCH_SIZE / PATTERNSEARCH_BLOCKS_IN_GRID)
 #define PATTERNLENGTH 400
 #define PATTERNCOUNT 400
 

cudampilib/apps/vecadd/app-streams-vecadd.c

@@ -22,13 +22,15 @@ LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OU
 #define ENABLE_OUTPUT_LOGS
 #include "utility.h"
 
-long long VECTORSIZE = VECADD_VECTOR_SIZE;
+struct __cudampi__arguments_type __cudampi__arguments;
+
+long long VECTORSIZE;
 
 double *vectora;
 double *vectorb;
 double *vectorc;
 
-int batchsize = VECADD_BATCH_SIZE;
+int batchsize;
 
 long long globalcounter = 0;
 
@@ -44,19 +46,11 @@ int main(int argc, char **argv)
 
   __cudampi__initializeMPI(argc, argv);
 
-  int alldevicescount = 0;
-
-  if (argc > 1) 
-  {
-    streamcount = atoi(argv[1]);
-  }
+  streamcount = __cudampi__arguments.number_of_streams;
+  batchsize = __cudampi__arguments.batch_size;
+  VECTORSIZE = __cudampi__arguments.problem_size;
 
-  if (argc > 2) 
-  {
-    powerlimit = atof(argv[2]);
-    log_message(LOG_INFO, "\nSetting power limit=%f\n", powerlimit);
-    __cudampi__setglobalpowerlimit(powerlimit);
-  }
+  int alldevicescount = 0;
 
   __cudampi__getDeviceCount(&alldevicescount);
 

cudampilib/apps/vecadd/appkernelvecadd.cu

@@ -33,17 +33,17 @@ __global__ void appkernel(void *devPtr)
   devPtrc[my_index] = devPtra[my_index] / 2 + devPtrb[my_index] / 3;
 }
 
-extern "C" void launchkernelinstream(void *devPtr, cudaStream_t stream) 
+extern "C" void launchkernelinstream(void *devPtr, int batchSize, cudaStream_t stream) 
 {
   dim3 blocksingrid(VECADD_BLOCKS_IN_GRID);
-  dim3 threadsinblock(VECADD_THREADS_IN_BLOCK);
+  dim3 threadsinblock(batchSize / VECADD_BLOCKS_IN_GRID);
 
-  log_message(LOG_DEBUG, "Launichng GPU Kernel with %i blocks in grid and %i threads in block.", VECADD_BLOCKS_IN_GRID, VECADD_THREADS_IN_BLOCK);
+  log_message(LOG_DEBUG, "Launichng GPU Kernel with %i blocks in grid and %i threads in block.", VECADD_BLOCKS_IN_GRID, batchSize / VECADD_BLOCKS_IN_GRID);
   appkernel<<<blocksingrid, threadsinblock, 0, stream>>>(devPtr);
 
   if (cudaSuccess != cudaGetLastError()) {
     log_message(LOG_ERROR, "Error during kernel launch in stream");
   }
 }
 
-extern "C" void launchkernel(void *devPtr) { launchkernelinstream(devPtr, 0); }
+extern "C" void launchkernel(void *devPtr, int batchSize) { launchkernelinstream(devPtr, batchSize, 0); }

cudampilib/apps/vecadd/cpukernelvecadd.c

@@ -32,9 +32,9 @@ void appkernel(void *devPtr, int num_elements, int num_threads)
     }
 }
 
-extern void launchcpukernel(void *devPtr, int num_threads)
+extern void launchcpukernel(void *devPtr, int batchSize, int num_threads)
 {
-    int num_elements = VECADD_BATCH_SIZE;
+    int num_elements = batchSize;
     log_message(LOG_DEBUG, "Launichng CPU Kernel with %i elements and %i threads.", num_elements, num_threads);
     appkernel(devPtr, num_elements, num_threads);
 }

cudampilib/apps/vecadd/vecadd_defines.h

@@ -1,9 +1,6 @@
 #ifndef VECADD_DEFINES_H
 #define VECADD_DEFINES_H
 
-#define VECADD_BATCH_SIZE 100000
 #define VECADD_BLOCKS_IN_GRID 100
-#define VECADD_THREADS_IN_BLOCK (VECADD_BATCH_SIZE / VECADD_BLOCKS_IN_GRID)
-#define VECADD_VECTOR_SIZE 80000000
 
 #endif // VECADD_DEFINES_H

cudampilib/apps/vecmaxdiv/app-streams-vecmaxdiv.c

@@ -23,13 +23,16 @@ LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OU
 #define ENABLE_OUTPUT_LOGS
 #include "utility.h"
 
-long long VECTORSIZE = VECMAXDIV_VECTORSIZE;
+
+struct __cudampi__arguments_type __cudampi__arguments;
+
+long long VECTORSIZE;
 
 double *vectora;
 double *vectorb;
 double *vectorc;
 
-int batchsize = VECMAXDIV_BATCH_SIZE;
+int batchsize;
 
 long long globalcounter = 0;
 
@@ -45,19 +48,12 @@ int main(int argc, char **argv)
 
   __cudampi__initializeMPI(argc, argv);
 
-  int alldevicescount = 0;
+  streamcount = __cudampi__arguments.number_of_streams;
+  batchsize = __cudampi__arguments.batch_size;
+  VECTORSIZE = __cudampi__arguments.problem_size;
 
-  if (argc > 1) 
-  {
-    streamcount = atoi(argv[1]);
-  }
+  int alldevicescount = 0;
 
-  if (argc > 2) 
-  {
-    powerlimit = atof(argv[2]);
-    log_message(LOG_INFO,"\nSetting power limit=%f\n", powerlimit);
-    __cudampi__setglobalpowerlimit(powerlimit);
-  }
   __cudampi__getDeviceCount(&alldevicescount);
 
   cudaHostAlloc((void **)&vectora, sizeof(double) * VECTORSIZE, cudaHostAllocDefault);

cudampilib/apps/vecmaxdiv/appkernelvecmaxdiv.cu

@@ -55,17 +55,17 @@ __global__ void appkernel(void *devPtr) {
   devPtrc[my_index] = result;
 }
 
-extern "C" void launchkernelinstream(void *devPtr, cudaStream_t stream) {
+extern "C" void launchkernelinstream(void *devPtr, int batchSize, cudaStream_t stream) {
 
   dim3 blocksingrid(VECMAXDIV_BLOCKS_IN_GRID);
-  dim3 threadsinblock(VECMAXDIV_THREADS_IN_BLOCK);
+  dim3 threadsinblock(batchSize / VECMAXDIV_BLOCKS_IN_GRID);
 
-  log_message(LOG_DEBUG, "Launichng GPU Kernel with %i blocks in grid and %i threads in block.", VECMAXDIV_BLOCKS_IN_GRID, VECMAXDIV_THREADS_IN_BLOCK);
+  log_message(LOG_DEBUG, "Launichng GPU Kernel with %i blocks in grid and %i threads in block.", VECMAXDIV_BLOCKS_IN_GRID, batchSize / VECMAXDIV_BLOCKS_IN_GRID);
   appkernel<<<blocksingrid, threadsinblock, 0, stream>>>(devPtr);
 
   if (cudaSuccess != cudaGetLastError()) {
     log_message(LOG_ERROR, "Error during kernel launch in stream");
   }
 }
 
-extern "C" void launchkernel(void *devPtr) { launchkernelinstream(devPtr, 0); }
+extern "C" void launchkernel(void *devPtr, int batchSize) { launchkernelinstream(devPtr, batchSize, 0); }

cudampilib/apps/vecmaxdiv/cpukernelvecmaxdiv.c

@@ -59,9 +59,9 @@ for (long my_index = 0 ; my_index < num_elements; my_index++)
 }
 }
 
-extern void launchcpukernel(void *devPtr, int num_threads) 
+extern void launchcpukernel(void *devPtr, int batchSize, int num_threads) 
 {
-    int num_elements = VECMAXDIV_BATCH_SIZE;
+    int num_elements = batchSize;
     log_message(LOG_DEBUG, "Launichng CPU Kernel with %i elements and %i threads.", num_elements, num_threads);
     appkernel(devPtr, num_elements, num_threads);
 }

cudampilib/apps/vecmaxdiv/vecmaxdiv_defines.h

@@ -1,9 +1,6 @@
 #ifndef VECMAXDIV_DEFINES_H
 #define VECMAXDIV_DEFINES_H
 
-#define VECMAXDIV_VECTORSIZE 200000000
-#define VECMAXDIV_BATCH_SIZE 50000
 #define VECMAXDIV_BLOCKS_IN_GRID 100
-#define VECMAXDIV_THREADS_IN_BLOCK (VECMAXDIV_BATCH_SIZE / VECMAXDIV_BLOCKS_IN_GRID)
 
 #endif // VECMAXDIV_DEFINES_H