diff --git a/.gitignore b/.gitignore index f88fb90..3cf923b 100644 --- a/.gitignore +++ b/.gitignore @@ -1,6 +1,7 @@ -CL +CL/ *.o *.swp mandelclassic clfract +clfractinteractive test diff --git a/CL b/CL deleted file mode 120000 index ec67360..0000000 --- a/CL +++ /dev/null @@ -1 +0,0 @@ -/opt/intel/opencl-1.2-3.0.67279/include/CL \ No newline at end of file diff --git a/Makefile b/Makefile index 57f4d68..523947d 100644 --- a/Makefile +++ b/Makefile @@ -10,15 +10,16 @@ CFLAGS=-c -Wall -O2 $(SDLFLAGS) # Libs! SDLLIBS=$(shell sdl2-config --libs) -lSDL2_ttf -OPENCLLIBS=-lOpenCL +OPENCLLIBS=-lOpenCL -L/opt/AMDAPPSDK-3.0/lib/x86_64 LIBS=-lm -lpthread $(SDLLIBS) # Includes! -INCLUDE=-I/usr/include/SDL2 -I./ +INCLUDE=-I/usr/include/SDL2 -I./ -I/opt/intel/opencl-sdk/include -all: mandelclassic clfract test clfractinteractive +# all: mandelclassic clfract test clfractinteractive +all: mandelclassic clfract clfractinteractive mandelclassic: mandel_classic.o $(CC) $(INCLUDE) mandel_classic.o $(LIBS) -o mandelclassic diff --git a/main.c b/main.c index 4521820..b99c5db 100644 --- a/main.c +++ b/main.c @@ -16,7 +16,7 @@ #define MAX_SOURCE_SIZE (0x100000) int main(int argn, char **argv) { - + // Init SDL if(SDL_Init(SDL_INIT_VIDEO) != 0) fprintf(stderr, "Could not initialize SDL: %s\n", SDL_GetError()); @@ -64,9 +64,9 @@ int main(int argn, char **argv) { //Initialize SDL_ttf if( TTF_Init() == -1 ) - { + { printf("Error setting up TTF module.\n"); - return 1; + return 1; } // Load a font @@ -79,7 +79,7 @@ int main(int argn, char **argv) { return 1; } - //The color of the font + //The color of the font SDL_Color textColor = { 255, 255, 255 }; // Prepare the resolution and sizes and colors... @@ -105,20 +105,20 @@ int main(int argn, char **argv) { // Get platform and device information cl_platform_id platform_id = NULL; - cl_device_id device_id = NULL; + cl_device_id device_id = NULL; cl_uint ret_num_devices; cl_uint ret_num_platforms; cl_int ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms); - ret = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_GPU, 1, + ret = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_GPU, 1, &device_id, &ret_num_devices); // Create an OpenCL context cl_context context = clCreateContext( NULL, 1, &device_id, NULL, NULL, &ret); // Create a command queue - cl_command_queue command_queue = clCreateCommandQueue(context, device_id, 0, &ret); + cl_command_queue command_queue = clCreateCommandQueueWithProperties(context, device_id, 0, &ret); - // Create memory buffers on the device for returning iterations + // Create memory buffers on the device for returning iterations // Input parameters cl_mem kernel_res_x = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(int), &res_x, &ret); @@ -127,36 +127,49 @@ int main(int argn, char **argv) { cl_mem kernel_current_line = clCreateBuffer(context, CL_MEM_READ_ONLY, sizeof(int), NULL, &ret); cl_mem kernel_zoom_level = clCreateBuffer(context, CL_MEM_READ_ONLY, - sizeof(float), NULL, &ret); + sizeof(float), NULL, &ret); // Output buffer - cl_mem graph_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY, + cl_mem graph_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY, res_x * sizeof(int), NULL, &ret); // Create a program from the kernel source - cl_program program = clCreateProgramWithSource(context, 1, + cl_program program = clCreateProgramWithSource(context, 1, (const char **)&source_str, (const size_t *)&source_size, &ret); // Build the program ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL); - // Check if it is correct - printf("clBuildProgram\n"); - cl_build_status build_status; - ret = clGetProgramBuildInfo(program, device_id, CL_PROGRAM_BUILD_STATUS, sizeof(cl_build_status), &build_status, NULL); + if (ret != CL_SUCCESS) { + printf("Program build failed, error code: %d", ret); - char *build_log; - size_t ret_val_size; - ret = clGetProgramBuildInfo(program, device_id, CL_PROGRAM_BUILD_LOG, 0, NULL, &ret_val_size); + // Check if it is correct + printf("clBuildProgram\n"); + cl_build_status build_status; + ret = clGetProgramBuildInfo(program, device_id, CL_PROGRAM_BUILD_STATUS, sizeof(cl_build_status), &build_status, NULL); + + char *build_log; + size_t ret_val_size; + ret = clGetProgramBuildInfo(program, device_id, CL_PROGRAM_BUILD_LOG, 0, NULL, &ret_val_size); + + build_log = (char *) malloc((ret_val_size + 1) * sizeof(char)); + ret = clGetProgramBuildInfo(program, device_id, CL_PROGRAM_BUILD_LOG, ret_val_size, build_log, NULL); + build_log[ret_val_size] = '\0'; + + printf("BUILD LOG: \n %s", build_log); + exit(1); + + } - build_log = (char *) malloc((ret_val_size + 1) * sizeof(char)); - ret = clGetProgramBuildInfo(program, device_id, CL_PROGRAM_BUILD_LOG, ret_val_size, build_log, NULL); - build_log[ret_val_size] = '\0'; - printf("BUILD LOG: \n %s", build_log); printf("program built\n"); // Create the OpenCL kernel cl_kernel kernel = clCreateKernel(program, "fractal_point", &ret); + if (ret != CL_SUCCESS) { + printf("Error when loading the kernel: %d", ret); + exit(1); + } + // Common kernel params ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *) &kernel_res_x); ret = clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *) &kernel_res_y); @@ -173,7 +186,7 @@ int main(int argn, char **argv) { clock_t start = clock(); - while(zoom > stop_point) + while(zoom > stop_point) { for (current_line = 0; current_line < res_y; current_line++) { @@ -187,25 +200,34 @@ int main(int argn, char **argv) { clFinish(command_queue); ret = clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *) &kernel_current_line); + if (ret != CL_SUCCESS) { + printf("Error setting current_line %d", ret); + exit(1); + } ret = clSetKernelArg(kernel, 3, sizeof(cl_mem), (void *) &kernel_zoom_level); - + if (ret != CL_SUCCESS) { + printf("Error setting zoom_level %d", ret); + exit(1); + } + // Execute the OpenCL kernel on the list size_t global_item_size = res_x; // Process the entire line - size_t local_item_size = 32; // Process in groups of 64 - ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, + size_t local_item_size = 16; // Process in groups of 64 + ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, &global_item_size, &local_item_size, 0, NULL, NULL); if (ret != CL_SUCCESS) { - // printf("Error while executing kernel\n"); - // printf("Error code %d\n", ret); + printf("Error while executing kernel\n"); + printf("Error code %d\n", ret); + exit(1); } // Wait for the computation to finish clFinish(command_queue); // Read the memory buffer graph_mem_obj on the device to the local variable graph_dots - ret = clEnqueueReadBuffer(command_queue, graph_mem_obj, CL_TRUE, 0, + ret = clEnqueueReadBuffer(command_queue, graph_mem_obj, CL_TRUE, 0, res_x * sizeof(int), graph_line, 0, NULL, NULL); if (ret != CL_SUCCESS) @@ -250,7 +272,7 @@ int main(int argn, char **argv) { 0, 0, 255); - } + } } }