diff --git a/OpenCL_Hello_World_Example/.gitignore b/OpenCL_Hello_World_Example/.gitignore new file mode 100644 index 0000000..ce01362 --- /dev/null +++ b/OpenCL_Hello_World_Example/.gitignore @@ -0,0 +1 @@ +hello diff --git a/OpenCL_Hello_World_Example/hello.c b/OpenCL_Hello_World_Example/hello.c index 49350af..3b12ed7 100644 --- a/OpenCL_Hello_World_Example/hello.c +++ b/OpenCL_Hello_World_Example/hello.c @@ -1,3 +1,5 @@ +// build and run with: +// (cd OpenCL_Hello_World_Example/; cc -o hello hello.c -lOpenCL -lm && ./hello y 1024) // // File: hello.c // @@ -50,6 +52,8 @@ //////////////////////////////////////////////////////////////////////////////// +#define CL_TARGET_OPENCL_VERSION 200 /* clCreateCommandQueueWithProperties */ + #include #include #include @@ -58,13 +62,11 @@ #include #include #include +#ifdef __APPLE__ #include - -//////////////////////////////////////////////////////////////////////////////// - -// Use a static data size for simplicity -// -#define DATA_SIZE (1024) +#else +#include +#endif //////////////////////////////////////////////////////////////////////////////// @@ -77,19 +79,129 @@ const char *KernelSource = "\n" \ " const unsigned int count) \n" \ "{ \n" \ " int i = get_global_id(0); \n" \ -" if(i < count) \n" \ -" output[i] = input[i] * input[i]; \n" \ +" if (i < count) \n" \ +" output[i] = sin(input[i]); \n" \ "} \n" \ "\n"; +const char* clGetErrorString(int errorCode) { + switch (errorCode) { + case 0: return "CL_SUCCESS"; + case -1: return "CL_DEVICE_NOT_FOUND"; + case -2: return "CL_DEVICE_NOT_AVAILABLE"; + case -3: return "CL_COMPILER_NOT_AVAILABLE"; + case -4: return "CL_MEM_OBJECT_ALLOCATION_FAILURE"; + case -5: return "CL_OUT_OF_RESOURCES"; + case -6: return "CL_OUT_OF_HOST_MEMORY"; + case -7: return "CL_PROFILING_INFO_NOT_AVAILABLE"; + case -8: return "CL_MEM_COPY_OVERLAP"; + case -9: return "CL_IMAGE_FORMAT_MISMATCH"; + case -10: return "CL_IMAGE_FORMAT_NOT_SUPPORTED"; + case -12: return "CL_MAP_FAILURE"; + case -13: return "CL_MISALIGNED_SUB_BUFFER_OFFSET"; + case -14: return "CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST"; + case -15: return "CL_COMPILE_PROGRAM_FAILURE"; + case -16: return "CL_LINKER_NOT_AVAILABLE"; + case -17: return "CL_LINK_PROGRAM_FAILURE"; + case -18: return "CL_DEVICE_PARTITION_FAILED"; + case -19: return "CL_KERNEL_ARG_INFO_NOT_AVAILABLE"; + case -30: return "CL_INVALID_VALUE"; + case -31: return "CL_INVALID_DEVICE_TYPE"; + case -32: return "CL_INVALID_PLATFORM"; + case -33: return "CL_INVALID_DEVICE"; + case -34: return "CL_INVALID_CONTEXT"; + case -35: return "CL_INVALID_QUEUE_PROPERTIES"; + case -36: return "CL_INVALID_COMMAND_QUEUE"; + case -37: return "CL_INVALID_HOST_PTR"; + case -38: return "CL_INVALID_MEM_OBJECT"; + case -39: return "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR"; + case -40: return "CL_INVALID_IMAGE_SIZE"; + case -41: return "CL_INVALID_SAMPLER"; + case -42: return "CL_INVALID_BINARY"; + case -43: return "CL_INVALID_BUILD_OPTIONS"; + case -44: return "CL_INVALID_PROGRAM"; + case -45: return "CL_INVALID_PROGRAM_EXECUTABLE"; + case -46: return "CL_INVALID_KERNEL_NAME"; + case -47: return "CL_INVALID_KERNEL_DEFINITION"; + case -48: return "CL_INVALID_KERNEL"; + case -49: return "CL_INVALID_ARG_INDEX"; + case -50: return "CL_INVALID_ARG_VALUE"; + case -51: return "CL_INVALID_ARG_SIZE"; + case -52: return "CL_INVALID_KERNEL_ARGS"; + case -53: return "CL_INVALID_WORK_DIMENSION"; + case -54: return "CL_INVALID_WORK_GROUP_SIZE"; + case -55: return "CL_INVALID_WORK_ITEM_SIZE"; + case -56: return "CL_INVALID_GLOBAL_OFFSET"; + case -57: return "CL_INVALID_EVENT_WAIT_LIST"; + case -58: return "CL_INVALID_EVENT"; + case -59: return "CL_INVALID_OPERATION"; + case -60: return "CL_INVALID_GL_OBJECT"; + case -61: return "CL_INVALID_BUFFER_SIZE"; + case -62: return "CL_INVALID_MIP_LEVEL"; + case -63: return "CL_INVALID_GLOBAL_WORK_SIZE"; + case -64: return "CL_INVALID_PROPERTY"; + case -65: return "CL_INVALID_IMAGE_DESCRIPTOR"; + case -66: return "CL_INVALID_COMPILER_OPTIONS"; + case -67: return "CL_INVALID_LINKER_OPTIONS"; + case -68: return "CL_INVALID_DEVICE_PARTITION_COUNT"; + case -69: return "CL_INVALID_PIPE_SIZE"; + case -70: return "CL_INVALID_DEVICE_QUEUE"; + case -71: return "CL_INVALID_SPEC_ID"; + case -72: return "CL_MAX_SIZE_RESTRICTION_EXCEEDED"; + case -1002: return "CL_INVALID_D3D10_DEVICE_KHR"; + case -1003: return "CL_INVALID_D3D10_RESOURCE_KHR"; + case -1004: return "CL_D3D10_RESOURCE_ALREADY_ACQUIRED_KHR"; + case -1005: return "CL_D3D10_RESOURCE_NOT_ACQUIRED_KHR"; + case -1006: return "CL_INVALID_D3D11_DEVICE_KHR"; + case -1007: return "CL_INVALID_D3D11_RESOURCE_KHR"; + case -1008: return "CL_D3D11_RESOURCE_ALREADY_ACQUIRED_KHR"; + case -1009: return "CL_D3D11_RESOURCE_NOT_ACQUIRED_KHR"; + case -1010: return "CL_INVALID_DX9_MEDIA_ADAPTER_KHR"; + case -1011: return "CL_INVALID_DX9_MEDIA_SURFACE_KHR"; + case -1012: return "CL_DX9_MEDIA_SURFACE_ALREADY_ACQUIRED_KHR"; + case -1013: return "CL_DX9_MEDIA_SURFACE_NOT_ACQUIRED_KHR"; + case -1093: return "CL_INVALID_EGL_OBJECT_KHR"; + case -1092: return "CL_EGL_RESOURCE_NOT_ACQUIRED_KHR"; + case -1001: return "CL_PLATFORM_NOT_FOUND_KHR"; + case -1057: return "CL_DEVICE_PARTITION_FAILED_EXT"; + case -1058: return "CL_INVALID_PARTITION_COUNT_EXT"; + case -1059: return "CL_INVALID_PARTITION_NAME_EXT"; + case -1094: return "CL_INVALID_ACCELERATOR_INTEL"; + case -1095: return "CL_INVALID_ACCELERATOR_TYPE_INTEL"; + case -1096: return "CL_INVALID_ACCELERATOR_DESCRIPTOR_INTEL"; + case -1097: return "CL_ACCELERATOR_TYPE_NOT_SUPPORTED_INTEL"; + case -1000: return "CL_INVALID_GL_SHAREGROUP_REFERENCE_KHR"; + case -1098: return "CL_INVALID_VA_API_MEDIA_ADAPTER_INTEL"; + case -1099: return "CL_INVALID_VA_API_MEDIA_SURFACE_INTEL"; + case -1100: return "CL_VA_API_MEDIA_SURFACE_ALREADY_ACQUIRED_INTEL"; + case -1101: return "CL_VA_API_MEDIA_SURFACE_NOT_ACQUIRED_INTEL"; + default: return "CL_UNKNOWN_ERROR"; + } +} + //////////////////////////////////////////////////////////////////////////////// int main(int argc, char** argv) { + if (argc != 3) { + fprintf(stderr, "Usage: %s \n", argv[0]); + return 1; + } + const int useGPU = argv[1][0] == 'y'; + const int nitems = atoi(argv[2]); + int err; // error code returned from api calls - float data[DATA_SIZE]; // original data set given to device - float results[DATA_SIZE]; // results returned from device + float *data = malloc(sizeof(float) * nitems); // original data set given to device + if (!data) { + printf("Error: Failed to allocate data\n"); + exit(1); + } + float *results = malloc(sizeof(float) * nitems); // results returned from device + if (!results) { + printf("Error: Failed to allocate results\n"); + exit(1); + } unsigned int correct; // number of correct results returned size_t global; // global domain size for our calculation @@ -107,14 +219,13 @@ int main(int argc, char** argv) // Fill our data set with random float values // int i = 0; - unsigned int count = DATA_SIZE; - for(i = 0; i < count; i++) + unsigned int count = nitems; + for (i = 0; i < count; i++) data[i] = rand() / (float)RAND_MAX; // Connect to a compute device // - int gpu = 1; - err = clGetDeviceIDs(NULL, gpu ? CL_DEVICE_TYPE_GPU : CL_DEVICE_TYPE_CPU, 1, &device_id, NULL); + err = clGetDeviceIDs(NULL, useGPU ? CL_DEVICE_TYPE_GPU : CL_DEVICE_TYPE_CPU, 1, &device_id, NULL); if (err != CL_SUCCESS) { printf("Error: Failed to create a device group!\n"); @@ -132,10 +243,10 @@ int main(int argc, char** argv) // Create a command commands // - commands = clCreateCommandQueue(context, device_id, 0, &err); + commands = clCreateCommandQueueWithProperties(context, device_id, NULL, &err); if (!commands) { - printf("Error: Failed to create a command commands!\n"); + printf("Error: Failed to create a command queue: %s\n", clGetErrorString(err)); return EXIT_FAILURE; } @@ -211,6 +322,27 @@ int main(int argc, char** argv) exit(1); } + cl_uint max_wi_ndims; + err = clGetDeviceInfo(device_id, CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS, sizeof(max_wi_ndims), &max_wi_ndims, NULL); + if (err) + { + printf("Error: Failed to get CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS: %s\n", clGetErrorString(err)); + return EXIT_FAILURE; + } + + size_t max_wi_dims[max_wi_ndims]; + err = clGetDeviceInfo(device_id, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(max_wi_dims), max_wi_dims, NULL); + if (err) + { + printf("Error: Failed to get CL_DEVICE_MAX_WORK_ITEM_SIZES: %s\n", clGetErrorString(err)); + return EXIT_FAILURE; + } + + if (local > count) /* don't ask for more than all */ + local = count; + if (local > max_wi_dims[0]) /* don't ask for more than max dim */ + local = max_wi_dims[0]; + // Execute the kernel over the entire range of our 1d input data set // using the maximum number of work group items for this device // @@ -218,7 +350,7 @@ int main(int argc, char** argv) err = clEnqueueNDRangeKernel(commands, kernel, 1, NULL, &global, &local, 0, NULL, NULL); if (err) { - printf("Error: Failed to execute kernel!\n"); + printf("Error: Failed to execute kernel: %s\n", clGetErrorString(err)); return EXIT_FAILURE; } @@ -238,9 +370,9 @@ int main(int argc, char** argv) // Validate our results // correct = 0; - for(i = 0; i < count; i++) + for (i = 0; i < count; i++) { - if(results[i] == data[i] * data[i]) + if (fabsf(results[i] - sin(data[i])) < 1e-6) correct++; } @@ -256,7 +388,8 @@ int main(int argc, char** argv) clReleaseKernel(kernel); clReleaseCommandQueue(commands); clReleaseContext(context); + free(data); + free(results); return 0; } - diff --git a/tools-master/cl-demo.c b/tools-master/cl-demo.c index f57bee0..91f43c8 100644 --- a/tools-master/cl-demo.c +++ b/tools-master/cl-demo.c @@ -1,15 +1,18 @@ +#define CL_TARGET_OPENCL_VERSION 300 + #include "timing.h" #include "cl-helper.h" +#define KERNEL_FILE "vec-add-soln.cl" int main(int argc, char **argv) { if (argc != 3) { - fprintf(stderr, "need two arguments!\n"); - abort(); + fprintf(stderr, "Usage: %s \n", argv[0]); + return 1; } const cl_long n = atol(argv[1]); @@ -17,14 +20,20 @@ int main(int argc, char **argv) cl_context ctx; cl_command_queue queue; - create_context_on(CHOOSE_INTERACTIVELY, CHOOSE_INTERACTIVELY, 0, &ctx, &queue, 0); + if (!create_context_on(CHOOSE_INTERACTIVELY, CHOOSE_INTERACTIVELY, 0, &ctx, &queue, 0)) + return 1; - print_device_info_from_queue(queue); + if (!print_device_info_from_queue(queue)) + return 1; // -------------------------------------------------------------------------- - // load kernels + // load kernels // -------------------------------------------------------------------------- - char *knl_text = read_file("vec-add-soln.cl"); + char *knl_text = read_file(KERNEL_FILE); + if (!knl_text) { + fprintf(stderr, "error with '%s'\n", KERNEL_FILE); + return 1; + } cl_kernel knl = kernel_from_string(ctx, knl_text, "sum", NULL); free(knl_text); @@ -32,11 +41,11 @@ int main(int argc, char **argv) // allocate and initialize CPU memory // -------------------------------------------------------------------------- float *a = (float *) malloc(sizeof(float) * n); - if (!a) { perror("alloc x"); abort(); } + if (!a) { perror("alloc x"); return 1; } float *b = (float *) malloc(sizeof(float) * n); - if (!b) { perror("alloc y"); abort(); } + if (!b) { perror("alloc y"); return 1; } float *c = (float *) malloc(sizeof(float) * n); - if (!c) { perror("alloc z"); abort(); } + if (!c) { perror("alloc z"); return 1; } for (size_t i = 0; i < n; ++i) { @@ -113,7 +122,7 @@ int main(int argc, char **argv) if (c[i] != 3*i) { printf("BAD %ld %f %f!\n", i, c[i], c[i] - 3*i); - abort(); + return 1; } puts("GOOD"); diff --git a/tools-master/cl-helper.c b/tools-master/cl-helper.c index 0642a1b..d93f9c9 100644 --- a/tools-master/cl-helper.c +++ b/tools-master/cl-helper.c @@ -22,6 +22,7 @@ +#define CL_TARGET_OPENCL_VERSION 300 #include "cl-helper.h" #include @@ -106,7 +107,7 @@ const char *cl_error_to_str(cl_int e) -void print_platforms_devices() +void *print_platforms_devices() { // get number of platforms cl_uint plat_count; @@ -207,7 +208,7 @@ const char *CHOOSE_INTERACTIVELY = "INTERACTIVE"; #define MIN(a,b) (((a)<(b))?(a):(b)) #define MAX(a,b) (((a)>(b))?(a):(b)) -void create_context_on(const char *plat_name, const char*dev_name, cl_uint idx, +void *create_context_on(const char *plat_name, const char*dev_name, cl_uint idx, cl_context *ctx, cl_command_queue *queue, int enable_profiling) { char dev_sel_buf[MAX_NAME_LEN]; @@ -352,7 +353,7 @@ void create_context_on(const char *plat_name, const char*dev_name, cl_uint idx, CHECK_CL_ERROR(status, "clCreateCommandQueue"); } - return; + return (void *)1; } else --idx; @@ -366,7 +367,7 @@ void create_context_on(const char *plat_name, const char*dev_name, cl_uint idx, free(platforms); fputs("create_context_on: specified device not found.\n", stderr); - abort(); + return NULL; } @@ -374,7 +375,7 @@ void create_context_on(const char *plat_name, const char*dev_name, cl_uint idx, char *read_file(const char *filename) { - FILE *f = fopen(filename, "r"); + FILE *f = fopen(filename, "rb"); CHECK_SYS_ERROR(!f, "read_file: opening file"); // figure out file size @@ -433,7 +434,7 @@ cl_kernel kernel_from_string(cl_context ctx, 0, NULL, &log_size)); bool do_print = status != CL_SUCCESS; - if (!do_print && log_size) + if (!do_print && log_size > 1) /* 1 is the null-terminator */ { if (getenv("CL_HELPER_PRINT_COMPILER_OUTPUT")) do_print = true; @@ -441,11 +442,11 @@ cl_kernel kernel_from_string(cl_context ctx, { if (!printed_compiler_output_message && !getenv("CL_HELPER_NO_COMPILER_OUTPUT_NAG")) { - fprintf(stderr, "*** Kernel compilation resulted in non-empty log message.\n" + fprintf(stderr, "*** Kernel compilation resulted in non-empty log message (length=%zd).\n" "*** Set environment variable CL_HELPER_PRINT_COMPILER_OUTPUT=1 to see more.\n" "*** NOTE: this may include compiler warnings and other important messages\n" "*** about your code.\n" - "*** Set CL_HELPER_NO_COMPILER_OUTPUT_NAG=1 to disable this message.\n"); + "*** Set CL_HELPER_NO_COMPILER_OUTPUT_NAG=1 to disable this message.\n", log_size); printed_compiler_output_message = true; } } @@ -462,7 +463,7 @@ cl_kernel kernel_from_string(cl_context ctx, CALL_CL_GUARDED(clGetProgramBuildInfo, (program, dev, CL_PROGRAM_BUILD_LOG, log_size, log, NULL)); - fprintf(stderr, "*** build of '%s' on '%s' said:\n%s\n*** (end of message)\n", + fprintf(stderr, "Error code was %s\n*** build of '%s' on '%s' said:\n%s\n*** (end of message)\n", cl_error_to_str(status), knl_name, devname, log); } } @@ -481,7 +482,7 @@ cl_kernel kernel_from_string(cl_context ctx, -void print_device_info(cl_device_id device) +void *print_device_info(cl_device_id device) { // adapted from http://graphics.stanford.edu/~yoel/notes/clInfo.c @@ -731,7 +732,7 @@ void print_device_info(cl_device_id device) -void print_device_info_from_queue(cl_command_queue queue) +void *print_device_info_from_queue(cl_command_queue queue) { cl_device_id dev; CALL_CL_GUARDED(clGetCommandQueueInfo, diff --git a/tools-master/cl-helper.h b/tools-master/cl-helper.h index a568320..3a336ea 100644 --- a/tools-master/cl-helper.h +++ b/tools-master/cl-helper.h @@ -74,7 +74,7 @@ /* An error check macro for Unix system functions. If "COND" is true, then the * last system error ("errno") is printed along with MSG, which is supposed to - * be a string describing what you were doing. + * be a string describing what you were doing, then returns NULL * * Example: * CHECK_SYS_ERROR(dave != 0, "opening hatch"); @@ -83,7 +83,7 @@ if (COND) \ { \ perror(MSG); \ - abort(); \ + return NULL; \ } /* Return a string describing the OpenCL error code 'e'. @@ -93,7 +93,7 @@ const char *cl_error_to_str(cl_int e); /* Print a list of available OpenCL platforms and devices * to standard output. */ -void print_platforms_devices(); +void *print_platforms_devices(); /* Create an OpenCL context and a matching command queue on a platform from a * vendor whose name contains 'plat_name' on a device whose name contains @@ -115,7 +115,7 @@ void print_platforms_devices(); * compiler option. */ extern const char *CHOOSE_INTERACTIVELY; -void create_context_on(const char *plat_name, const char*dev_name, cl_uint +void *create_context_on(const char *plat_name, const char*dev_name, cl_uint idx, cl_context *ctx, cl_command_queue *queue, int enable_profiling); /* Read contents of file 'filename'. @@ -142,8 +142,8 @@ cl_kernel kernel_from_string(cl_context ctx, /* Print information about a device, found from either the * queue or the device_id. */ -void print_device_info(cl_device_id device); -void print_device_info_from_queue(cl_command_queue queue); +void *print_device_info(cl_device_id device); +void *print_device_info_from_queue(cl_command_queue queue); #define SET_1_KERNEL_ARG(knl, arg0) \ CALL_CL_GUARDED(clSetKernelArg, (knl, 0, sizeof(arg0), &arg0)); diff --git a/tools-master/print-devices.c b/tools-master/print-devices.c index 390d90b..9d8adf0 100644 --- a/tools-master/print-devices.c +++ b/tools-master/print-devices.c @@ -1,7 +1,8 @@ +#define CL_TARGET_OPENCL_VERSION 300 + #include "cl-helper.h" int main(int argc, char **argv) { - print_platforms_devices(); - return 0; + return print_platforms_devices() ? 0 : 1; }