Skip to content
Browse files

Initial commit

  • Loading branch information...
0 parents commit d8bcdf7078fca95f4053f1deb0b46d77230bf3f6 @matze committed Jul 19, 2012
Showing with 433 additions and 0 deletions.
  1. +82 −0 check.c
  2. +243 −0 ocl.c
  3. +34 −0 ocl.h
  4. +23 −0 programs.cl
  5. BIN waf
  6. +51 −0 wscript
82 check.c
@@ -0,0 +1,82 @@
+
+#include <stdarg.h>
+#include <glib.h>
+#include <glib/gprintf.h>
+#include "ocl.h"
+
+static void
+check_kernel (cl_program program, const gchar *name, int *errcode)
+{
+ cl_kernel kernel;
+
+ kernel = clCreateKernel (program, name, errcode);
+
+ if (kernel != NULL)
+ clReleaseKernel (kernel);
+}
+
+static void
+print_check (const gchar *format, int errcode, ...)
+{
+ va_list args;
+
+ va_start (args, errcode);
+ g_vprintf (format, args);
+ va_end (args);
+
+ if (errcode == CL_SUCCESS) {
+ g_print (": OK\n");
+ }
+ else
+ g_print (": Error: %s\n", ocl_error (errcode));
+}
+
+int main(int argc, char *argv[])
+{
+ OCL *ocl;
+ GOptionContext *context;
+ GError *error = NULL;
+ int errcode = CL_SUCCESS;
+ cl_program program;
+
+ static gint first_device = 0;
+ static gint last_device = OCL_ALL_DEVICES;
+
+ static GOptionEntry entries[] =
+ {
+ { "first", 'f', 0, G_OPTION_ARG_INT, &first_device, "First device to use", "N" },
+ { "last", 'l', 0, G_OPTION_ARG_INT, &last_device, "Last device to use", "M" },
+ { NULL }
+ };
+
+ static const gchar *kernels[] =
+ {
+ "assign",
+ "two_const_params",
+ "three_const_params",
+ "four_const_params",
+ NULL
+ };
+
+ context = g_option_context_new ("");
+ g_option_context_add_main_entries (context, entries, NULL);
+
+ if (!g_option_context_parse (context, &argc, &argv, &error)) {
+ g_print ("Option parsing failed: %s\n", error->message);
+ return 1;
+ }
+
+ ocl = ocl_new (first_device, last_device, &errcode);
+ print_check ("Initialization", errcode);
+
+ program = ocl_program_new_from_file (ocl, "programs.cl", "", &errcode);
+ print_check ("Creating backproject program", errcode);
+
+ for (guint i = 0; kernels[i] != NULL; i++) {
+ check_kernel (program, kernels[i], &errcode);
+ print_check ("Creating kernel `%s`", errcode, kernels[i]);
+ }
+
+ ocl_free (ocl);
+ return 0;
+}
243 ocl.c
@@ -0,0 +1,243 @@
+#include <stdio.h>
+#include "ocl.h"
+
+#define check_and_return_ocl_errcode(code) \
+ if (code != CL_SUCCESS) { g_warning ("OpenCL error at %s:%i\n", __FILE__, __LINE__); \
+ *ocl_errcode = code; return NULL; }
+
+static const gchar *error_msgs[];
+
+const gchar *
+ocl_error (int ocl_errcode)
+{
+ if (ocl_errcode >= -14)
+ return error_msgs[-ocl_errcode];
+
+ if (ocl_errcode <= -30)
+ return error_msgs[-ocl_errcode-15];
+
+ return NULL;
+}
+
+static gchar *
+ocl_read_program (const gchar *filename)
+{
+ FILE *fp = fopen (filename, "r");
+ if (fp == NULL)
+ return NULL;
+
+ fseek (fp, 0, SEEK_END);
+ const size_t length = ftell (fp);
+ rewind (fp);
+
+ gchar *buffer = (gchar *) g_malloc0(length+1);
+ if (buffer == NULL) {
+ fclose (fp);
+ return NULL;
+ }
+
+ size_t buffer_length = fread (buffer, 1, length, fp);
+ fclose (fp);
+ if (buffer_length != length) {
+ g_free (buffer);
+ return NULL;
+ }
+
+ return buffer;
+}
+
+cl_program
+ocl_program_new_from_file (OCL *ocl, const gchar *filename, const gchar *options, int *ocl_errcode)
+{
+ cl_program program;
+ gchar *buffer;
+
+ g_return_val_if_fail (ocl != NULL && filename != NULL && ocl_errcode != NULL, NULL);
+
+ buffer = ocl_read_program (filename);
+
+ if (buffer == NULL)
+ return NULL;
+
+ program = ocl_program_new_from_source (ocl, buffer, options, ocl_errcode);
+ g_free (buffer);
+ return program;
+}
+
+cl_program
+ocl_program_new_from_source (OCL *ocl, const gchar *source, const gchar *options, int *ocl_errcode)
+{
+ cl_program program;
+ int errcode;
+
+ g_return_val_if_fail (ocl != NULL && source != NULL && ocl_errcode != NULL, NULL);
+
+ program = clCreateProgramWithSource (ocl->context, 1, (const char **) &source, NULL, &errcode);
+ check_and_return_ocl_errcode (errcode);
+ errcode = clBuildProgram (program, ocl->num_devices, ocl->devices, options, NULL, NULL);
+
+ if (errcode != CL_SUCCESS) {
+ const int LOG_SIZE = 4096;
+ gchar* log = (gchar *) g_malloc0(LOG_SIZE * sizeof (char));
+
+ clGetProgramBuildInfo (program, ocl->devices[0], CL_PROGRAM_BUILD_LOG, LOG_SIZE, (void*) log, NULL);
+ g_print ("\n=== Build log for ===%s\n\n", log);
+ g_free (log);
+ return NULL;
+ }
+
+ return program;
+}
+
+static cl_platform_id
+get_platform (gchar *prefered_platform)
+{
+ gchar result[256];
+ cl_platform_id *platforms = NULL;
+ cl_uint num_platforms = 0;
+ cl_platform_id platform = NULL;
+
+ cl_int errcode = clGetPlatformIDs (0, NULL, &num_platforms);
+ if (errcode != CL_SUCCESS)
+ return NULL;
+
+ platforms = (cl_platform_id *) g_malloc0(num_platforms * sizeof (cl_platform_id));
+ errcode = clGetPlatformIDs (num_platforms, platforms, NULL);
+
+ if (errcode == CL_SUCCESS)
+ platform = platforms[0];
+
+ for (int i = 0; i < num_platforms; i++) {
+ clGetPlatformInfo (platforms[i], CL_PLATFORM_VENDOR, 256, result, NULL);
+ if (g_strstr_len (result, -1, prefered_platform) != NULL) {
+ platform = platforms[i];
+ break;
+ }
+ }
+
+ g_free (platforms);
+ return platform;
+}
+
+OCL *
+ocl_new (guint from, guint to, int *ocl_errcode)
+{
+ OCL *ocl;
+ int errcode;
+ const size_t len = 256;
+ char string_buffer[len];
+
+ g_return_val_if_fail (ocl_errcode != NULL, NULL);
+ g_return_val_if_fail (from <= to, NULL);
+
+ ocl = g_malloc0(sizeof (OCL));
+ cl_platform_id platform = get_platform ("");
+
+ if (platform == NULL)
+ return NULL;
+
+ /* Get all device types */
+ errcode = clGetDeviceIDs (platform, CL_DEVICE_TYPE_ALL, 0, NULL, &ocl->num_all_devices);
+ check_and_return_ocl_errcode (errcode);
+
+ ocl->all_devices = g_malloc0 (ocl->num_all_devices * sizeof (cl_device_id));
+ errcode = clGetDeviceIDs (platform, CL_DEVICE_TYPE_ALL, ocl->num_all_devices, ocl->all_devices, NULL);
+ check_and_return_ocl_errcode (errcode);
+
+ /* Create context spanning devices within range from:to */
+ to = MIN (to, ocl->num_all_devices - 1);
+ ocl->num_devices = to - from + 1;
+ ocl->devices = g_malloc0 (ocl->num_devices * sizeof (cl_device_id));
+
+ for (guint i = from; i < to + 1; i++)
+ ocl->devices[i - from] = ocl->all_devices[i];
+
+ ocl->context = clCreateContext (NULL, ocl->num_devices, ocl->devices, NULL, NULL, &errcode);
+ check_and_return_ocl_errcode (errcode);
+ ocl->cmd_queues = g_malloc0(ocl->num_devices * sizeof (cl_command_queue));
+ cl_command_queue_properties queue_properties = CL_QUEUE_PROFILING_ENABLE;
+
+ /* Print platform info */
+ errcode = clGetPlatformInfo (platform, CL_PLATFORM_VERSION, len, string_buffer, NULL);
+ check_and_return_ocl_errcode (errcode);
+ g_print ("# Platform: %s\n", string_buffer);
+
+ for (int i = 0; i < ocl->num_devices; i++) {
+ clGetDeviceInfo (ocl->devices[i], CL_DEVICE_NAME, len, string_buffer, NULL);
+ g_print ("# Device %i: %s\n", i, string_buffer);
+
+ ocl->cmd_queues[i] = clCreateCommandQueue (ocl->context, ocl->devices[i], queue_properties, &errcode);
+ check_and_return_ocl_errcode (errcode);
+ }
+
+ return ocl;
+}
+
+void
+ocl_free (OCL *ocl)
+{
+ for (int i = 0; i < ocl->num_devices; i++)
+ clReleaseCommandQueue (ocl->cmd_queues[i]);
+
+ clReleaseContext (ocl->context);
+
+ g_free (ocl->all_devices);
+ g_free (ocl->devices);
+ g_free (ocl->cmd_queues);
+ g_free (ocl);
+}
+
+static const gchar *error_msgs[] = {
+ "CL_SUCCESS",
+ "CL_DEVICE_NOT_FOUND",
+ "CL_DEVICE_NOT_AVAILABLE",
+ "CL_COMPILER_NOT_AVAILABLE",
+ "CL_MEM_OBJECT_ALLOCATION_FAILURE",
+ "CL_OUT_OF_RESOURCES",
+ "CL_OUT_OF_HOST_MEMORY",
+ "CL_PROFILING_INFO_NOT_AVAILABLE",
+ "CL_MEM_COPY_OVERLAP",
+ "CL_IMAGE_FORMAT_MISMATCH",
+ "CL_IMAGE_FORMAT_NOT_SUPPORTED",
+ "CL_BUILD_PROGRAM_FAILURE",
+ "CL_MAP_FAILURE",
+ "CL_MISALIGNED_SUB_BUFFER_OFFSET",
+ "CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST",
+
+ /* next IDs start at 30! */
+ "CL_INVALID_VALUE",
+ "CL_INVALID_DEVICE_TYPE",
+ "CL_INVALID_PLATFORM",
+ "CL_INVALID_DEVICE",
+ "CL_INVALID_CONTEXT",
+ "CL_INVALID_QUEUE_PROPERTIES",
+ "CL_INVALID_COMMAND_QUEUE",
+ "CL_INVALID_HOST_PTR",
+ "CL_INVALID_MEM_OBJECT",
+ "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR",
+ "CL_INVALID_IMAGE_SIZE",
+ "CL_INVALID_SAMPLER",
+ "CL_INVALID_BINARY",
+ "CL_INVALID_BUILD_OPTIONS",
+ "CL_INVALID_PROGRAM",
+ "CL_INVALID_PROGRAM_EXECUTABLE",
+ "CL_INVALID_KERNEL_NAME",
+ "CL_INVALID_KERNEL_DEFINITION",
+ "CL_INVALID_KERNEL",
+ "CL_INVALID_ARG_INDEX",
+ "CL_INVALID_ARG_VALUE",
+ "CL_INVALID_ARG_SIZE",
+ "CL_INVALID_KERNEL_ARGS",
+ "CL_INVALID_WORK_DIMENSION",
+ "CL_INVALID_WORK_GROUP_SIZE",
+ "CL_INVALID_WORK_ITEM_SIZE",
+ "CL_INVALID_GLOBAL_OFFSET",
+ "CL_INVALID_EVENT_WAIT_LIST",
+ "CL_INVALID_EVENT",
+ "CL_INVALID_OPERATION",
+ "CL_INVALID_GL_OBJECT",
+ "CL_INVALID_BUFFER_SIZE",
+ "CL_INVALID_MIP_LEVEL",
+ "CL_INVALID_GLOBAL_WORK_SIZE"
+};
+
34 ocl.h
@@ -0,0 +1,34 @@
+#ifndef OCL_H
+#define OCL_H
+
+#include <CL/cl.h>
+#include <glib.h>
+
+#define OCL_ALL_DEVICES 0xFFFF
+
+typedef struct _OCL {
+ cl_context context;
+ cl_uint num_all_devices;
+ cl_device_id *all_devices;
+ cl_uint num_devices;
+ cl_device_id *devices;
+ cl_command_queue *cmd_queues;
+} OCL;
+
+OCL *ocl_new (guint device_from,
+ guint device_to,
+ int *ocl_errcode);
+void ocl_free (OCL *ocl);
+cl_program ocl_program_new_from_source
+ (OCL *ocl,
+ const gchar *source,
+ const gchar *options,
+ int *ocl_errcode);
+cl_program ocl_program_new_from_file
+ (OCL *ocl,
+ const gchar *filename,
+ const gchar *options,
+ int *ocl_errcode);
+const gchar *ocl_error (int ocl_errcode);
+
+#endif
23 programs.cl
@@ -0,0 +1,23 @@
+__kernel void assign(__global float *input, __global float *output)
+{
+ const int idx = get_global_id (0);
+ input[idx] = output[idx];
+}
+
+__kernel void two_const_params(__constant float *c_param_1,
+ __constant float *c_param_2)
+{
+}
+
+__kernel void three_const_params(__constant float *c_param_1,
+ __constant float *c_param_2,
+ __constant float *c_param_3)
+{
+}
+
+__kernel void four_const_params(__constant float *c_param_1,
+ __constant float *c_param_2,
+ __constant float *c_param_3,
+ __constant float *c_param_4)
+{
+}
BIN waf
Binary file not shown.
51 wscript
@@ -0,0 +1,51 @@
+SOURCES = [
+ 'ocl.c',
+ 'check.c',
+]
+
+OPENCL_INC_PATHS = [
+ '/usr/local/cuda/include',
+ '/opt/cuda/include'
+]
+
+
+def guess_cl_include_path():
+ import os
+
+ try:
+ OPENCL_INC_PATHS.append(os.environ['CUDA_INC_PATH'])
+ except:
+ pass
+
+ return filter(lambda d: os.path.exists(d), OPENCL_INC_PATHS)
+
+
+def options(opt):
+ opt.load('compiler_c')
+
+
+def configure(conf):
+ conf.load('compiler_c')
+ conf.env.append_unique('CFLAGS', ['-g', '-ggdb', '-std=c99', '-O3', '-Wall', '-Werror' ])
+
+ # Check OpenCL include paths ...
+ conf.start_msg('Checking for OpenCL include path')
+ incs = guess_cl_include_path()
+
+ if incs:
+ conf.env.OPENCL_INC_PATH = incs[0]
+ conf.end_msg('yes')
+ else:
+ conf.fatal('OpenCL include path not found')
+
+ # ... and library
+ conf.check_cc(lib='OpenCL', uselib_store='CL')
+
+ # Check libraries using pkg-config
+ conf.check_cfg(package='glib-2.0', args='--cflags --libs', uselib_store='GLIB2')
+
+
+def build(bld):
+ srcs = ' '.join(SOURCES)
+ incs = ' '.join(OPENCL_INC_PATHS)
+ bld.program(source=srcs, target='check', use=['GLIB2', 'CL'], includes=bld.env.OPENCL_INC_PATH)

0 comments on commit d8bcdf7

Please sign in to comment.
Something went wrong with that request. Please try again.