diff --git a/CMakeLists.txt b/CMakeLists.txt index 82c23dd..24ddff4 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -27,7 +27,7 @@ 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 ${OPENCL_LIBRARIES}) +target_link_libraries(${PROJECT_NAME} ${GLIB_LDFLAGS} ${GTK3_LDFLAGS} ${CAIRO_LDFLAGS} m pthread ${OPENCL_LIBRARIES} epoxy) install (TARGETS ${PROJECT_NAME} DESTINATION bin) message(STATUS "OpenCL found: ${OPENCL_FOUND}") diff --git a/glade/main.glade b/glade/main.glade index 1935ad6..6b42667 100644 --- a/glade/main.glade +++ b/glade/main.glade @@ -11,10 +11,7 @@ - - True - False - + diff --git a/mandelbrot.cl b/mandelbrot.cl index b7fd9c5..d33b52e 100644 --- a/mandelbrot.cl +++ b/mandelbrot.cl @@ -1,4 +1,4 @@ -__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) +__kernel void mandelbrot(__global unsigned char *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); @@ -6,6 +6,8 @@ __kernel void mandelbrot(__global unsigned int *out_buff, unsigned int iter, dou double z_real, z_imag; double z2_real, z2_imag; double pt_real, pt_imag; + double color_scale; + unsigned int r, g, b; const unsigned int pix_x = idx % width; const unsigned int pix_y = (unsigned int)idx / width; @@ -30,5 +32,15 @@ __kernel void mandelbrot(__global unsigned int *out_buff, unsigned int iter, dou break; } - out_buff[idx] = i; + color_scale = (double)(i) / (double)iter; + if (i != iter) { + r = (unsigned char)(255.0 * sqrt((1-color_scale))); + g = (unsigned char)(255.0 * color_scale*color_scale); + b = (unsigned char)(255.0 * color_scale); + } else { + r = g = b = 0; + } + out_buff[idx*3] = r; + out_buff[idx*3+1] = g; + out_buff[idx*3+2] = b; } diff --git a/src/main.c b/src/main.c index 8b9678f..80b8373 100644 --- a/src/main.c +++ b/src/main.c @@ -4,9 +4,35 @@ #include #include #include +#include +#include + + +static const char * const vertex_shader = "#version 330 core\n \ + layout (location = 0) in vec3 aPos;\n \ + layout (location = 1) in vec2 aTexCoord;\n\ + out vec2 texCoord;\n \ + void main()\n \ + {\n \ + gl_Position = vec4(aPos.x, aPos.y, aPos.z, 1.0);\n \ + texCoord = aTexCoord;\n\ + }\n"; + +static const char * const orange_framgment_shader = " \ + #version 330 core\n \ + out vec4 FragColor;\n \ + in vec4 vertexColor;\n \ + in vec2 texCoord;\n \ + uniform sampler2D ourTexture;\n \ + \n \ + void main()\n \ + {\n \ + FragColor = texture(ourTexture, texCoord);\n \ + }"; + struct canvas_buffer { - unsigned int *mandelbrot_buffer; + unsigned char *mandelbrot_buffer; unsigned int iterations; unsigned int height; unsigned int width; @@ -58,51 +84,6 @@ cl_program create_program(char *src, cl_context context) 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; - int width; - int height; - GtkStyleContext *style_context; - int x, y; - double color_scale; - - style_context = gtk_widget_get_style_context(widget); - width = gtk_widget_get_allocated_width(widget); - height = gtk_widget_get_allocated_height(widget); - - cairo_save(cr); - /* Drawing code start */ - 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, - ((double)x - (double)(buff->width - 1) / 2), - ((double)y - (double)(buff->height - 1) / 2), - 1, - 1); - if (data_array[y * buff->width + x] == buff->iterations) { - cairo_set_source_rgb (cr, 0, 0, 0); - } else { - 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); - - } - } - - - /* Drawing code end */ - cairo_restore(cr); - - return FALSE; -} - static int calculate_mandelbrot(struct canvas_buffer *buff) { @@ -122,12 +103,15 @@ static int calculate_mandelbrot(struct canvas_buffer *buff) char *source_code = NULL; char temp_buff[100]; size_t data_size; - size_t work_size = 64; + size_t work_size = 10; data_size = buff->width*buff->height; - data_array = (unsigned int *)malloc(sizeof(data_array[0])*data_size); - if (!data_array) + + if (!buff->mandelbrot_buffer) + buff->mandelbrot_buffer = (unsigned char *)malloc(data_size*3); + + if (!buff->mandelbrot_buffer) return -ENOMEM; @@ -178,7 +162,7 @@ static int calculate_mandelbrot(struct canvas_buffer *buff) dat_buff = clCreateBuffer(context, CL_MEM_WRITE_ONLY | CL_MEM_COPY_HOST_PTR, - sizeof(data_array[0])*data_size, data_array, &error); + 3*data_size, buff->mandelbrot_buffer, &error); check_error(error); load_kernel_from_file("cl_kernels/mandelbrot.cl", &source_code); @@ -201,9 +185,9 @@ static int calculate_mandelbrot(struct canvas_buffer *buff) 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); + check_error(ret); - clEnqueueReadBuffer(queue, dat_buff, CL_TRUE, 0, data_size*sizeof(data_array[0]), data_array, 0, NULL, NULL); + clEnqueueReadBuffer(queue, dat_buff, CL_TRUE, 0, data_size*3, buff->mandelbrot_buffer, 0, NULL, NULL); ret = clFlush(queue); ret = clFinish(queue); @@ -219,39 +203,173 @@ static int calculate_mandelbrot(struct canvas_buffer *buff) } +float vertices[] = { + // Vertex Texture + -1.0f, -1.0f, 0.0f, 0.0f, 0.0f, // bottom left + -1.0f, 1.0f, 0.0f, 0.0f, 1.0f, // top left + 1.0f, -1.0f, 0.0f, 1.0f, 0.0f, // bottom right + 1.0f, 1.0f, 0.0f, 1.0f, 1.0f, // top right +}; + +unsigned int indices[] = { + 0, 1, 2, + 2, 1, 3 +}; + +static unsigned int vertex_buffer_object; +static unsigned int element_buffer_object; +static unsigned int vertex_array_object; +static unsigned int shader_program_id; +static unsigned int texture_id; + +gboolean render(GtkGLArea *gl_area, GdkGLContext *context, gpointer user_data) +{ + + gtk_gl_area_make_current(gl_area); + + glClear(GL_COLOR_BUFFER_BIT); + + glUseProgram(shader_program_id); + //glPolygonMode(GL_FRONT_AND_BACK, GL_LINE); + glBindTexture(GL_TEXTURE_2D, texture_id); + glBindVertexArray(vertex_array_object); + //glDrawArrays(GL_TRIANGLES, 0, 3); + glDrawElements(GL_TRIANGLES, 6, GL_UNSIGNED_INT, 0); + + glFlush(); + + return TRUE; +} + +void realize(GtkGLArea *gl_area, gpointer user_data) +{ + + unsigned int vertex_shader_id; + unsigned int fragment_shader_id; + struct canvas_buffer *mandelbrot_buff = (struct canvas_buffer *)user_data; + + int success; + char info_log[512]; + + gtk_gl_area_make_current(gl_area); + glClearColor(0, 0, 0, 1); + + glGenVertexArrays(1, &vertex_array_object); + glBindVertexArray(vertex_array_object); + + glGenBuffers(1, &vertex_buffer_object); + glBindBuffer(GL_ARRAY_BUFFER, vertex_buffer_object); + glBufferData(GL_ARRAY_BUFFER, sizeof(vertices), vertices, GL_STATIC_DRAW); + + glGenBuffers(1, &element_buffer_object); + glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, element_buffer_object); + glBufferData(GL_ELEMENT_ARRAY_BUFFER, sizeof(indices), indices, GL_STATIC_DRAW); + + vertex_shader_id = glCreateShader(GL_VERTEX_SHADER); + glShaderSource(vertex_shader_id, 1, &vertex_shader, NULL); + glCompileShader(vertex_shader_id); + glGetShaderiv(vertex_shader_id, GL_COMPILE_STATUS, &success); + + if (!success) { + glGetShaderInfoLog(vertex_shader_id, 512, NULL, info_log); + printf("%s\n", info_log); + } else { + printf("Vertex shader compilation successful\n"); + } + + fragment_shader_id = glCreateShader(GL_FRAGMENT_SHADER); + glShaderSource(fragment_shader_id, 1, &orange_framgment_shader, NULL); + glCompileShader(fragment_shader_id); + glGetShaderiv(fragment_shader_id, GL_COMPILE_STATUS, &success); + + if (!success) { + glGetShaderInfoLog(fragment_shader_id, 512, NULL, info_log); + printf("%s\n", info_log); + } else { + printf("Fragment shader compilation successful\n"); + } + + shader_program_id = glCreateProgram(); + glAttachShader(shader_program_id, vertex_shader_id); + glAttachShader(shader_program_id, fragment_shader_id); + glLinkProgram(shader_program_id); + + glGetShaderiv(shader_program_id, GL_LINK_STATUS, &success); + if (!success) { + glGetShaderInfoLog(shader_program_id, 512, NULL, info_log); + printf("%s\n", info_log); + } else { + printf("Shader Core linked successfully\n"); + } + + glUseProgram(shader_program_id); + glDeleteShader(vertex_shader_id); + glDeleteShader(fragment_shader_id); + glVertexAttribPointer(0, 3, GL_FLOAT, GL_FALSE, 5 * sizeof(float), (void*)0); + glEnableVertexAttribArray(0); + + glVertexAttribPointer(1, 2, GL_FLOAT, GL_FALSE, 5 * sizeof(float), (void*)(3 * sizeof(float))); + glEnableVertexAttribArray(1); + + glGenTextures(1, &texture_id); + glBindTexture(GL_TEXTURE_2D, texture_id); + glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_REPEAT); + glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_REPEAT); + glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR); + glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR); + glTexImage2D(GL_TEXTURE_2D, 0, GL_RGB, mandelbrot_buff->width, mandelbrot_buff->height, + 0, GL_RGB, GL_UNSIGNED_BYTE, mandelbrot_buff->mandelbrot_buffer); + glGenerateMipmap(GL_TEXTURE_2D); + + + + //glClear(GL_COLOR_BUFFER_BIT); +} int main(int argc, char **argv) { - GtkDrawingArea *drawing_area; + GtkGLArea *gl_area; GtkWindow *window; GtkBuilder *builder; + GtkWidget *temp; static struct canvas_buffer mandelbrot_buff; gtk_init(&argc, &argv); builder = gtk_builder_new_from_resource("/main.glade"); - drawing_area = GTK_DRAWING_AREA(gtk_builder_get_object(builder, "drawing-area")); window = GTK_WINDOW(gtk_builder_get_object(builder, "main-window")); - g_signal_connect(window, "delete-event", G_CALLBACK(on_main_window_close), NULL); - g_signal_connect(G_OBJECT(drawing_area), - "draw", G_CALLBACK(drawing_callback), (gpointer)&mandelbrot_buff); + + + + mandelbrot_buff.x_span = 0.075;//2.8; + mandelbrot_buff.y_span = 0.075;//2.25; + mandelbrot_buff.center_x = -0.6; + mandelbrot_buff.center_y = 0.45; + mandelbrot_buff.width = 3000; + mandelbrot_buff.height = 3000; + mandelbrot_buff.iterations = 200; + mandelbrot_buff.mandelbrot_buffer = NULL; + calculate_mandelbrot(&mandelbrot_buff); + + + + gl_area = GTK_GL_AREA(gtk_gl_area_new()); + gtk_gl_area_set_auto_render(gl_area, TRUE); + gtk_widget_set_vexpand(GTK_WIDGET(gl_area), TRUE); + gtk_widget_set_hexpand(GTK_WIDGET(gl_area), TRUE); + g_signal_connect(gl_area, "realize", G_CALLBACK(realize), &mandelbrot_buff); + g_signal_connect(gl_area, "render", G_CALLBACK(render), NULL); + + gtk_container_add(GTK_CONTAINER(window), GTK_WIDGET(gl_area)); + gtk_widget_show(GTK_WIDGET(gl_area)); g_object_unref(builder); - - 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); + gtk_gl_area_make_current(gl_area); gtk_main();