changeset 0:0e6e76dbdb0f

add file
author Yutaka_Kinjyo
date Tue, 12 Jul 2011 11:12:51 +0900
parents
children 403e35dd9b6d
files WordCount/Makefile WordCount/main.cc WordCount/oclUtils.cc WordCount/oclUtils.h WordCount/word_count.cl hello/Makefile hello/hello.cc hello/hello.cl hello/oclUtils.cc hello/oclUtils.h
diffstat 10 files changed, 714 insertions(+), 0 deletions(-) [+]
line wrap: on
line diff
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/WordCount/Makefile	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,23 @@
+TARGET= word_count
+CC = g++
+WARN = -Wall
+CFLAGS = -isysroot /Developer/SDKs/MacOSX10.6.sdk
+LIBS = -framework OpenCL #-lclsdk
+
+HEADERS = $(shell ls *.h)
+SRCS = $(shell ls *.cc)
+OBJS = $(SRCS:.cc=.o)
+
+.SUFFIXES: .cc .o
+
+.cc.o:
+	$(CC) $(CCFLAGS) $(INCLUDE) -c $< -o $@
+
+all: $(TARGET)
+
+$(TARGET): $(OBJS)
+	$(CC) -o $@ $(OBJS) $(LIBS)
+
+clean:
+	rm -f $(TARGET) $(OBJS)
+	rm -f *~ \#*
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/WordCount/main.cc	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,232 @@
+#include <stdio.h>
+#include <stdlib.h>
+#include <string.h>
+#include <sys/stat.h>
+#include <sys/mman.h>
+#include <sys/types.h>
+#include <fcntl.h>
+#include <unistd.h>
+#include <OpenCL/opencl.h>
+#include <oclUtils.h>
+
+#define OUT_PARAM_NUM 2
+
+typedef struct {
+    caddr_t file_mmap;
+    off_t size;
+} st_mmap_t;
+
+/*与えられたsizeをfix_byte_sizeの倍数にする(丸め込むっていうのかな?)*/
+static int
+fix_byte(int size,int fix_byte_size)
+{
+    size = (size/fix_byte_size)*fix_byte_size  + ((size%fix_byte_size)!= 0)*fix_byte_size;
+    
+    return size;
+}
+
+
+static st_mmap_t
+my_mmap(char *filename)
+{
+
+    /*マッピングだよ!*/
+    int fd = -1;
+    int map = MAP_PRIVATE;
+    st_mmap_t st_mmap;
+    struct stat sb;
+    
+    if ((fd=open(filename,O_RDONLY,0666))==0) {
+	fprintf(stderr,"can't open %s\n",filename);
+    }
+    
+    if (fstat(fd,&sb)) {
+	fprintf(stderr,"can't fstat %s\n",filename);
+    }
+
+    printf("file size %d\n",(int)sb.st_size);
+   
+    /*sizeをページングサイズの倍数にあわせる*/
+    st_mmap.size = fix_byte(sb.st_size,4096);
+
+    printf("fix 4096byte file size %d\n",(int)st_mmap.size);
+
+    st_mmap.file_mmap = (char*)mmap(NULL,st_mmap.size,PROT_READ,map,fd,(off_t)0);
+    if (st_mmap.file_mmap == (caddr_t)-1) {
+	fprintf(stderr,"Can't mmap file\n");
+	perror(NULL);
+	exit(0);
+    }
+
+    return st_mmap;
+
+}
+
+
+
+int main(int args, char *argv[]) 
+{
+
+    char *filename = 0;
+
+    for (int i = 1; argv[i]; ++i) {	
+	if (strcmp(argv[i], "-file") == 0) {
+	    filename = argv[i+1];
+	} else if (strcmp(argv[i], "-help")) {
+            printf("Usage: ./word_count [-file filename]\n");
+        }
+    }
+
+    if (filename == 0) {
+        printf("Usage: ./word_count [-file filename]\n");
+        return 0;
+    }
+
+    //指定されたファイルをメモリにmap
+    st_mmap_t st_mmap = my_mmap(filename);
+
+    //kernelファイルの大きさ取得して、メモリ確保
+    int fd = -1;
+    const char *kernel_filename =  "./word_count.cl";
+    if ((fd=open(kernel_filename,O_RDONLY,0666))==0) {
+        fprintf(stderr,"can't open %s\n",kernel_filename);
+    }
+
+    struct stat sb;
+    if (fstat(fd,&sb)) {
+        fprintf(stderr,"can't fstat %s\n",filename);
+    }
+
+    size_t source_size = sb.st_size;
+    void *source_str = malloc(source_size);
+
+    int err = read(fd, source_str, source_size);
+    if (err == -1) {
+        fprintf(stderr,"can't read %s\n",filename);
+    }
+
+    close(fd);
+
+    cl_platform_id platform_id = NULL;
+    cl_uint ret_num_platforms = NULL;
+    cl_int ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);
+
+    cl_device_id device_id = NULL;
+    cl_uint ret_num_devices = NULL;
+    // CL_DEVICE_TYPE_CPU, CL_DEVICE_TYPE_GPU, と指定できる
+    // CL_DEVICE_TYPE_DEFAULT はどうなるのか
+    ret = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_DEFAULT, 
+                          1, &device_id, &ret_num_devices);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    //OpenCLコンテキストの作成
+    cl_context context = clCreateContext( NULL, 1, &device_id, NULL, NULL, &ret);
+    oclCheckError(ret, CL_SUCCESS);
+    cl_command_queue command_queue = clCreateCommandQueue(context, device_id, 0, &ret);
+    oclCheckError(ret, CL_SUCCESS);
+    // カーネルプログラムを読み込む
+    cl_program program = clCreateProgramWithSource(context, 1, (const char **)&source_str,
+                                                   (const size_t *)&source_size, &ret);
+    ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    //カーネルプログラムをビルド
+    //Task選択にあたる
+    cl_kernel kernel = clCreateKernel(program, "word_count", &ret);
+    oclCheckError(ret, CL_SUCCESS);
+
+    //カウントするテキストデータのメモリオブジェクト
+    cl_mem text_memobj = clCreateBuffer(context, CL_MEM_READ_ONLY, 
+                                        st_mmap.size * sizeof(char), NULL, &ret);
+
+    ret = clEnqueueWriteBuffer(command_queue, text_memobj,
+                               CL_TRUE, 0, st_mmap.size, (char*)st_mmap.file_mmap,
+                               0, NULL, NULL);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    // 必要なパラメータのオブジェクト
+    cl_mem param_memobj = clCreateBuffer(context, CL_MEM_READ_ONLY, 
+                                         sizeof(int), NULL, &ret);
+
+    ret = clEnqueueWriteBuffer(command_queue, param_memobj,
+                               CL_TRUE, 0, sizeof(int), (int*)(&st_mmap.size),
+                               0, NULL, NULL);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    
+    /*
+     * 並列度の計算
+     * wcするファイルの大きさに合わせる
+     *
+     */
+
+    // このdivi_size はどうやって決めるよ
+    int divi_size = 1024;
+    st_mmap.size / 1024;
+    
+
+    size_t global_work_size = 4;
+
+    // 行数、単語数を格納する2のint配列
+    int out_size = sizeof(int) * OUT_PARAM_NUM * global_work_size;
+    cl_mem out_memobj = clCreateBuffer(context, CL_MEM_WRITE_ONLY, 
+                                       out_size * sizeof(char), NULL, &ret);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+
+    // 引数のSet
+    // memory object にしなくてもできるsetできるかも
+    ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&text_memobj);
+    ret |= clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *)&param_memobj);
+    ret |= clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *)&out_memobj);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    /* 
+     * kernel実行
+   * 並列に処理せずに work-item ひとつで動かしたい場合は、clEnqueueNDRangeKernel の簡易版 clEnqueueTask が使える
+     *
+    */   
+
+    //ret = clEnqueueTask(command_queue, kernel, 0, NULL, NULL);
+    // global_work_size は配列。並列動作させる時の次元数にあわせて、配列の次元数も決まるはず
+    ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, &global_work_size, NULL, 0, NULL, NULL);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    int *out_data = (int*)malloc(out_size);
+
+    // 演算結果の読み込み
+    ret = clEnqueueReadBuffer(command_queue, out_memobj, CL_TRUE, 0, 
+                              out_size * sizeof(char), out_data, 0, NULL, NULL);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    for (int i = 0; i < global_work_size; i++) {
+        for (int j = 0; j < OUT_PARAM_NUM; j++) {
+            printf("%d ", out_data[i*OUT_PARAM_NUM+j]);
+        }
+        printf("\n");
+    }
+
+    clFlush(command_queue);
+    clFinish(command_queue);
+    clReleaseKernel(kernel);
+    clReleaseProgram(program);
+    clReleaseMemObject(text_memobj);
+    clReleaseMemObject(param_memobj);
+    clReleaseMemObject(out_memobj);
+    clReleaseCommandQueue(command_queue);
+    clReleaseContext(context);
+
+    free(source_str);
+    free(out_data);
+
+    return 0;
+
+}
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/WordCount/oclUtils.cc	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,85 @@
+#include "oclUtils.h"
+#include <fstream>
+#include <vector>
+#include <iostream>
+#include <algorithm>
+#include <stdarg.h>
+
+// Helper function to get OpenCL error string from constant
+// *********************************************************************
+const char* oclErrorString(cl_int error)
+{
+    static const char* errorString[] = {
+        "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_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",
+    };
+
+    const int errorCount = sizeof(errorString) / sizeof(errorString[0]);
+
+    const int index = -error;
+
+    return (index >= 0 && index < errorCount) ? errorString[index] : "Unspecified Error";
+}
+
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/WordCount/oclUtils.h	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,50 @@
+#ifndef OCL_UTILS_H
+#define OCL_UTILS_H
+
+#include <OpenCL/opencl.h>
+#include <stdio.h>
+#include <stdlib.h>
+
+// SDK Revision #
+#define OCL_SDKREVISION "6161726"
+
+// Error and Exit Handling Macros... 
+// *********************************************************************
+// Full error handling macro with Cleanup() callback (if supplied)... 
+// (Companion Inline Function lower on page)
+#define oclCheckErrorEX(a, b, c) __oclCheckErrorEX(a, b, c, __FILE__ , __LINE__) 
+
+// Short version without Cleanup() callback pointer
+// Both Input (a) and Reference (b) are specified as args
+#define oclCheckError(a, b) oclCheckErrorEX(a, b, 0) 
+
+extern "C" const char* oclErrorString(cl_int error);
+
+inline void __oclCheckErrorEX(cl_int iSample, cl_int iReference, void (*pCleanup)(int), const char* cFile, const int iLine)
+{
+    // An error condition is defined by the sample/test value not equal to the reference
+    if (iReference != iSample)
+    {
+        // If the sample/test value isn't equal to the ref, it's an error by defnition, so override 0 sample/test value
+        iSample = (iSample == 0) ? -9999 : iSample; 
+
+        // Log the error info
+        //shrLog("\n !!! Error # %i (%s) at line %i , in file %s !!!\n\n", iSample, oclErrorString(iSample), iLine, cFile);
+
+	printf("\n !!! Error # %i (%s) at line %i , in file %s !!!\n\n", iSample, oclErrorString(iSample), iLine, cFile);
+
+        // Cleanup and exit, or just exit if no cleanup function pointer provided.  Use iSample (error code in this case) as process exit code.
+        if (pCleanup != NULL)
+        {
+            pCleanup(iSample);
+        }
+        else 
+        {
+	  //shrLogEx(LOGBOTH | CLOSELOG, 0, "Exiting...\n");
+            exit(iSample);
+        }
+    }
+}
+
+#endif
+
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/WordCount/word_count.cl	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,50 @@
+#pragma OPENCL EXTENSION cl_khr_byte_addressable_store : enable
+
+__kernel void word_count(__global char* text_data,  __global int* size, __global int* out_data)
+{
+
+  //unsigned long long *head_tail_flag = o_data +2;
+
+  uint gid = get_global_id(0);
+
+  int length = *size;
+  int word_flag = 0;
+  int word_num = 0;
+  int line_num = 0;
+  int i = 0;
+    
+  int start = gid*length/2;
+  int end = start + length/2;
+  
+  //head_tail_flag[0] = (i_data[0] != 0x20) && (i_data[0] != 0x0A);
+  //word_num -= 1-head_tail_flag[0];
+
+  for (i = 0; i < length; i++) {
+  //for (i = start; i < end; i++) {
+    if (text_data[i] == 0x20) {
+      word_flag = 1;
+    } else if (text_data[i] == 0x0A) {
+      line_num += 1;
+      word_flag = 1;
+    } else {
+      word_num += word_flag;
+      word_flag = 0;
+    }
+  }
+
+  word_num += word_flag;
+
+  //head_tail_flag[1] = (i_data[i-1] != 0x20) && (i_data[i-1] != 0x0A);
+  
+  // s->printf("SPE word %d line %d\n",word_num,line_num);
+
+  int index = gid*2;
+  
+  out_data[index] = word_num;
+  out_data[index+1] = line_num;
+
+
+  
+  return 0;
+
+}
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/hello/Makefile	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,23 @@
+TARGET= hello
+CC = g++
+WARN = -Wall
+CFLAGS = -isysroot /Developer/SDKs/MacOSX10.6.sdk
+LIBS = -framework OpenCL #-lclsdk
+
+HEADERS = $(shell ls *.h)
+SRCS = $(shell ls *.cc)
+OBJS = $(SRCS:.cc=.o)
+
+.SUFFIXES: .cc .o
+
+.cc.o:
+	$(CC) $(CCFLAGS) $(INCLUDE) -c $< -o $@
+
+all: $(TARGET)
+
+$(TARGET): $(OBJS)
+	$(CC) -o $@ $(OBJS) $(LIBS)
+
+clean:
+	rm -f $(TARGET) $(OBJS)
+	rm -f *~ \#*
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/hello/hello.cc	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,97 @@
+#include <stdio.h>
+#include <stdlib.h>
+#include <sys/stat.h>
+#include <fcntl.h>
+#include <unistd.h>
+#include <OpenCL/opencl.h>
+#include <oclUtils.h>
+
+
+#define MEM_SIZE (128)
+
+int main(int args, char *argv[]) 
+{
+
+    int fd = -1;
+    const char *filename =  "./hello.cl";
+    if ((fd=open(filename,O_RDONLY,0666))==0) {
+        fprintf(stderr,"can't open %s\n",filename);
+    }
+
+    struct stat sb;
+    if (fstat(fd,&sb)) {
+        fprintf(stderr,"can't fstat %s\n",filename);
+    }
+
+    size_t source_size = sb.st_size;
+    void *source_str = malloc(source_size);
+
+    int err = read(fd, source_str, source_size);
+    if (err == -1) {
+        fprintf(stderr,"can't read %s\n",filename);
+    }
+
+    close(fd);
+
+    cl_platform_id platform_id = NULL;
+    cl_device_id device_id = NULL;
+    cl_uint ret_num_platforms = NULL;
+    cl_uint ret_num_devices = NULL;
+
+    cl_int ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);
+
+    // CL_DEVICE_TYPE_CPU, CL_DEVICE_TYPE_GPU, と指定できる
+    // CL_DEVICE_TYPE_DEFAULT はどうなるのか
+    ret = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_DEFAULT, 
+                          1, &device_id, &ret_num_devices);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    //OpenCLコンテキストの作成
+    cl_context context = clCreateContext( NULL, 1, &device_id, NULL, NULL, &ret);
+
+    cl_command_queue command_queue = clCreateCommandQueue(context, device_id, 0, &ret);
+
+    cl_mem memobj = clCreateBuffer(context, CL_MEM_READ_WRITE, 
+                                   MEM_SIZE * sizeof(char), NULL, &ret);
+
+    // カーネルプログラムを読み込む
+    cl_program program = clCreateProgramWithSource(context, 1, (const char **)&source_str,
+                                                   (const size_t *)&source_size, &ret);
+
+    ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    //カーネルプログラムをビルド
+    cl_kernel kernel = clCreateKernel(program, "hello", &ret);
+
+    ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&memobj);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    ret = clEnqueueTask(command_queue, kernel, 0, NULL, NULL);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    char string[MEM_SIZE];
+    ret = clEnqueueReadBuffer(command_queue, memobj, CL_TRUE, 0, 
+                              MEM_SIZE * sizeof(char), string, 0, NULL, NULL);
+
+    oclCheckError(ret, CL_SUCCESS);
+
+    puts(string);
+
+    clFlush(command_queue);
+    clFinish(command_queue);
+    clReleaseKernel(kernel);
+    clReleaseProgram(program);
+    clReleaseMemObject(memobj);
+    clReleaseCommandQueue(command_queue);
+    clReleaseContext(context);
+
+    free(source_str);
+
+    return 0;
+
+}
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/hello/hello.cl	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,19 @@
+#pragma OPENCL EXTENSION cl_khr_byte_addressable_store : enable
+
+__kernel void hello(__global char* string)
+{
+   string[0] = 'H';
+   string[1] = 'e';
+   string[2] = 'l';
+   string[3] = 'l';
+   string[4] = 'o';
+   string[5] = ',';
+   string[6] = ' ';
+   string[7] = 'W';
+   string[8] = 'o';
+   string[9] = 'r';
+   string[10] = 'l';
+   string[11] = 'd';
+   string[12] = '!';
+   string[13] = '\0';
+}
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/hello/oclUtils.cc	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,85 @@
+#include "oclUtils.h"
+#include <fstream>
+#include <vector>
+#include <iostream>
+#include <algorithm>
+#include <stdarg.h>
+
+// Helper function to get OpenCL error string from constant
+// *********************************************************************
+const char* oclErrorString(cl_int error)
+{
+    static const char* errorString[] = {
+        "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_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",
+    };
+
+    const int errorCount = sizeof(errorString) / sizeof(errorString[0]);
+
+    const int index = -error;
+
+    return (index >= 0 && index < errorCount) ? errorString[index] : "Unspecified Error";
+}
+
--- /dev/null	Thu Jan 01 00:00:00 1970 +0000
+++ b/hello/oclUtils.h	Tue Jul 12 11:12:51 2011 +0900
@@ -0,0 +1,50 @@
+#ifndef OCL_UTILS_H
+#define OCL_UTILS_H
+
+#include <OpenCL/opencl.h>
+#include <stdio.h>
+#include <stdlib.h>
+
+// SDK Revision #
+#define OCL_SDKREVISION "6161726"
+
+// Error and Exit Handling Macros... 
+// *********************************************************************
+// Full error handling macro with Cleanup() callback (if supplied)... 
+// (Companion Inline Function lower on page)
+#define oclCheckErrorEX(a, b, c) __oclCheckErrorEX(a, b, c, __FILE__ , __LINE__) 
+
+// Short version without Cleanup() callback pointer
+// Both Input (a) and Reference (b) are specified as args
+#define oclCheckError(a, b) oclCheckErrorEX(a, b, 0) 
+
+extern "C" const char* oclErrorString(cl_int error);
+
+inline void __oclCheckErrorEX(cl_int iSample, cl_int iReference, void (*pCleanup)(int), const char* cFile, const int iLine)
+{
+    // An error condition is defined by the sample/test value not equal to the reference
+    if (iReference != iSample)
+    {
+        // If the sample/test value isn't equal to the ref, it's an error by defnition, so override 0 sample/test value
+        iSample = (iSample == 0) ? -9999 : iSample; 
+
+        // Log the error info
+        //shrLog("\n !!! Error # %i (%s) at line %i , in file %s !!!\n\n", iSample, oclErrorString(iSample), iLine, cFile);
+
+	printf("\n !!! Error # %i (%s) at line %i , in file %s !!!\n\n", iSample, oclErrorString(iSample), iLine, cFile);
+
+        // Cleanup and exit, or just exit if no cleanup function pointer provided.  Use iSample (error code in this case) as process exit code.
+        if (pCleanup != NULL)
+        {
+            pCleanup(iSample);
+        }
+        else 
+        {
+	  //shrLogEx(LOGBOTH | CLOSELOG, 0, "Exiting...\n");
+            exit(iSample);
+        }
+    }
+}
+
+#endif
+