Add get_global_id example

3 files changed, 128 insertions, 1 deletions

diff --git a/Makefile b/Makefile
index 85f7f53..a298383 100644
--- a/Makefile
+++ b/Makefile
@@ -1,9 +1,12 @@
 CFLAGS=-g
 
-all: hello_world math-int
+all: hello_world math-int get-global-id
 
 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
+
+get-global-id: get_global_id.o util.o
+	gcc -o get-global-id $^ -L/usr/local/lib/ -lOpenCL
diff --git a/get_global_id.c b/get_global_id.c
new file mode 100644
index 0000000..65e20d4
--- /dev/null
+++ b/get_global_id.c
@@ -0,0 +1,117 @@
+#include <stdio.h>
+#include <stdlib.h>
+
+#include <CL/cl.h>
+
+
+int main(int argc, char ** argv)
+{
+   cl_int error;
+   cl_device_id device_id;
+
+   cl_context context;
+
+   cl_command_queue command_queue;
+
+   cl_kernel kernel;
+
+   cl_mem out_buffer;
+
+   unsigned i;
+   int out_data[10];
+   size_t global_work_size = 10;
+
+   if (!cluInitGpuDevice(&device_id)) {
+      return EXIT_FAILURE;
+   }
+
+   if (!cluCreateContext(&context, device_id)) {
+      return EXIT_FAILURE;
+   }
+
+   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, "global_id")) {
+      return EXIT_FAILURE;
+   }
+
+   out_buffer = clCreateBuffer(context,
+                               CL_MEM_WRITE_ONLY, /* Flags */
+                               sizeof(out_data), /* 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)) {
+      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(out_data), /* Bytes to read */
+                                out_data, /* 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");
+
+   for (i = 0; i < 10; i++) {
+      fprintf(stderr, "id %u = %u\n", i, out_data[i]);
+   }
+
+/*
+
+   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/global_id.cl b/global_id.cl
new file mode 100644
index 0000000..720bebb
--- /dev/null
+++ b/global_id.cl
@@ -0,0 +1,7 @@
+
+__kernel void global_id(__global int * out)
+{
+  unsigned id = get_global_id(0);
+  out[id] = id;
+}
+