Add math-int program for testing simple math operations

Usage: ./math-int kernel_name arg0 arg1 expected_value

6 files changed, 314 insertions, 0 deletions

diff --git a/Makefile b/Makefile
index 02e399d..85f7f53 100644
--- a/Makefile
+++ b/Makefile
@@ -1,4 +1,9 @@
 CFLAGS=-g
 
+all: hello_world math-int
+
 hello_world: hello_world.o util.o
 	gcc -o hello_world $^ -L/usr/local/lib/ -lOpenCL
+
+math-int: math-int.o util.o
+	gcc -o math-int $^ -L/usr/local/lib/ -lOpenCL
diff --git a/add.cl b/add.cl
new file mode 100644
index 0000000..2fa9378
--- /dev/null
+++ b/add.cl
@@ -0,0 +1,4 @@
+__kernel void add(__global int * out, int arg0, int arg1)
+{
+  out[0] = arg0 + arg1;
+}
diff --git a/math-int.c b/math-int.c
new file mode 100644
index 0000000..e02439e
--- /dev/null
+++ b/math-int.c
@@ -0,0 +1,127 @@
+#include <stdio.h>
+#include <stdlib.h>
+
+#include <CL/cl.h>
+
+
+int main(int argc, char ** argv)
+{
+   const char * kernel_name = argv[1];
+   int arg0, arg1, expected, result;
+   cl_int error;
+   cl_device_id device_id;
+
+   cl_context context;
+
+   cl_command_queue command_queue;
+
+   cl_kernel kernel;
+
+   cl_mem out_buffer;
+   int out_value = 0;
+   size_t global_work_size = 1;
+
+   /* Parse command line args */
+   arg0 = atoi(argv[2]);
+   arg1 = atoi(argv[3]);
+   expected = atoi(argv[4]);
+
+   if (!cluInitGpuDevice(&device_id)) {
+      return EXIT_FAILURE;
+   }
+
+   context = clCreateContext(NULL, /* Properties */
+                           1, /* Number of devices */
+                           &device_id, /* Device pointer */
+                           NULL, /* Callback for reporting errors */
+                           NULL, /* User data to pass to error callback */
+                           &error); /* Error code */
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clCreateContext() failed: %s\n", cluErrorString(error));
+      return EXIT_FAILURE;
+   }
+
+   fprintf(stderr, "clCreateContext() succeeded.\n");
+
+   command_queue = clCreateCommandQueue(context,
+                                        device_id,
+                                        0, /* Command queue properties */
+                                        &error); /* Error code */
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clCreateCommandQueue() failed: %s\n",
+                      cluErrorString(error));
+      return EXIT_FAILURE;
+   }
+
+   fprintf(stderr, "clCreateCommandQueue() succeeded.\n");
+
+   if (!cluCreateKernel(context, device_id, &kernel, kernel_name)) {
+      return EXIT_FAILURE;
+   }
+
+   out_buffer = clCreateBuffer(context,
+                               CL_MEM_WRITE_ONLY, /* Flags */
+                               sizeof(int), /* Size of buffer */
+                               NULL, /* Pointer to the data */
+                               &error); /* error code */
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clCreateBuffer() failed: %s\n", cluErrorString(error));
+      return EXIT_FAILURE;
+   }
+
+   fprintf(stderr, "clCreateBuffer() succeeded.\n");
+
+  if (   !cluKernelSetArg(kernel, 0, sizeof(cl_mem), &out_buffer)
+      || !cluKernelSetArg(kernel, 1, sizeof(int), &arg0)
+      || !cluKernelSetArg(kernel, 2, sizeof(int), &arg1)) {
+      return EXIT_FAILURE;
+   }
+
+   error = clEnqueueNDRangeKernel(command_queue,
+                                  kernel,
+                                  1, /* Number of dimensions */
+                                  NULL, /* Global work offset */
+                                  &global_work_size,
+                                  &global_work_size, /* local work size */
+                                  0, /* Events in wait list */
+                                  NULL, /* Wait list */
+                                  NULL); /* Event object for this event */
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clEnqueueNDRangeKernel() failed: %s\n",
+                      cluErrorString(error));
+      return EXIT_FAILURE;
+   }
+
+   fprintf(stderr, "clEnqueueNDRangeKernel() suceeded.\n");
+
+   error = clEnqueueReadBuffer(command_queue,
+                                out_buffer,
+                                CL_TRUE, /* TRUE means it is a blocking read. */
+                                0, /* Buffer offset to read from. */
+                                sizeof(int), /* Bytes to read */
+                                &out_value, /* Pointer to store the data */
+                                0, /* Events in wait list */
+                                NULL, /* Wait list */
+                                NULL); /* Event object */
+
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clEnqueueReadBuffer() failed: %s\n",
+                      cluErrorString(error));
+      return EXIT_FAILURE;
+   }
+
+   fprintf(stderr, "clEnqueueReadBuffer() suceeded.\n");
+
+   if (out_value == expected) {
+      fprintf(stderr, "Pass\n");
+      return EXIT_SUCCESS;
+   } else {
+      fprintf(stderr, "Expected %d, but got %d\n", expected, out_value);
+      return EXIT_FAILURE;
+   }
+}
diff --git a/mod.cl b/mod.cl
new file mode 100644
index 0000000..a881e00
--- /dev/null
+++ b/mod.cl
@@ -0,0 +1,5 @@
+
+__kernel void mod(__global int * out, int arg0, int arg1)
+{
+  out[0] = arg0 % arg1;
+}
diff --git a/util.c b/util.c
index 6b493af..33031cb 100644
--- a/util.c
+++ b/util.c
@@ -1,5 +1,9 @@
 #include <CL/cl.h>
 
+#include <fcntl.h>
+#include <stdio.h>
+#include <string.h>
+
 #include "util.h"
 
 #define CASE_ERR(ec) case ec: return #ec;
@@ -45,3 +49,167 @@ const char * cluErrorString(cl_int error)
 
    }
 }
+
+unsigned cluInitGpuDevice(cl_device_id * device_id)
+{
+   cl_int error;
+
+   cl_uint total_platforms;
+   cl_platform_id platform_id;
+
+   cl_uint total_gpu_devices;
+
+   error = clGetPlatformIDs(
+                           1, /* Max number of platform IDs to return */
+                           &platform_id, /* Pointer to platform_id */
+                           &total_platforms); /* Total number of platforms
+                                               * found on the system */
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clGetPlatformIDs() failed: %s\n", cluErrorString(error));
+      return 0;
+   }
+
+   fprintf(stderr, "There are %u platforms.\n", total_platforms);
+
+   error = clGetDeviceIDs(platform_id,
+                          CL_DEVICE_TYPE_GPU,
+                          1,
+                          device_id,
+                          &total_gpu_devices);
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clGetDeviceIDs() failed: %s\n", cluErrorString(error));
+      return 0;
+   }
+
+   fprintf(stderr, "There are %u GPU devices.\n", total_gpu_devices);
+
+   return 1;
+}
+
+#define CODE_CHUNK 64
+
+unsigned cluCreateKernel(cl_context context, cl_device_id device_id,
+                         cl_kernel * kernel, const char * kernel_name)
+{
+   char * filename;
+   char * code = NULL;
+   size_t code_len = 0;
+   int fd;
+   int bytes_read;
+
+   /* +3 .cl
+    * +1 NULL byte
+    * --
+    * +4
+    */
+   unsigned filename_len = strlen(kernel_name) + 4;
+
+   cl_int error;
+   cl_program program;
+
+   /* Determine file name */
+   filename = malloc (filename_len + 4);
+   if (!filename) {
+      fprintf(stderr, "Failed to malloc filename.\n");
+      return 0;
+   }
+
+   snprintf(filename, filename_len, "%s.cl", kernel_name);
+
+   /* Open file */
+   fd = open(filename, O_RDONLY);
+   if (fd < 0) {
+      fprintf(stderr, "Failed to open file: %s\n", filename);
+      return 0;
+   }
+
+   /* Read code */
+   do {
+      code = realloc(code, (bytes_read + CODE_CHUNK) * sizeof(unsigned char));
+      if (!code) {
+         fprintf(stderr, "Failed to realloc code.\n");
+         return 0;
+      }
+
+      bytes_read = read(fd, code + bytes_read, CODE_CHUNK);
+      if (bytes_read < 0) {
+         fprintf(stderr, "Failed to read code.\n");
+         return 0;
+      }
+      code_len += bytes_read;
+   } while(bytes_read == CODE_CHUNK);
+
+   /* Create program */
+   program = clCreateProgramWithSource(context,
+                                       1, /* Number of strings */
+                                       &code,
+                                       &code_len, /* String lengths, 0 means all the
+                                              * strings are NULL terminated. */
+                                       &error);
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clCreateProgramWithSource() failed: %s\n",
+                      cluErrorString(error));
+      return 0;
+   }
+
+   fprintf(stderr, "clCreateProgramWithSource() suceeded.\n");
+
+   /* Build program */
+   error = clBuildProgram(program,
+                          1, /* Number of devices */
+                          &device_id,
+                          NULL, /* options */
+                          NULL, /* callback function when compile is complete */
+                          NULL); /* user data for callback */
+
+
+   if (error != CL_SUCCESS) {
+      char build_str[10000];
+      error = clGetProgramBuildInfo(program,
+                                    device_id,
+                                    CL_PROGRAM_BUILD_LOG,
+                                    10000, /* Size of output string */
+                                    build_str, /* pointer to write the log to */
+                                    NULL); /* Number of bytes written to the log */
+      if (error != CL_SUCCESS) {
+         fprintf(stderr, "clGetProgramBuildInfo() failed: %s\n",
+                          cluErrorString(error));
+      } else {
+         fprintf(stderr, "Build Log: \n%s\n\n", build_str);
+      }
+      return 0;
+   }
+
+   fprintf(stderr, "clBuildProgram() suceeded.\n");
+
+   *kernel = clCreateKernel(program, kernel_name, &error);
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clCreateKernel() failed: %s\n", cluErrorString(error));
+      return 0;
+   }
+
+   fprintf(stderr, "clCreateKernel() suceeded.\n");
+
+   return 1;
+}
+
+unsigned cluKernelSetArg(cl_kernel kernel, cl_uint index, size_t size,
+                         const void * value)
+{
+   cl_int error;
+
+   error = clSetKernelArg(kernel, index, size, value);
+
+   if (error != CL_SUCCESS) {
+      fprintf(stderr, "clSetKernelArg failed: %s\n", cluErrorString(error));
+      return 0;
+   }
+
+   fprintf(stderr, "clSetKernelArg() succeeded.\n");
+
+   return 1;
+}
diff --git a/util.h b/util.h
index 5293d32..922ac8d 100644
--- a/util.h
+++ b/util.h
@@ -1,2 +1,7 @@
 
 const char * cluErrorString(cl_int error);
+unsigned cluInitGpuDevice(cl_device_id * device_id);
+unsigned cluCreateKernel(cl_context context, cl_device_id device_id,
+                         cl_kernel * kernel, const char * kernel_name);
+unsigned cluKernelSetArg(cl_kernel kernel, cl_uint index, size_t size,
+                         const void * value);