From 2ac2de2ad094252a0a6fa0671af702c15729a1cd Mon Sep 17 00:00:00 2001 From: =?UTF-8?q?Mario=20H=C3=BCttel?= Date: Mon, 11 Mar 2019 21:58:38 +0100 Subject: [PATCH] Add example floating point calculation --- .gitignore | 1 + CMakeLists.txt | 2 +- main.c | 92 +++++++++++++++++++++++++++++++++++++++----------- saxpy.cl | 2 +- 4 files changed, 75 insertions(+), 22 deletions(-) diff --git a/.gitignore b/.gitignore index db885a0..faee5cc 100644 --- a/.gitignore +++ b/.gitignore @@ -1,3 +1,4 @@ *.o build +*.user.* *.user diff --git a/CMakeLists.txt b/CMakeLists.txt index 4e15ca2..b009860 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -12,5 +12,5 @@ message(STATUS "OpenCL libraries: ${OPENCL_LIBRARIES}") add_executable(${PROJECT_NAME} "main.c") configure_file("saxpy.cl" "cl_kernels/saxpy.cl" COPYONLY) -target_link_libraries(${PROJECT_NAME} ${OPENCL_LIBRARIES}) +target_link_libraries(${PROJECT_NAME} m ${OPENCL_LIBRARIES}) include_directories(${PROJECT_NAME} ${OPENCL_INCLUDE_DIR}) diff --git a/main.c b/main.c index b528e54..d7979d5 100644 --- a/main.c +++ b/main.c @@ -3,7 +3,7 @@ #include #include #include - +#include void check_error(cl_int error) { @@ -42,23 +42,43 @@ cl_program create_program(char *src, cl_context context) return program; } -float test_a[] = {21.321, 213.213, 3213.23}; -float test_b[] = {21.21, 13.213, 17.2553}; +float *test_a, *test_b; +static void prepare_data() +{ + int i; + test_a = malloc(sizeof(float)*1000000000); + test_b = malloc(sizeof(float)*1000000000); + + for (i = 0; i < 1000000000; i++) { + test_a[i] = (float)i; + test_b[i] = 0.0f; + } + +} + +float mult = 1.0f; int main(void) { - + int i; cl_int error; + cl_int ret; cl_platform_id *platform_ids = NULL; cl_device_id *device_ids = NULL; cl_uint platform_id_count = 0; cl_uint device_id_count = 0; - cl_context_properties context_properties[4]; cl_context context; cl_program program; cl_kernel kernel; + cl_command_queue queue; cl_mem a_buff, b_buff; char *source_code = NULL; + char temp_buff[100]; + size_t data_size = 1000000000; + size_t work_size = 64; + + + prepare_data(); clGetPlatformIDs(0, NULL, &platform_id_count); if (platform_id_count == 0) @@ -73,6 +93,11 @@ int main(void) printf("Platforms available: %u\n", (unsigned int)platform_id_count); + for (i = 0; i < platform_id_count; i++) { + temp_buff[0] = '\0'; + clGetPlatformInfo(platform_ids[i], CL_PLATFORM_NAME, sizeof(temp_buff), temp_buff, NULL); + printf("Platform ID %d: %s\n", i, temp_buff); + } clGetDeviceIDs(platform_ids[0], CL_DEVICE_TYPE_ALL, 0, NULL, &device_id_count); @@ -88,16 +113,32 @@ int main(void) clGetDeviceIDs(platform_ids[0], CL_DEVICE_TYPE_ALL, device_id_count, device_ids, NULL); - context_properties[0] = CL_CONTEXT_PLATFORM; - context_properties[1] = (cl_context_properties)platform_ids[0]; - context_properties[2] = context_properties[3] = 0; + for (i = 0; i < device_id_count; i++) { + temp_buff[0] = '\0'; + clGetDeviceInfo(device_ids[i], CL_DEVICE_NAME, sizeof(temp_buff), temp_buff, NULL); + printf("Device ID %d: %s\n", i, temp_buff); + } - context = clCreateContext(context_properties, device_id_count, - device_ids, NULL, - NULL, &error); + context = clCreateContext(NULL, 1, &device_ids[0], NULL, NULL, &error); check_error(error); + queue = clCreateCommandQueue(context, device_ids[0], 0, &error); + + + /* Create memory buffers for data */ + a_buff = clCreateBuffer(context, CL_MEM_READ_ONLY | + CL_MEM_COPY_HOST_PTR, + sizeof(test_a[0])*data_size, test_a, &error); + check_error(error); + + b_buff = clCreateBuffer(context, CL_MEM_READ_WRITE | + CL_MEM_COPY_HOST_PTR, + sizeof(test_b[0])*data_size, test_b, &error); + check_error(error); + + + load_kernel_from_file("cl_kernels/saxpy.cl", &source_code); program = create_program(source_code, context); @@ -108,17 +149,28 @@ int main(void) kernel = clCreateKernel(program, "SAXPY", &error); + /* Setting the arguments */ + clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&a_buff); + clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *)&b_buff); + clSetKernelArg(kernel, 2, sizeof(float), (void *)&mult); - /* Create memory buffers for data */ - a_buff = clCreateBuffer(context, CL_MEM_READ_ONLY | - CL_MEM_COPY_HOST_PTR, - sizeof(test_a), test_a, &error); - check_error(error); + ret = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &data_size, &work_size, 0, NULL, NULL); + check_error(ret); - b_buff = clCreateBuffer(context, CL_MEM_READ_WRITE | - CL_MEM_COPY_HOST_PTR, - sizeof(test_b), test_b, &error); - check_error(error); + clEnqueueReadBuffer(queue, b_buff, CL_TRUE, 0, 1000000000*sizeof(float), test_b, 0, NULL, NULL); + + ret = clFlush(queue); + ret = clFinish(queue); + ret = clReleaseKernel(kernel); + ret = clReleaseProgram(program); + ret = clReleaseMemObject(a_buff); + ret = clReleaseMemObject(b_buff); + ret = clReleaseCommandQueue(queue); + ret = clReleaseContext(context); + free(test_a); + free(test_b); + free(platform_ids); + free(device_ids); return 0; } diff --git a/saxpy.cl b/saxpy.cl index 2a661a0..c296777 100644 --- a/saxpy.cl +++ b/saxpy.cl @@ -1,5 +1,5 @@ __kernel void SAXPY (__global float* x, __global float* y, float a) { const int i = get_global_id (0); - y [i] += a * x [i]; + y[i] += a * x[i]; }