From 1f513c54ef5ba597bc696c6e9eb09f61df8ba293 Mon Sep 17 00:00:00 2001 From: Tom Stellard Date: Mon, 5 Mar 2012 17:03:43 -0500 Subject: Add math-int program for testing simple math operations Usage: ./math-int kernel_name arg0 arg1 expected_value --- Makefile | 5 ++ add.cl | 4 ++ math-int.c | 127 ++++++++++++++++++++++++++++++++++++++++++++++ mod.cl | 5 ++ util.c | 168 +++++++++++++++++++++++++++++++++++++++++++++++++++++++++++++ util.h | 5 ++ 6 files changed, 314 insertions(+) create mode 100644 add.cl create mode 100644 math-int.c create mode 100644 mod.cl 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 +#include + +#include + + +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 +#include +#include +#include + #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); -- cgit v1.2.3