aboutsummaryrefslogtreecommitdiff
diff options
context:
space:
mode:
-rw-r--r--Release/cl/mandelbrot64.cl34
-rw-r--r--Release/glade/settings.glade263
-rw-r--r--src/mandelbrot-zoom.c10
-rw-r--r--src/mandelbrot-zoom.h4
-rw-r--r--src/render.c22
-rw-r--r--src/render.h1
-rw-r--r--src/render_cpu.c63
-rw-r--r--src/render_cpu.h4
-rw-r--r--src/render_opencl.c89
-rw-r--r--src/render_opencl.h9
10 files changed, 335 insertions, 164 deletions
diff --git a/Release/cl/mandelbrot64.cl b/Release/cl/mandelbrot64.cl
index be38b4a..c89f17c 100644
--- a/Release/cl/mandelbrot64.cl
+++ b/Release/cl/mandelbrot64.cl
@@ -1,11 +1,15 @@
+#define native_log2 log2
+#define native_cos cos
+
#ifdef KHR_DP_EXTENSION
#pragma OPENCL EXTENSION cl_khr_fp64 : enable
#else
#pragma OPENCL EXTENSION cl_amd_fp64 : enable
#endif
-__kernel void calculate(__global uchar4 * mandelbrotImage, const double posx, const double posy, const double stepSizeX, const double stepSizeY, const uint maxIterations, const uint width)
+__kernel void calculate(__global uchar4 * mandelbrotImage, const double posx, const double posy, const double stepSizeX, const double stepSizeY, const uint maxIterations, const int width)
{
+//printf("%f, %f, %f, %f, %u, %d", posx, posy, stepSizeX, stepSizeY, maxIterations, width);
int tid = get_global_id(0);
int i = tid % (width / 4);
@@ -162,13 +166,6 @@ __kernel void calculate(__global uchar4 * mandelbrotImage, const double posx, co
color[0].s1 = 0;
color[0].s2 = 0;
}
- if (bench)
- {
- color[0].s0 = ccount.s0 & 0xff;
- color[0].s1 = (ccount.s0 & 0xff00) >> 8;
- color[0].s2 = (ccount.s0 & 0xff0000) >> 16;
- color[0].s3 = (ccount.s0 & 0xff000000) >> 24;
- }
mandelbrotImage[4 * tid] = color[0];
c = fc.s1 * 2.0 * 3.1416 / 256.0;
color[1].s0 = ((1.0 + native_cos(c)) * 0.5) * 255;
@@ -181,13 +178,6 @@ __kernel void calculate(__global uchar4 * mandelbrotImage, const double posx, co
color[1].s1 = 0;
color[1].s2 = 0;
}
- if (bench)
- {
- color[1].s0 = ccount.s1 & 0xff;
- color[1].s1 = (ccount.s1 & 0xff00) >> 8;
- color[1].s2 = (ccount.s1 & 0xff0000) >> 16;
- color[1].s3 = (ccount.s1 & 0xff000000) >> 24;
- }
mandelbrotImage[4 * tid + 1] = color[1];
c = fc.s2 * 2.0 * 3.1416 / 256.0;
color[2].s0 = ((1.0 + native_cos(c)) * 0.5) * 255;
@@ -200,13 +190,6 @@ __kernel void calculate(__global uchar4 * mandelbrotImage, const double posx, co
color[2].s1 = 0;
color[2].s2 = 0;
}
- if (bench)
- {
- color[2].s0 = ccount.s2 & 0xff;
- color[2].s1 = (ccount.s2 & 0xff00) >> 8;
- color[2].s2 = (ccount.s2 & 0xff0000) >> 16;
- color[2].s3 = (ccount.s2 & 0xff000000) >> 24;
- }
mandelbrotImage[4 * tid + 2] = color[2];
c = fc.s3 * 2.0 * 3.1416 / 256.0;
color[3].s0 = ((1.0 + native_cos(c)) * 0.5) * 255;
@@ -219,12 +202,5 @@ __kernel void calculate(__global uchar4 * mandelbrotImage, const double posx, co
color[3].s1 = 0;
color[3].s2 = 0;
}
- if (bench)
- {
- color[3].s0 = ccount.s3 & 0xff;
- color[3].s1 = (ccount.s3 & 0xff00) >> 8;
- color[3].s2 = (ccount.s3 & 0xff0000) >> 16;
- color[3].s3 = (ccount.s3 & 0xff000000) >> 24;
- }
mandelbrotImage[4 * tid + 3] = color[3];
} \ No newline at end of file
diff --git a/Release/glade/settings.glade b/Release/glade/settings.glade
index ff11a2c..8029d00 100644
--- a/Release/glade/settings.glade
+++ b/Release/glade/settings.glade
@@ -95,7 +95,6 @@
<property name="can_focus">False</property>
<property name="n_rows">6</property>
<property name="n_columns">2</property>
- <property name="homogeneous">True</property>
<child>
<object class="GtkLabel" id="iterationsLbl">
<property name="visible">True</property>
@@ -220,107 +219,6 @@
</packing>
</child>
<child>
- <object class="GtkLabel" id="zoomLbl">
- <property name="visible">True</property>
- <property name="can_focus">False</property>
- <property name="label" translatable="yes">Zoom to (temp)</property>
- </object>
- <packing>
- <property name="top_attach">3</property>
- <property name="bottom_attach">4</property>
- </packing>
- </child>
- <child>
- <object class="GtkHBox" id="hbox5">
- <property name="visible">True</property>
- <property name="can_focus">False</property>
- <child>
- <object class="GtkHBox" id="hbox6">
- <property name="visible">True</property>
- <property name="can_focus">False</property>
- <child>
- <object class="GtkLabel" id="zoomToXLbl">
- <property name="visible">True</property>
- <property name="can_focus">False</property>
- <property name="label" translatable="yes">X</property>
- </object>
- <packing>
- <property name="expand">True</property>
- <property name="fill">True</property>
- <property name="position">0</property>
- </packing>
- </child>
- <child>
- <object class="GtkEntry" id="zoomToXEntry">
- <property name="visible">True</property>
- <property name="can_focus">True</property>
- <property name="invisible_char">●</property>
- <property name="primary_icon_activatable">False</property>
- <property name="secondary_icon_activatable">False</property>
- <property name="primary_icon_sensitive">True</property>
- <property name="secondary_icon_sensitive">True</property>
- </object>
- <packing>
- <property name="expand">True</property>
- <property name="fill">True</property>
- <property name="position">1</property>
- </packing>
- </child>
- </object>
- <packing>
- <property name="expand">True</property>
- <property name="fill">True</property>
- <property name="position">0</property>
- </packing>
- </child>
- <child>
- <object class="GtkHBox" id="hbox7">
- <property name="visible">True</property>
- <property name="can_focus">False</property>
- <child>
- <object class="GtkLabel" id="zoomToYLbl">
- <property name="visible">True</property>
- <property name="can_focus">False</property>
- <property name="label" translatable="yes">Y</property>
- </object>
- <packing>
- <property name="expand">True</property>
- <property name="fill">True</property>
- <property name="position">0</property>
- </packing>
- </child>
- <child>
- <object class="GtkEntry" id="zoomToYEntry">
- <property name="visible">True</property>
- <property name="can_focus">True</property>
- <property name="invisible_char">●</property>
- <property name="primary_icon_activatable">False</property>
- <property name="secondary_icon_activatable">False</property>
- <property name="primary_icon_sensitive">True</property>
- <property name="secondary_icon_sensitive">True</property>
- </object>
- <packing>
- <property name="expand">True</property>
- <property name="fill">True</property>
- <property name="position">1</property>
- </packing>
- </child>
- </object>
- <packing>
- <property name="expand">True</property>
- <property name="fill">True</property>
- <property name="position">1</property>
- </packing>
- </child>
- </object>
- <packing>
- <property name="left_attach">1</property>
- <property name="right_attach">2</property>
- <property name="top_attach">3</property>
- <property name="bottom_attach">4</property>
- </packing>
- </child>
- <child>
<object class="GtkLabel" id="speedLbl">
<property name="visible">True</property>
<property name="can_focus">False</property>
@@ -420,6 +318,132 @@
<property name="right_attach">2</property>
</packing>
</child>
+ <child>
+ <object class="GtkLabel" id="zoomLbl">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ <property name="label" translatable="yes">Zoom</property>
+ </object>
+ <packing>
+ <property name="top_attach">3</property>
+ <property name="bottom_attach">4</property>
+ </packing>
+ </child>
+ <child>
+ <object class="GtkFrame" id="zoomFrame">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ <property name="label_xalign">0</property>
+ <child>
+ <object class="GtkAlignment" id="alignment7">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ <property name="left_padding">12</property>
+ <child>
+ <object class="GtkVBox" id="vbox3">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ <child>
+ <object class="GtkCheckButton" id="zoomSpecificCb">
+ <property name="label" translatable="yes">Zoom to specific coordinates</property>
+ <property name="visible">True</property>
+ <property name="can_focus">True</property>
+ <property name="receives_default">False</property>
+ <property name="draw_indicator">True</property>
+ </object>
+ <packing>
+ <property name="expand">True</property>
+ <property name="fill">True</property>
+ <property name="position">0</property>
+ </packing>
+ </child>
+ <child>
+ <object class="GtkHBox" id="hbox5">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ <child>
+ <object class="GtkLabel" id="zoomToXLbl">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ <property name="label" translatable="yes">X</property>
+ </object>
+ <packing>
+ <property name="expand">True</property>
+ <property name="fill">True</property>
+ <property name="position">0</property>
+ </packing>
+ </child>
+ <child>
+ <object class="GtkEntry" id="zoomToXEntry">
+ <property name="visible">True</property>
+ <property name="can_focus">True</property>
+ <property name="invisible_char">●</property>
+ <property name="primary_icon_activatable">False</property>
+ <property name="secondary_icon_activatable">False</property>
+ <property name="primary_icon_sensitive">True</property>
+ <property name="secondary_icon_sensitive">True</property>
+ </object>
+ <packing>
+ <property name="expand">True</property>
+ <property name="fill">True</property>
+ <property name="position">1</property>
+ </packing>
+ </child>
+ <child>
+ <object class="GtkLabel" id="zoomToYLbl">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ <property name="label" translatable="yes">Y</property>
+ </object>
+ <packing>
+ <property name="expand">True</property>
+ <property name="fill">True</property>
+ <property name="position">2</property>
+ </packing>
+ </child>
+ <child>
+ <object class="GtkEntry" id="zoomToYEntry">
+ <property name="visible">True</property>
+ <property name="can_focus">True</property>
+ <property name="invisible_char">●</property>
+ <property name="primary_icon_activatable">False</property>
+ <property name="secondary_icon_activatable">False</property>
+ <property name="primary_icon_sensitive">True</property>
+ <property name="secondary_icon_sensitive">True</property>
+ </object>
+ <packing>
+ <property name="expand">True</property>
+ <property name="fill">True</property>
+ <property name="position">3</property>
+ </packing>
+ </child>
+ </object>
+ <packing>
+ <property name="expand">True</property>
+ <property name="fill">True</property>
+ <property name="position">1</property>
+ </packing>
+ </child>
+ </object>
+ </child>
+ </object>
+ </child>
+ <child type="label">
+ <object class="GtkLabel" id="zoomLbl1">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ <property name="label" translatable="yes">Zoom precision</property>
+ <property name="use_markup">True</property>
+ </object>
+ </child>
+ </object>
+ <packing>
+ <property name="left_attach">1</property>
+ <property name="right_attach">2</property>
+ <property name="top_attach">3</property>
+ <property name="bottom_attach">4</property>
+ </packing>
+ </child>
</object>
</child>
</object>
@@ -521,7 +545,7 @@
<object class="GtkLabel" id="fpuLbl">
<property name="visible">True</property>
<property name="can_focus">False</property>
- <property name="label" translatable="yes">FPU float depth</property>
+ <property name="label" translatable="yes">FPU float precision</property>
</object>
<packing>
<property name="expand">True</property>
@@ -863,20 +887,20 @@
<object class="GtkTable" id="table1">
<property name="visible">True</property>
<property name="can_focus">False</property>
- <property name="n_rows">2</property>
+ <property name="n_rows">3</property>
<property name="n_columns">2</property>
<child>
<object class="GtkLabel" id="inLbl">
<property name="visible">True</property>
<property name="can_focus">False</property>
- <property name="label" translatable="yes">Zoom in</property>
+ <property name="label" translatable="yes">Start zoom</property>
</object>
</child>
<child>
- <object class="GtkLabel" id="outLbl">
+ <object class="GtkLabel" id="incIterLbl">
<property name="visible">True</property>
<property name="can_focus">False</property>
- <property name="label" translatable="yes">Zoom out</property>
+ <property name="label" translatable="yes">Increase iterations</property>
</object>
<packing>
<property name="top_attach">1</property>
@@ -884,6 +908,17 @@
</packing>
</child>
<child>
+ <object class="GtkLabel" id="decIterLbl">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ <property name="label" translatable="yes">Decrease iterations</property>
+ </object>
+ <packing>
+ <property name="top_attach">2</property>
+ <property name="bottom_attach">3</property>
+ </packing>
+ </child>
+ <child>
<object class="GtkComboBoxText" id="inCombo">
<property name="visible">True</property>
<property name="can_focus">False</property>
@@ -894,7 +929,7 @@
</packing>
</child>
<child>
- <object class="GtkComboBoxText" id="outCombo">
+ <object class="GtkComboBoxText" id="incIterCombo">
<property name="visible">True</property>
<property name="can_focus">False</property>
</object>
@@ -905,6 +940,18 @@
<property name="bottom_attach">2</property>
</packing>
</child>
+ <child>
+ <object class="GtkComboBoxText" id="decIterCombo">
+ <property name="visible">True</property>
+ <property name="can_focus">False</property>
+ </object>
+ <packing>
+ <property name="left_attach">1</property>
+ <property name="right_attach">2</property>
+ <property name="top_attach">2</property>
+ <property name="bottom_attach">3</property>
+ </packing>
+ </child>
</object>
</child>
</object>
diff --git a/src/mandelbrot-zoom.c b/src/mandelbrot-zoom.c
index 72b54ba..f009e10 100644
--- a/src/mandelbrot-zoom.c
+++ b/src/mandelbrot-zoom.c
@@ -22,6 +22,7 @@ int main(int argc, char **argv)
ui_settings.threadsSp = GTK_SPIN_BUTTON(gtk_builder_get_object(builder, "threadsSp"));
ui_settings.colorFromBtn = GTK_BUTTON(gtk_builder_get_object(builder, "colorFromBtn"));
ui_settings.colorToBtn = GTK_BUTTON(gtk_builder_get_object(builder, "colorToBtn"));
+ ui_settings.zoomSpecificCb = GTK_CHECK_BUTTON(gtk_builder_get_object(builder, "zoomSpecificCb"));
ui_settings.zoomToXEntry = GTK_ENTRY(gtk_builder_get_object(builder, "zoomToXEntry"));
ui_settings.zoomToYEntry = GTK_ENTRY(gtk_builder_get_object(builder, "zoomToYEntry"));
ui_settings.speedEntry = GTK_ENTRY(gtk_builder_get_object(builder, "speedEntry"));
@@ -40,6 +41,9 @@ int main(int argc, char **argv)
ui_settings.startBtn = GTK_BUTTON(gtk_builder_get_object(builder, "startBtn"));
ui_settings.exitBtn = GTK_BUTTON(gtk_builder_get_object(builder, "exitBtn"));
ui_settings.colorDialog = GTK_COLOR_SELECTION_DIALOG(gtk_builder_get_object(builder, "colorDialog"));
+ ui_settings.inCombo = GTK_COMBO_BOX_TEXT(gtk_builder_get_object(builder, "inCombo"));
+ ui_settings.incIterCombo = GTK_COMBO_BOX_TEXT(gtk_builder_get_object(builder, "incIterCombo"));
+ ui_settings.decIterCombo = GTK_COMBO_BOX_TEXT(gtk_builder_get_object(builder, "decIterCombo"));
gtk_spin_button_set_increments(ui_settings.iterationsSp, 1, 2);
gtk_spin_button_set_increments(ui_settings.threadsSp, 1, 2);
@@ -72,6 +76,8 @@ int main(int argc, char **argv)
gtk_combo_box_text_append(ui_settings.fpuCombo, NULL, "64 bit");
gtk_combo_box_text_append(ui_settings.fpuCombo, NULL, "128 bit");
+
+
g_signal_connect(ui_settings.colorFromBtn, "clicked", G_CALLBACK(on_colorFromBtn_clicked), NULL);
g_signal_connect(ui_settings.colorToBtn, "clicked", G_CALLBACK(on_colorToBtn_clicked), NULL);
g_signal_connect(ui_settings.startBtn, "clicked", G_CALLBACK(on_startBtn_clicked), NULL);
@@ -100,14 +106,18 @@ void on_startBtn_clicked()
config.config_cpu.iterations = gtk_spin_button_get_value(ui_settings.iterationsSp);
config.config_opencl.iterations = gtk_spin_button_get_value(ui_settings.iterationsSp);
+ config.zoomSpecific = gtk_toggle_button_get_active(GTK_TOGGLE_BUTTON(ui_settings.zoomSpecificCb));
+
config.config_cpu.threads = gtk_spin_button_get_value(ui_settings.threadsSp);
config.config_opencl.fma = gtk_toggle_button_get_active(GTK_TOGGLE_BUTTON(ui_settings.fmaCb));
config.video = gtk_toggle_button_get_active(GTK_TOGGLE_BUTTON(ui_settings.exportCb));
+
config.path = CHAR_PTR(gtk_entry_get_text(ui_settings.exportTf));
+
config.width = gtk_spin_button_get_value(ui_settings.widthSp);
config.config_cpu.width = gtk_spin_button_get_value(ui_settings.widthSp);
config.config_opencl.width = gtk_spin_button_get_value(ui_settings.widthSp);
diff --git a/src/mandelbrot-zoom.h b/src/mandelbrot-zoom.h
index 8309885..5aa4dc4 100644
--- a/src/mandelbrot-zoom.h
+++ b/src/mandelbrot-zoom.h
@@ -21,6 +21,7 @@ typedef struct Ui_settings {
GtkSpinButton *threadsSp;
GtkButton *colorFromBtn;
GtkButton *colorToBtn;
+ GtkCheckButton *zoomSpecificCb;
GtkEntry *zoomToXEntry;
GtkEntry *zoomToYEntry;
GtkEntry *speedEntry;
@@ -39,6 +40,9 @@ typedef struct Ui_settings {
GtkButton *startBtn;
GtkButton *exitBtn;
GtkColorSelectionDialog *colorDialog;
+ GtkComboBoxText *inCombo;
+ GtkComboBoxText *incIterCombo;
+ GtkComboBoxText *decIterCombo;
} Ui_settings;
Ui_settings ui_settings;
diff --git a/src/render.c b/src/render.c
index 2ebb7c3..6c01029 100644
--- a/src/render.c
+++ b/src/render.c
@@ -44,12 +44,30 @@ void init_render(Config *config)
case MODE_CPU:
init_cpu(&_config->config_cpu);
glutDisplayFunc(render_cpu);
- glutIdleFunc(idle_cpu);
+ if (_config->zoomSpecific)
+ {
+ glutIdleFunc(idle_cpu);
+ }
+ else
+ {
+ glutIdleFunc(idle_cpu_dummy);
+ glutKeyboardFunc(keyboard_cpu);
+ glutMouseFunc(mouse_cpu);
+ }
break;
case MODE_OPENCL:
init_opencl(&_config->config_opencl);
glutDisplayFunc(render_opencl);
- glutIdleFunc(idle_opencl);
+ if (_config->zoomSpecific)
+ {
+ glutIdleFunc(idle_opencl);
+ }
+ else
+ {
+ glutIdleFunc(idle_opencl_dummy);
+ glutKeyboardFunc(keyboard_opencl);
+ glutMouseFunc(mouse_opencl);
+ }
break;
}
diff --git a/src/render.h b/src/render.h
index a50027a..52be446 100644
--- a/src/render.h
+++ b/src/render.h
@@ -17,6 +17,7 @@ typedef struct config {
CpuConfig config_cpu;
OpenCLConfig config_opencl;
u8 mode;
+ u8 zoomSpecific;
u8 video;
u8 filetype;
u16 width;
diff --git a/src/render_cpu.c b/src/render_cpu.c
index 5065d4d..64fc9ec 100644
--- a/src/render_cpu.c
+++ b/src/render_cpu.c
@@ -8,6 +8,15 @@
#include "render_cpu.h"
#define HAVE_STRUCT_TIMESPEC
+void idle_cpu_dummy(void)
+{
+// glBindTexture(GL_TEXTURE_2D, config_cpu->tex);
+// glTexImage2D(GL_TEXTURE_2D, 0, GL_RGBA, config_cpu->width,
+// config_cpu->height, 0, GL_RGBA, GL_UNSIGNED_BYTE, config_cpu->arr);
+// glBindTexture(GL_TEXTURE_2D, 0);
+ glutPostRedisplay();
+}
+
void init_cpu(CpuConfig *config)
{
config_cpu = config;
@@ -21,10 +30,15 @@ void init_cpu(CpuConfig *config)
y_max = y_max_s;
calculate(x_min, y_min, x_max, y_max, config_cpu->set_func, config_cpu->arr);
+ t_old_cpu = glutGet(GLUT_ELAPSED_TIME);
}
void render_cpu(void)
{
+ glBindTexture(GL_TEXTURE_2D, config_cpu->tex);
+ glTexImage2D(GL_TEXTURE_2D, 0, GL_RGBA, config_cpu->width,
+ config_cpu->height, 0, GL_RGBA, GL_UNSIGNED_BYTE, config_cpu->arr);
+ glBindTexture(GL_TEXTURE_2D, 0);
glClear(GL_COLOR_BUFFER_BIT | GL_DEPTH_BUFFER_BIT);
glBindTexture(GL_TEXTURE_2D, config_cpu->tex);
glEnable(GL_TEXTURE_2D);
@@ -66,6 +80,9 @@ void calculate_t(void *args)
x_math = _args->x_min + ((d64) x * (_args->x_max - _args->x_min)) / config_cpu->width;
y_math = _args->y_min + ((d64) (config_cpu->height - y) * (_args->y_max - _args->y_min)) / config_cpu->height;
iterations = _args->sfunc(x_math, y_math, config_cpu->iterations);
+// color[2].s0 = ((1.0f + native_cos(c)) * 0.5f) * 255;
+// color[2].s1 = ((1.0f + native_cos(2.0f * c + 2.0f * 3.1416f / 3.0f)) * 0.5f) * 255;
+// color[2].s2 = ((1.0f + native_cos(c - 2.0f * 3.1416f / 3.0f)) * 0.5f) * 255;
_args->arr[COORDS(x, y, config_cpu->width)] = (((1<<24)-1)*iterations)/config_cpu->iterations;
}
}
@@ -73,15 +90,14 @@ void calculate_t(void *args)
void idle_cpu(void)
{
- static int t_old;
int t = 0, delta = 0;
do
{
t = glutGet(GLUT_ELAPSED_TIME);
- delta = t - t_old;
+ delta = t - t_old_cpu;
}
while(delta < 16); // TODO: Hardcoded FPS
- t_old = t;
+ t_old_cpu = t;
calculate(x_min, y_min, x_max, y_max, config_cpu->set_func, config_cpu->arr);
//glGenTextures(1, &tex);
@@ -97,3 +113,44 @@ void idle_cpu(void)
y_max = y_max_s - config_cpu->zoom_func(ft, (d64)1.0 - config_cpu->to_y);
glutPostRedisplay();
}
+
+void keyboard_cpu(unsigned char key, int mouseX, int mouseY)
+{
+ switch(key)
+ {
+ case 'i':
+ config_cpu->iterations++;
+ break;
+ case 'd':
+ config_cpu->iterations--;
+ break;
+ }
+}
+
+void mouse_cpu(int button, int state, int x, int y)
+{
+ if (state == GLUT_DOWN)
+ {
+ switch (button)
+ {
+ case GLUT_LEFT_BUTTON:
+ if (config_cpu->speed < 0)
+ config_cpu->speed = (-1) * config_cpu->speed;
+ break;
+ case GLUT_RIGHT_BUTTON:
+ if (config_cpu->speed > 0)
+ config_cpu->speed = (-1) * config_cpu->speed;
+ break;
+ }
+ config_cpu->to_x = x_min
+ + ((d64) x * (x_max - x_min)) / config_cpu->width;
+ config_cpu->to_y = y_min
+ + ((d64) y * (y_max - y_min)) / config_cpu->height;
+ t_old_cpu = glutGet(GLUT_ELAPSED_TIME);
+ glutIdleFunc(idle_cpu);
+ }
+ else if (state == GLUT_UP)
+ {
+ glutIdleFunc(idle_cpu_dummy);
+ }
+}
diff --git a/src/render_cpu.h b/src/render_cpu.h
index 5d32661..7469ca2 100644
--- a/src/render_cpu.h
+++ b/src/render_cpu.h
@@ -49,6 +49,7 @@ typedef struct t_args {
CpuConfig *config_cpu;
u32 rendercnt;
float ft;
+int t_old_cpu;
d64 x_min, x_max, y_min, y_max;
d64 x_min_s, x_max_s, y_min_s, y_max_s;
@@ -56,6 +57,9 @@ d64 x_min_s, x_max_s, y_min_s, y_max_s;
void init_cpu(CpuConfig *config);
void render_cpu(void);
void idle_cpu(void);
+void idle_cpu_dummy(void);
+void keyboard_cpu(unsigned char key, int mouseX, int mouseY);
+void mouse_cpu(int button, int state, int x, int y);
void calculate(d64 x_min, d64 y_min, d64 x_max, d64 y_max, u32 (*sfunc) (d64, d64, u32), u32 *arr);
void calculate_t(void *args);
diff --git a/src/render_opencl.c b/src/render_opencl.c
index 84fc314..055a7f3 100644
--- a/src/render_opencl.c
+++ b/src/render_opencl.c
@@ -7,16 +7,21 @@
#include "render_opencl.h"
+void idle_opencl_dummy(void)
+{
+ glutPostRedisplay();
+}
+
void init_opencl(OpenCLConfig *config)
{
- x_min_s = -2.0;
- x_max_s = 1.0;
- y_min_s = -1.0;
- y_max_s = 1.0;
- x_min = x_min_s;
- x_max = x_max_s;
- y_min = y_min_s;
- y_max = y_max_s;
+ x_min_s_cl = -2.0;
+ x_max_s_cl = 1.0;
+ y_min_s_cl = -1.0;
+ y_max_s_cl = 1.0;
+ x_min_cl = x_min_s_cl;
+ x_max_cl = x_max_s_cl;
+ y_min_cl = y_min_s_cl;
+ y_max_cl = y_max_s_cl;
config_opencl = config;
output = (cl_uint *) malloc((config_opencl->width) * (config_opencl->height) * sizeof(cl_uchar4));
@@ -126,6 +131,7 @@ void init_opencl(OpenCLConfig *config)
{
flags = strcat(flags, "-D MUL_ADD=mad ");
}
+ printf("flags: %s\n", flags);
program = clCreateProgramWithSource(context, 1, (const char **)&cl_src, (const size_t *)&cl_src_sz, &ret);
ret = clBuildProgram(program, num_devices, devices, flags, NULL, NULL);
for (cl_uint i = 0; i < num_devices; i++)
@@ -147,12 +153,12 @@ void render_opencl(void)
cl_double y_max_t;
cl_float y_max_t_f;
- cl_double x_delta = ((x_max - x_min) / (double) config_opencl->width);
- cl_double y_delta = -((y_max - y_min) / (double) config_opencl->height);
+ cl_double x_delta = ((x_max_cl - x_min_cl) / (double) config_opencl->width);
+ cl_double y_delta = -((y_max_cl - y_min_cl) / (double) config_opencl->height);
cl_float x_delta_f = (float) x_delta;
cl_float y_delta_f = (float) y_delta;
- cl_float x_min_f = (float) x_min;
+ cl_float x_min_f = (float) x_min_cl;
globalThreads[0] = ((config_opencl->width) * (config_opencl->height))
/ num_devices;
@@ -171,7 +177,7 @@ void render_opencl(void)
localThreads[0] = kernelWorkGroupSize;
}
- y_max_t = (((y_min + y_max) / 2.0) + (y_max - y_min) / 2.0 - ((double) i * (y_max - y_min)) / (double) num_devices);
+ y_max_t = (((y_min_cl + y_max_cl) / 2.0) + (y_max_cl - y_min_cl) / 2.0 - ((double) i * (y_max_cl - y_min_cl)) / (double) num_devices);
ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *) &outputBuffer[i]);
@@ -189,7 +195,7 @@ void render_opencl(void)
ret = clSetKernelArg(kernel, 4, sizeof(cl_float), (void *) &y_delta_f);
break;
case OPENCL_FPU_64:
- ret = clSetKernelArg(kernel, 1, sizeof(cl_double), (void *) &x_min);
+ ret = clSetKernelArg(kernel, 1, sizeof(cl_double), (void *) &x_min_cl);
ret = clSetKernelArg(kernel, 2, sizeof(cl_double), (void *) &y_max_t);
ret = clSetKernelArg(kernel, 3, sizeof(cl_double), (void *) &x_delta);
ret = clSetKernelArg(kernel, 4, sizeof(cl_double), (void *) &y_delta);
@@ -238,6 +244,11 @@ void render_opencl(void)
ret = clReleaseEvent(events[num_devices - i - 1]);
}
+ glBindTexture(GL_TEXTURE_2D, config_opencl->tex);
+ glTexImage2D(GL_TEXTURE_2D, 0, GL_RGBA, config_opencl->width,
+ config_opencl->height, 0, GL_RGBA, GL_UNSIGNED_BYTE, config_opencl->arr);
+ glBindTexture(GL_TEXTURE_2D, 0);
+
glClear(GL_COLOR_BUFFER_BIT | GL_DEPTH_BUFFER_BIT);
glBindTexture(GL_TEXTURE_2D, config_opencl->tex);
glEnable(GL_TEXTURE_2D);
@@ -256,14 +267,13 @@ void render_opencl(void)
void idle_opencl(void)
{
- static int t_old;
int t = 0, delta = 0;
do
{
t = glutGet(GLUT_ELAPSED_TIME);
- delta = t - t_old;
+ delta = t - t_old_opencl;
} while (delta < 16); // TODO: Hardcoded FPS
- t_old = t;
+ t_old_opencl = t;
//glGenTextures(1, &tex);
glBindTexture(GL_TEXTURE_2D, config_opencl->tex);
@@ -273,9 +283,48 @@ void idle_opencl(void)
config_opencl->height, 0, GL_RGBA, GL_UNSIGNED_BYTE, config_opencl->arr);
glBindTexture(GL_TEXTURE_2D, 0);
cl_ft += (config_opencl->speed * (delta / 1000.0));
- x_min = x_min_s + config_opencl->zoom_func(cl_ft, (d64) 2.0 + config_opencl->to_x);
- y_min = y_min_s + config_opencl->zoom_func(cl_ft, (d64) 1.0 + config_opencl->to_y);
- x_max = x_max_s - config_opencl->zoom_func(cl_ft, (d64) 1.0 - config_opencl->to_x);
- y_max = y_max_s - config_opencl->zoom_func(cl_ft, (d64) 1.0 - config_opencl->to_y);
+ x_min_cl = x_min_s_cl + config_opencl->zoom_func(cl_ft, (cl_double) 2.0 + config_opencl->to_x);
+ y_min_cl = y_min_s_cl + config_opencl->zoom_func(cl_ft, (cl_double) 1.0 + config_opencl->to_y);
+ x_max_cl = x_max_s_cl - config_opencl->zoom_func(cl_ft, (cl_double) 1.0 - config_opencl->to_x);
+ y_max_cl = y_max_s_cl - config_opencl->zoom_func(cl_ft, (cl_double) 1.0 - config_opencl->to_y);
glutPostRedisplay();
}
+
+void keyboard_opencl(unsigned char key, int mouseX, int mouseY)
+{
+ switch (key)
+ {
+ case 'i':
+ config_opencl->iterations++;
+ break;
+ case 'd':
+ config_opencl->iterations--;
+ break;
+ }
+}
+
+void mouse_opencl(int button, int state, int x, int y)
+{
+ if (state == GLUT_DOWN)
+ {
+ switch (button)
+ {
+ case GLUT_LEFT_BUTTON:
+ if (config_opencl->speed < 0)
+ config_opencl->speed = (-1) * config_opencl->speed;
+ break;
+ case GLUT_RIGHT_BUTTON:
+ if (config_opencl->speed > 0)
+ config_opencl->speed = (-1) * config_opencl->speed;
+ break;
+ }
+ config_opencl->to_x = x_min_cl + ((d64) x * (x_max_cl - x_min_cl)) / config_opencl->width;
+ config_opencl->to_y = y_min_cl + ((d64) y * (y_max_cl - y_min_cl)) / config_opencl->height;
+ t_old_opencl = glutGet(GLUT_ELAPSED_TIME);
+ glutIdleFunc(idle_opencl);
+ }
+ else if(state == GLUT_UP)
+ {
+ glutIdleFunc(idle_opencl_dummy);
+ }
+}
diff --git a/src/render_opencl.h b/src/render_opencl.h
index 720678b..bca77ad 100644
--- a/src/render_opencl.h
+++ b/src/render_opencl.h
@@ -42,8 +42,8 @@ typedef struct config_opencl {
OpenCLConfig *config_opencl;
-d64 x_min, x_max, y_min, y_max;
-d64 x_min_s, x_max_s, y_min_s, y_max_s;
+cl_double x_min_cl, x_max_cl, y_min_cl, y_max_cl;
+cl_double x_min_s_cl, x_max_s_cl, y_min_s_cl, y_max_s_cl;
float cl_ft;
cl_uint *output;
@@ -58,10 +58,15 @@ cl_mem outputBuffer[MAX_DEVICES];
cl_int width_cl;
+int t_old_opencl;
+
d64 zoom_func(d64 ft, d64 s);
void init_opencl(OpenCLConfig *config);
void render_opencl(void);
void idle_opencl(void);
+void idle_opencl_dummy(void);
+void keyboard_opencl(unsigned char key, int mouseX, int mouseY);
+void mouse_opencl(int button, int state, int x, int y);
#endif /* RENDER_OPENCL_H_ */