opencl example
This commit is contained in:
@@ -1,22 +1,36 @@
|
||||
project(mandelbrot)
|
||||
|
||||
set(CMAKE_MODULE_PATH "${CMAKE_SOURCE_DIR}/cmake")
|
||||
|
||||
|
||||
cmake_minimum_required(VERSION 2.8)
|
||||
find_package(PkgConfig REQUIRED)
|
||||
pkg_search_module(GLIB REQUIRED glib-2.0)
|
||||
pkg_check_modules(GTK3 REQUIRED gtk+-3.0)
|
||||
pkg_check_modules(CAIRO REQUIRED cairo)
|
||||
|
||||
find_package(OpenCL REQUIRED)
|
||||
|
||||
|
||||
|
||||
add_subdirectory(glade)
|
||||
|
||||
include_directories(${GLIB_INCLUDE_DIRS} ${GTK3_INCLUDE_DIRS} ${CAIRO_INCLUDE_DIRS})
|
||||
include_directories(${GLIB_INCLUDE_DIRS} ${GTK3_INCLUDE_DIRS} ${CAIRO_INCLUDE_DIRS} ${OPENCL_INCLUDE_DIRS})
|
||||
link_directories(${GLIB_LINK_DIRS} ${GTK3_LINK_DIRS} ${CAIRO_LINK_DIRS})
|
||||
add_definitions(${GLIB2_CFLAGS_OTHER})
|
||||
|
||||
configure_file("mandelbrot.cl" "cl_kernels/mandelbrot.cl" COPYONLY)
|
||||
|
||||
aux_source_directory("src" SOURCES)
|
||||
add_compile_options(-Wall)
|
||||
|
||||
add_executable(${PROJECT_NAME} ${SOURCES} ${CMAKE_CURRENT_BINARY_DIR}/glade/resources.c)
|
||||
add_dependencies(${PROJECT_NAME} glib-resources)
|
||||
SET_SOURCE_FILES_PROPERTIES(${CMAKE_CURRENT_BINARY_DIR}/glade/resources.c PROPERTIES GENERATED 1)
|
||||
target_link_libraries(${PROJECT_NAME} ${GLIB_LDFLAGS} ${GTK3_LDFLAGS} ${CAIRO_LDFLAGS} m pthread)
|
||||
target_link_libraries(${PROJECT_NAME} ${GLIB_LDFLAGS} ${GTK3_LDFLAGS} ${CAIRO_LDFLAGS} m pthread ${OPENCL_LIBRARIES})
|
||||
install (TARGETS ${PROJECT_NAME} DESTINATION bin)
|
||||
|
||||
message(STATUS "OpenCL found: ${OPENCL_FOUND}")
|
||||
message(STATUS "OpenCL includes: ${OPENCL_INCLUDE_DIRS}")
|
||||
message(STATUS "OpenCL CXX includes: ${OPENCL_HAS_CPP_BINDINGS}")
|
||||
message(STATUS "OpenCL libraries: ${OPENCL_LIBRARIES}")
|
||||
|
||||
171
cmake/FindOpenCL.cmake
Normal file
171
cmake/FindOpenCL.cmake
Normal file
@@ -0,0 +1,171 @@
|
||||
# Find OpenCL
|
||||
#
|
||||
# To set manually the paths, define these environment variables:
|
||||
# OpenCL_INCPATH - Include path (e.g. OpenCL_INCPATH=/opt/cuda/4.0/cuda/include)
|
||||
# OpenCL_LIBPATH - Library path (e.h. OpenCL_LIBPATH=/usr/lib64/nvidia)
|
||||
#
|
||||
# Once done this will define
|
||||
# OPENCL_FOUND - system has OpenCL
|
||||
# OPENCL_INCLUDE_DIRS - the OpenCL include directory
|
||||
# OPENCL_LIBRARIES - link these to use OpenCL
|
||||
# OPENCL_HAS_CPP_BINDINGS - system has also cl.hpp
|
||||
|
||||
FIND_PACKAGE(PackageHandleStandardArgs)
|
||||
|
||||
SET (OPENCL_VERSION_STRING "0.1.0")
|
||||
SET (OPENCL_VERSION_MAJOR 0)
|
||||
SET (OPENCL_VERSION_MINOR 1)
|
||||
SET (OPENCL_VERSION_PATCH 0)
|
||||
|
||||
IF (APPLE)
|
||||
|
||||
# IF OpenCL_LIBPATH is given use it and don't use default path
|
||||
IF (DEFINED ENV{OpenCL_LIBPATH})
|
||||
FIND_LIBRARY(OPENCL_LIBRARIES OpenCL PATHS ENV OpenCL_LIBPATH NO_DEFAULT_PATH)
|
||||
ELSE ()
|
||||
FIND_LIBRARY(OPENCL_LIBRARIES OpenCL DOC "OpenCL lib for OSX")
|
||||
ENDIF ()
|
||||
|
||||
# IF OpenCL_INCPATH is given use it and find for CL/cl.h and OpenCL/cl.h do not try to find default paths
|
||||
IF (DEFINED ENV{OpenCL_INCPATH})
|
||||
FIND_PATH(OPENCL_INCLUDE_DIRS CL/cl.h OpenCL/cl.h PATHS ENV OpenCL_INCPATH NO_DEFAULT_PATH)
|
||||
FIND_PATH(_OPENCL_CPP_INCLUDE_DIRS CL/cl.hpp OpenCL/cl.hpp PATHS ${OPENCL_INCLUDE_DIRS} NO_DEFAULT_PATH)
|
||||
ELSE ()
|
||||
FIND_PATH(OPENCL_INCLUDE_DIRS OpenCL/cl.h DOC "Include for OpenCL on OSX")
|
||||
FIND_PATH(_OPENCL_CPP_INCLUDE_DIRS OpenCL/cl.hpp DOC "Include for OpenCL CPP bindings on OSX")
|
||||
ENDIF ()
|
||||
|
||||
ELSE (APPLE)
|
||||
|
||||
IF (WIN32)
|
||||
|
||||
# Find OpenCL includes and libraries from environment variables provided by vendor
|
||||
SET(OPENCL_INCLUDE_SEARCH_PATHS)
|
||||
SET(OPENCL_LIBRARY_SEARCH_PATHS)
|
||||
SET(OPENCL_LIBRARY_64_SEARCH_PATHS)
|
||||
|
||||
# Nvidia
|
||||
IF (DEFINED ENV{CUDA_INC_PATH})
|
||||
SET(OPENCL_INCLUDE_SEARCH_PATHS ${OPENCL_INCLUDE_SEARCH_PATHS} $ENV{CUDA_INC_PATH})
|
||||
SET(OPENCL_LIBRARY_64_SEARCH_PATHS ${OPENCL_LIBRARY_64_SEARCH_PATHS} $ENV{CUDA_LIB_PATH}/../lib64)
|
||||
SET(OPENCL_LIBRARY_SEARCH_PATHS ${OPENCL_LIBRARY_SEARCH_PATHS} $ENV{CUDA_LIB_PATH}/../lib)
|
||||
ENDIF()
|
||||
IF (DEFINED ENV{CUDA_PATH})
|
||||
SET(OPENCL_INCLUDE_SEARCH_PATHS ${OPENCL_INCLUDE_SEARCH_PATHS} $ENV{CUDA_INC_PATH})
|
||||
SET(OPENCL_LIBRARY_64_SEARCH_PATHS ${OPENCL_LIBRARY_64_SEARCH_PATHS} $ENV{CUDA_PATH}/lib/x64/)
|
||||
SET(OPENCL_LIBRARY_SEARCH_PATHS ${OPENCL_LIBRARY_SEARCH_PATHS} $ENV{CUDA_PATH}/lib/Win32/)
|
||||
ENDIF()
|
||||
|
||||
# Intel SDK
|
||||
IF (DEFINED ENV{INTELOCSDKROOT})
|
||||
SET(OPENCL_INCLUDE_SEARCH_PATHS ${OPENCL_INCLUDE_SEARCH_PATHS} $ENV{INTELOCSDKROOT}/include)
|
||||
SET(OPENCL_LIBRARY_64_SEARCH_PATHS ${OPENCL_LIBRARY_64_SEARCH_PATHS} $ENV{INTELOCSDKROOT}/lib/x64)
|
||||
SET(OPENCL_LIBRARY_SEARCH_PATHS ${OPENCL_LIBRARY_SEARCH_PATHS} $ENV{INTELOCSDKROOT}/lib/x86)
|
||||
ENDIF()
|
||||
|
||||
# AMD SDK
|
||||
IF (DEFINED ENV{AMDAPPSDKROOT})
|
||||
SET(OPENCL_INCLUDE_SEARCH_PATHS ${OPENCL_INCLUDE_SEARCH_PATHS} $ENV{AMDAPPSDKROOT}/include)
|
||||
SET(OPENCL_LIBRARY_64_SEARCH_PATHS ${OPENCL_LIBRARY_64_SEARCH_PATHS} $ENV{AMDAPPSDKROOT}/lib/x86_64)
|
||||
SET(OPENCL_LIBRARY_SEARCH_PATHS ${OPENCL_LIBRARY_SEARCH_PATHS} $ENV{AMDAPPSDKROOT}/lib/x86)
|
||||
ENDIF()
|
||||
|
||||
# Override search paths with OpenCL_INCPATH env variable
|
||||
IF (DEFINED ENV{OpenCL_INCPATH})
|
||||
SET(OPENCL_INCLUDE_SEARCH_PATHS $ENV{OpenCL_INCPATH})
|
||||
ENDIF ()
|
||||
|
||||
# Override search paths with OpenCL_LIBPATH env variable
|
||||
IF (DEFINED ENV{OpenCL_LIBPATH})
|
||||
SET(OPENCL_LIBRARY_SEARCH_PATHS $ENV{OpenCL_LIBPATH})
|
||||
SET(OPENCL_LIBRARY_64_SEARCH_PATHS $ENV{OpenCL_LIBPATH})
|
||||
ENDIF ()
|
||||
|
||||
FIND_PATH(OPENCL_INCLUDE_DIRS CL/cl.h PATHS ${OPENCL_INCLUDE_SEARCH_PATHS})
|
||||
FIND_PATH(_OPENCL_CPP_INCLUDE_DIRS CL/cl.hpp PATHS ${OPENCL_INCLUDE_SEARCH_PATHS})
|
||||
|
||||
FIND_LIBRARY(_OPENCL_32_LIBRARIES OpenCL.lib HINTS ${OPENCL_LIBRARY_SEARCH_PATHS} PATHS ${OPENCL_LIB_DIR} ENV PATH)
|
||||
FIND_LIBRARY(_OPENCL_64_LIBRARIES OpenCL.lib HINTS ${OPENCL_LIBRARY_64_SEARCH_PATHS} PATHS ${OPENCL_LIB_DIR} ENV PATH)
|
||||
|
||||
# Check if 64bit or 32bit versions links fine
|
||||
SET (_OPENCL_VERSION_SOURCE "${CMAKE_BINARY_DIR}${CMAKE_FILES_DIRECTORY}/openclversion.c")
|
||||
#SET (_OPENCL_VERSION_SOURCE "${CMAKE_BINARY_DIR}/test.c")
|
||||
FILE (WRITE "${_OPENCL_VERSION_SOURCE}"
|
||||
"
|
||||
#if __APPLE__
|
||||
#include <OpenCL/cl.h>
|
||||
#else /* !__APPLE__ */
|
||||
#include <CL/cl.h>
|
||||
#endif /* __APPLE__ */
|
||||
int main()
|
||||
{
|
||||
cl_int result;
|
||||
cl_platform_id id;
|
||||
result = clGetPlatformIDs(1, &id, NULL);
|
||||
return result != CL_SUCCESS;
|
||||
}
|
||||
")
|
||||
|
||||
TRY_COMPILE(_OPENCL_64_COMPILE_SUCCESS ${CMAKE_BINARY_DIR} "${_OPENCL_VERSION_SOURCE}"
|
||||
CMAKE_FLAGS
|
||||
"-DINCLUDE_DIRECTORIES:STRING=${OPENCL_INCLUDE_DIRS}"
|
||||
CMAKE_FLAGS
|
||||
"-DLINK_LIBRARIES:STRING=${_OPENCL_64_LIBRARIES}"
|
||||
)
|
||||
|
||||
IF(_OPENCL_64_COMPILE_SUCCESS)
|
||||
message(STATUS "OpenCL 64bit lib found.")
|
||||
SET(OPENCL_LIBRARIES ${_OPENCL_64_LIBRARIES})
|
||||
ELSE()
|
||||
TRY_COMPILE(_OPENCL_32_COMPILE_SUCCESS ${CMAKE_BINARY_DIR} "${_OPENCL_VERSION_SOURCE}"
|
||||
CMAKE_FLAGS
|
||||
"-DINCLUDE_DIRECTORIES:STRING=${OPENCL_INCLUDE_DIRS}"
|
||||
CMAKE_FLAGS
|
||||
"-DLINK_LIBRARIES:STRING=${_OPENCL_32_LIBRARIES}"
|
||||
)
|
||||
IF(_OPENCL_32_COMPILE_SUCCESS)
|
||||
message(STATUS "OpenCL 32bit lib found.")
|
||||
SET(OPENCL_LIBRARIES ${_OPENCL_32_LIBRARIES})
|
||||
ELSE()
|
||||
message(STATUS "Couldn't link opencl..")
|
||||
ENDIF()
|
||||
ENDIF()
|
||||
|
||||
|
||||
ELSE (WIN32)
|
||||
|
||||
IF (CYGWIN)
|
||||
SET (CMAKE_FIND_LIBRARY_SUFFIXES .lib)
|
||||
SET (OCL_LIB_SUFFIX .lib)
|
||||
ENDIF (CYGWIN)
|
||||
|
||||
# Unix style platforms
|
||||
FIND_LIBRARY(OPENCL_LIBRARIES OpenCL${OCL_LIB_SUFFIX}
|
||||
PATHS ENV LD_LIBRARY_PATH ENV OpenCL_LIBPATH
|
||||
)
|
||||
|
||||
GET_FILENAME_COMPONENT(OPENCL_LIB_DIR ${OPENCL_LIBRARIES} PATH)
|
||||
GET_FILENAME_COMPONENT(_OPENCL_INC_CAND ${OPENCL_LIB_DIR}/../../include ABSOLUTE)
|
||||
|
||||
# The AMD SDK currently does not place its headers
|
||||
# in /usr/include, therefore also search relative
|
||||
# to the library
|
||||
FIND_PATH(OPENCL_INCLUDE_DIRS CL/cl.h PATHS ${_OPENCL_INC_CAND} "/usr/local/cuda/include" "/opt/AMDAPP/include" ENV OpenCL_INCPATH)
|
||||
FIND_PATH(_OPENCL_CPP_INCLUDE_DIRS CL/cl.hpp PATHS ${_OPENCL_INC_CAND} "/usr/local/cuda/include" "/opt/AMDAPP/include" ENV OpenCL_INCPATH)
|
||||
|
||||
ENDIF (WIN32)
|
||||
|
||||
ENDIF (APPLE)
|
||||
|
||||
FIND_PACKAGE_HANDLE_STANDARD_ARGS(OpenCL DEFAULT_MSG OPENCL_LIBRARIES OPENCL_INCLUDE_DIRS)
|
||||
|
||||
IF(_OPENCL_CPP_INCLUDE_DIRS)
|
||||
SET( OPENCL_HAS_CPP_BINDINGS TRUE )
|
||||
LIST( APPEND OPENCL_INCLUDE_DIRS ${_OPENCL_CPP_INCLUDE_DIRS} )
|
||||
# This is often the same, so clean up
|
||||
LIST( REMOVE_DUPLICATES OPENCL_INCLUDE_DIRS )
|
||||
ENDIF(_OPENCL_CPP_INCLUDE_DIRS)
|
||||
|
||||
MARK_AS_ADVANCED(
|
||||
OPENCL_INCLUDE_DIRS
|
||||
)
|
||||
34
mandelbrot.cl
Normal file
34
mandelbrot.cl
Normal file
@@ -0,0 +1,34 @@
|
||||
__kernel void mandelbrot(__global unsigned int *out_buff, unsigned int iter, double cx, double cy, double xs, double ys, unsigned int width, unsigned int height)
|
||||
{
|
||||
const int idx = get_global_id(0);
|
||||
|
||||
unsigned int i;
|
||||
double z_real, z_imag;
|
||||
double z2_real, z2_imag;
|
||||
double pt_real, pt_imag;
|
||||
|
||||
const unsigned int pix_x = idx % width;
|
||||
const unsigned int pix_y = (unsigned int)idx / width;
|
||||
const double val_per_x_pixel = xs / (width - 1);
|
||||
const double val_per_y_pixel = ys / (height - 1);
|
||||
|
||||
|
||||
pt_real = (((double)pix_x - (double)(width - 1) / 2)) * val_per_x_pixel + cx;
|
||||
pt_imag = (((double)pix_y - (double)(height - 1) / 2)) * val_per_y_pixel + cy;
|
||||
|
||||
z_real = 0.0;
|
||||
z_imag = 0.0;
|
||||
|
||||
for (i = 0; i < iter; i++) {
|
||||
z2_real = z_real * z_real - z_imag * z_imag;
|
||||
z2_imag = 2.0 * z_real * z_imag;
|
||||
|
||||
z_real = z2_real + pt_real;
|
||||
z_imag = z2_imag + pt_imag;
|
||||
|
||||
if ((z_real * z_real + z_imag * z_imag) >= 4.0)
|
||||
break;
|
||||
}
|
||||
|
||||
out_buff[idx] = i;
|
||||
}
|
||||
202
src/main.c
202
src/main.c
@@ -1,7 +1,9 @@
|
||||
#include <stdio.h>
|
||||
#include <gtk/gtk.h>
|
||||
#include <math.h>
|
||||
#include <threads.h>
|
||||
#include <errno.h>
|
||||
#include <stdlib.h>
|
||||
#include <CL/cl.h>
|
||||
|
||||
struct canvas_buffer {
|
||||
unsigned int *mandelbrot_buffer;
|
||||
@@ -20,6 +22,43 @@ static gboolean on_main_window_close(GtkWidget *window, gpointer data)
|
||||
return FALSE;
|
||||
}
|
||||
|
||||
void check_error(cl_int error)
|
||||
{
|
||||
if (error != CL_SUCCESS) {
|
||||
printf("OpenCL call failed with error %d\n", (int)error);
|
||||
exit(error);
|
||||
}
|
||||
}
|
||||
|
||||
void load_kernel_from_file(char *file, char **src)
|
||||
{
|
||||
long fsize;
|
||||
FILE *f = fopen(file, "rb");
|
||||
|
||||
fseek(f, 0, SEEK_END);
|
||||
fsize = ftell(f);
|
||||
fseek(f, 0, SEEK_SET); //same as rewind(f);
|
||||
*src = (char *)malloc(fsize + 1);
|
||||
fread(*src, fsize, 1, f);
|
||||
fclose(f);
|
||||
(*src)[fsize] = 0;
|
||||
}
|
||||
|
||||
cl_program create_program(char *src, cl_context context)
|
||||
{
|
||||
cl_program program;
|
||||
size_t lengths[1];
|
||||
cl_int error = 0;
|
||||
|
||||
lengths[0] = strlen(src);
|
||||
printf("strlen: %u\n", (unsigned int)lengths[0]);
|
||||
program = clCreateProgramWithSource(context, 1, (const char **)&src,
|
||||
lengths, &error);
|
||||
check_error(error);
|
||||
|
||||
return program;
|
||||
}
|
||||
unsigned int *data_array = NULL;
|
||||
static gboolean drawing_callback(GtkWidget *widget, cairo_t *cr, gpointer data)
|
||||
{
|
||||
struct canvas_buffer *buff = (struct canvas_buffer *)data;
|
||||
@@ -38,6 +77,7 @@ static gboolean drawing_callback(GtkWidget *widget, cairo_t *cr, gpointer data)
|
||||
cairo_scale(cr, 1, -1);
|
||||
cairo_translate(cr, (double)width/2.0, -(double)height/2.0);
|
||||
|
||||
|
||||
for (x = 0; x < buff->width; x++) {
|
||||
for (y = 0; y < buff->height; y++) {
|
||||
cairo_rectangle(cr,
|
||||
@@ -45,10 +85,10 @@ static gboolean drawing_callback(GtkWidget *widget, cairo_t *cr, gpointer data)
|
||||
((double)y - (double)(buff->height - 1) / 2),
|
||||
1,
|
||||
1);
|
||||
if (buff->mandelbrot_buffer[y * buff->width + x] == buff->iterations) {
|
||||
if (data_array[y * buff->width + x] == buff->iterations) {
|
||||
cairo_set_source_rgb (cr, 0, 0, 0);
|
||||
} else {
|
||||
color_scale = (double)buff->mandelbrot_buffer[y * buff->width + x] / buff->iterations;
|
||||
color_scale = (double)data_array[y * buff->width + x] / buff->iterations;
|
||||
cairo_set_source_rgb (cr, (sqrt(1-color_scale)), (color_scale*color_scale), color_scale);
|
||||
}
|
||||
cairo_fill(cr);
|
||||
@@ -57,65 +97,130 @@ static gboolean drawing_callback(GtkWidget *widget, cairo_t *cr, gpointer data)
|
||||
}
|
||||
|
||||
|
||||
|
||||
/* Drawing code end */
|
||||
cairo_restore(cr);
|
||||
|
||||
return FALSE;
|
||||
}
|
||||
|
||||
static void calculate_mandelbrot(struct canvas_buffer *buff)
|
||||
|
||||
static int calculate_mandelbrot(struct canvas_buffer *buff)
|
||||
{
|
||||
double val_per_x_pixel;
|
||||
double val_per_y_pixel;
|
||||
unsigned int i;
|
||||
unsigned int pixel_x;
|
||||
unsigned int pixel_y;
|
||||
double point_real, point_imag;
|
||||
double z_real, z_imag;
|
||||
double z2_real, z2_imag;
|
||||
|
||||
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 context;
|
||||
cl_program program;
|
||||
cl_kernel kernel;
|
||||
cl_command_queue queue;
|
||||
cl_mem dat_buff;
|
||||
char *source_code = NULL;
|
||||
char temp_buff[100];
|
||||
size_t data_size;
|
||||
size_t work_size = 64;
|
||||
|
||||
|
||||
if (!buff)
|
||||
return;
|
||||
data_size = buff->width*buff->height;
|
||||
data_array = (unsigned int *)malloc(sizeof(data_array[0])*data_size);
|
||||
if (!data_array)
|
||||
return -ENOMEM;
|
||||
|
||||
/* Setup the calculation data */
|
||||
val_per_x_pixel = (buff->x_span) / (buff->width - 1);
|
||||
val_per_y_pixel = (buff->y_span) / (buff->height - 1);
|
||||
|
||||
for (pixel_y = 0; pixel_y < buff->height; pixel_y++) {
|
||||
for (pixel_x = 0; pixel_x < buff->width; pixel_x++) {
|
||||
point_real = (((double)pixel_x - (double)(buff->width - 1) / 2)) * val_per_x_pixel + buff->center_x;
|
||||
point_imag = (((double)pixel_y - (double)(buff->height - 1) / 2)) * val_per_y_pixel + buff->center_y;
|
||||
clGetPlatformIDs(0, NULL, &platform_id_count);
|
||||
if (platform_id_count == 0)
|
||||
return -ENOMEM;
|
||||
|
||||
z_real = 0.0;
|
||||
z_imag = 0.0;
|
||||
platform_ids = (cl_platform_id *)malloc(platform_id_count *
|
||||
sizeof(cl_platform_id));
|
||||
if (platform_ids == NULL)
|
||||
return -ENOMEM;
|
||||
|
||||
/* Check for convergence */
|
||||
for (i = 0; i < buff->iterations; i++) {
|
||||
clGetPlatformIDs(platform_id_count, platform_ids, NULL);
|
||||
|
||||
z2_real = z_real * z_real - z_imag * z_imag;
|
||||
z2_imag = 2.0 * z_real * z_imag;
|
||||
printf("Platforms available: %u\n", (unsigned int)platform_id_count);
|
||||
|
||||
z_real = z2_real + point_real;
|
||||
z_imag = z2_imag + point_imag;
|
||||
|
||||
if ((z_real * z_real + z_imag * z_imag) >= 4.0)
|
||||
break;
|
||||
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);
|
||||
}
|
||||
|
||||
buff->mandelbrot_buffer[pixel_y * buff->width + pixel_x] = i;
|
||||
}
|
||||
clGetDeviceIDs(platform_ids[0], CL_DEVICE_TYPE_ALL, 0, NULL,
|
||||
&device_id_count);
|
||||
printf("Device in Platform 0 count: %u\n",
|
||||
(unsigned int)device_id_count);
|
||||
|
||||
if (device_id_count == 0)
|
||||
return -1;
|
||||
|
||||
device_ids = (cl_device_id *) malloc(sizeof(cl_device_id)
|
||||
*device_id_count);
|
||||
|
||||
clGetDeviceIDs(platform_ids[0], CL_DEVICE_TYPE_ALL, device_id_count,
|
||||
device_ids, NULL);
|
||||
|
||||
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(NULL, 1, &device_ids[0], NULL, NULL, &error);
|
||||
check_error(error);
|
||||
|
||||
queue = clCreateCommandQueue(context, device_ids[0], 0, &error);
|
||||
|
||||
dat_buff = clCreateBuffer(context, CL_MEM_WRITE_ONLY |
|
||||
CL_MEM_COPY_HOST_PTR,
|
||||
sizeof(data_array[0])*data_size, data_array, &error);
|
||||
check_error(error);
|
||||
|
||||
load_kernel_from_file("cl_kernels/mandelbrot.cl", &source_code);
|
||||
program = create_program(source_code, context);
|
||||
|
||||
/* Compile the source code */
|
||||
check_error(clBuildProgram(program, device_id_count,
|
||||
device_ids, NULL, NULL, NULL));
|
||||
kernel = clCreateKernel(program, "mandelbrot", &error);
|
||||
|
||||
|
||||
/* Setting the arguments */
|
||||
clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&dat_buff);
|
||||
clSetKernelArg(kernel, 1, sizeof(unsigned int), (void *)&buff->iterations);
|
||||
clSetKernelArg(kernel, 2, sizeof(double), (void *)&buff->center_x);
|
||||
clSetKernelArg(kernel, 3, sizeof(double), (void *)&buff->center_y);
|
||||
clSetKernelArg(kernel, 4, sizeof(double), (void *)&buff->x_span);
|
||||
clSetKernelArg(kernel, 5, sizeof(double), (void *)&buff->y_span);
|
||||
clSetKernelArg(kernel, 6, sizeof(unsigned int), (void *)&buff->width);
|
||||
clSetKernelArg(kernel, 7, sizeof(unsigned int), (void *)&buff->height);
|
||||
|
||||
ret = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &data_size, &work_size, 0, NULL, NULL);
|
||||
check_error(ret);
|
||||
|
||||
clEnqueueReadBuffer(queue, dat_buff, CL_TRUE, 0, data_size*sizeof(data_array[0]), data_array, 0, NULL, NULL);
|
||||
|
||||
ret = clFlush(queue);
|
||||
ret = clFinish(queue);
|
||||
ret = clReleaseKernel(kernel);
|
||||
ret = clReleaseProgram(program);
|
||||
ret = clReleaseMemObject(dat_buff);
|
||||
ret = clReleaseCommandQueue(queue);
|
||||
ret = clReleaseContext(context);
|
||||
free(platform_ids);
|
||||
free(device_ids);
|
||||
|
||||
return 0;
|
||||
|
||||
}
|
||||
|
||||
|
||||
int f(void * a)
|
||||
{
|
||||
return 1;
|
||||
}
|
||||
|
||||
|
||||
int main(int argc, char **argv)
|
||||
{
|
||||
@@ -137,18 +242,15 @@ int main(int argc, char **argv)
|
||||
|
||||
g_object_unref(builder);
|
||||
|
||||
thrd_t t;
|
||||
thrd_create(&t, f, NULL);
|
||||
|
||||
|
||||
mandelbrot_buff.x_span = 2.8;
|
||||
mandelbrot_buff.y_span = 2.25;
|
||||
mandelbrot_buff.center_x = -0.5;
|
||||
mandelbrot_buff.center_y = 0.0;
|
||||
mandelbrot_buff.width = 1900;
|
||||
mandelbrot_buff.height = 1000;
|
||||
mandelbrot_buff.iterations = 300;
|
||||
mandelbrot_buff.mandelbrot_buffer = (unsigned int *)malloc(sizeof(unsigned int) * mandelbrot_buff.height * mandelbrot_buff.width);
|
||||
mandelbrot_buff.x_span = 1;//2.8;
|
||||
mandelbrot_buff.y_span = 0.5;//2.25;
|
||||
mandelbrot_buff.center_x = -0.6;
|
||||
mandelbrot_buff.center_y = 0.6;
|
||||
mandelbrot_buff.width = 4000;
|
||||
mandelbrot_buff.height = 2000;
|
||||
mandelbrot_buff.iterations = 70;
|
||||
mandelbrot_buff.mandelbrot_buffer = NULL;
|
||||
calculate_mandelbrot(&mandelbrot_buff);
|
||||
|
||||
|
||||
|
||||
Reference in New Issue
Block a user