[gegl] error handling in gegl-buffer-cl-iterator
- From: Ãyvind KolÃs <ok src gnome org>
- To: commits-list gnome org
- Cc:
- Subject: [gegl] error handling in gegl-buffer-cl-iterator
- Date: Tue, 20 Mar 2012 13:51:49 +0000 (UTC)
commit 8e77631ade779d6cfef7c672b128444ae3162153
Author: Victor Oliveira <victormatheus gmail com>
Date: Mon Jan 16 13:38:04 2012 -0200
error handling in gegl-buffer-cl-iterator
gegl/buffer/gegl-buffer-cl-iterator.c | 84 ++++++++++++++-----------
gegl/buffer/gegl-buffer-cl-iterator.h | 2 +-
gegl/operation/gegl-operation-point-filter.c | 34 +++++-----
gegl/operation/gegl-operation-point-filter.h | 2 +-
operations/common/brightness-contrast.c | 30 ++++-----
5 files changed, 79 insertions(+), 73 deletions(-)
---
diff --git a/gegl/buffer/gegl-buffer-cl-iterator.c b/gegl/buffer/gegl-buffer-cl-iterator.c
index 4a2d588..df042ff 100644
--- a/gegl/buffer/gegl-buffer-cl-iterator.c
+++ b/gegl/buffer/gegl-buffer-cl-iterator.c
@@ -14,7 +14,7 @@
#include "gegl-tile-storage.h"
#include "gegl-utils.h"
-#define CL_ERROR {g_printf("[OpenCL] Error in %s:%d %s - %s\n", __FILE__, __LINE__, __func__, gegl_cl_errstring(errcode)); g_assert(FALSE);}
+#define CL_ERROR {g_printf("[OpenCL] Error in %s:%d %s - %s\n", __FILE__, __LINE__, __func__, gegl_cl_errstring(cl_err)); goto error;}
typedef struct GeglBufferClIterators
{
@@ -129,11 +129,12 @@ gegl_buffer_cl_iterator_add (GeglBufferClIterator *iterator,
}
gboolean
-gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
+gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator, gboolean *err)
{
GeglBufferClIterators *i = (gpointer)iterator;
gboolean result = FALSE;
- gint no;
+ gint no, j;
+ cl_int cl_err = 0;
const size_t origin_zero[3] = {0, 0, 0};
@@ -157,17 +158,14 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
}
else
{
- gint j;
- cl_int errcode = 0;
-
/* complete pending write work */
for (no=0; no<i->iterators;no++)
{
if (i->flags[no] == GEGL_CL_BUFFER_WRITE)
{
/* Wait Processing */
- errcode = gegl_clEnqueueBarrier(gegl_cl_get_command_queue());
- if (errcode != CL_SUCCESS) CL_ERROR;
+ cl_err = gegl_clEnqueueBarrier(gegl_cl_get_command_queue());
+ if (cl_err != CL_SUCCESS) CL_ERROR;
/* color conversion in the GPU (output) */
if (i->conv[no] == CL_COLOR_CONVERT)
@@ -179,17 +177,17 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
&i->buf_cl_format[no],
i->roi[no][j].width,
i->roi[no][j].height,
- 0, NULL, &errcode);
- if (errcode != CL_SUCCESS) CL_ERROR;
+ 0, NULL, &cl_err);
+ if (cl_err != CL_SUCCESS) CL_ERROR;
- errcode = gegl_cl_color_conv (&i->tex[no][j], &i->tex_aux[no][j], i->size[no][j],
+ cl_err = gegl_cl_color_conv (&i->tex[no][j], &i->tex_aux[no][j], i->size[no][j],
i->format[no], i->buffer[no]->format);
- if (errcode != CL_SUCCESS) CL_ERROR;
+ if (cl_err == FALSE) CL_ERROR;
}
/* Wait Processing */
- errcode = gegl_clEnqueueBarrier(gegl_cl_get_command_queue());
- if (errcode != CL_SUCCESS) CL_ERROR;
+ cl_err = gegl_clEnqueueBarrier(gegl_cl_get_command_queue());
+ if (cl_err != CL_SUCCESS) CL_ERROR;
/* GPU -> CPU */
for (j=0; j < i->n; j++)
@@ -201,8 +199,8 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
data = gegl_clEnqueueMapImage(gegl_cl_get_command_queue(), i->tex[no][j], CL_TRUE,
CL_MAP_READ,
origin_zero, region, &pitch, NULL,
- 0, NULL, NULL, &errcode);
- if (errcode != CL_SUCCESS) CL_ERROR;
+ 0, NULL, NULL, &cl_err);
+ if (cl_err != CL_SUCCESS) CL_ERROR;
/* tile-ize */
if (i->conv[no] == CL_COLOR_NOT_SUPPORTED)
@@ -212,16 +210,16 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
/* color conversion has already been performed in the GPU */
gegl_buffer_set (i->buffer[no], &i->roi[no][j], i->buffer[no]->format, data, pitch);
- errcode = gegl_clEnqueueUnmapMemObject (gegl_cl_get_command_queue(), i->tex[no][j], data,
+ cl_err = gegl_clEnqueueUnmapMemObject (gegl_cl_get_command_queue(), i->tex[no][j], data,
0, NULL, NULL);
- if (errcode != CL_SUCCESS) CL_ERROR;
+ if (cl_err != CL_SUCCESS) CL_ERROR;
}
}
}
/* Run! */
- errcode = gegl_clFinish(gegl_cl_get_command_queue());
- if (errcode != CL_SUCCESS) CL_ERROR;
+ cl_err = gegl_clFinish(gegl_cl_get_command_queue());
+ if (cl_err != CL_SUCCESS) CL_ERROR;
for (no=0; no < i->iterators; no++)
for (j=0; j < i->n; j++)
@@ -242,9 +240,6 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
/* then we iterate all */
for (no=0; no<i->iterators;no++)
{
- int j;
- int errcode = 0;
-
for (j = 0; j < i->n; j++)
{
GeglRectangle r = {i->rect[no].x + i->roi_all[i->roi_no+j].x,
@@ -273,15 +268,15 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
&i->op_cl_format [no],
i->roi[no][j].width,
i->roi[no][j].height,
- 0, NULL, &errcode);
- if (errcode != CL_SUCCESS) CL_ERROR;
+ 0, NULL, &cl_err);
+ if (cl_err != CL_SUCCESS) CL_ERROR;
/* pre-pinned memory */
data = gegl_clEnqueueMapImage(gegl_cl_get_command_queue(), i->tex[no][j], CL_TRUE,
CL_MAP_WRITE,
origin_zero, region, &pitch, NULL,
- 0, NULL, NULL, &errcode);
- if (errcode != CL_SUCCESS) CL_ERROR;
+ 0, NULL, NULL, &cl_err);
+ if (cl_err != CL_SUCCESS) CL_ERROR;
/* un-tile */
if (i->conv[no] == CL_COLOR_NOT_SUPPORTED)
@@ -291,13 +286,13 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
/* color conversion will be performed in the GPU later */
gegl_buffer_get (i->buffer[no], 1.0, &i->roi[no][j], i->buffer[no]->format, data, pitch);
- errcode = gegl_clEnqueueUnmapMemObject (gegl_cl_get_command_queue(), i->tex[no][j], data,
+ cl_err = gegl_clEnqueueUnmapMemObject (gegl_cl_get_command_queue(), i->tex[no][j], data,
0, NULL, NULL);
- if (errcode != CL_SUCCESS) CL_ERROR;
+ if (cl_err != CL_SUCCESS) CL_ERROR;
}
- errcode = gegl_clEnqueueBarrier(gegl_cl_get_command_queue());
- if (errcode != CL_SUCCESS) CL_ERROR;
+ cl_err = gegl_clEnqueueBarrier(gegl_cl_get_command_queue());
+ if (cl_err != CL_SUCCESS) CL_ERROR;
/* color conversion in the GPU (input) */
if (i->conv[no] == CL_COLOR_CONVERT)
@@ -310,12 +305,12 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
&i->op_cl_format[no],
i->roi[no][j].width,
i->roi[no][j].height,
- 0, NULL, &errcode);
- if (errcode != CL_SUCCESS) CL_ERROR;
+ 0, NULL, &cl_err);
+ if (cl_err != CL_SUCCESS) CL_ERROR;
- errcode = gegl_cl_color_conv (&i->tex[no][j], &i->tex_aux[no][j], i->size[no][j],
+ cl_err = gegl_cl_color_conv (&i->tex[no][j], &i->tex_aux[no][j], i->size[no][j],
i->buffer[no]->format, i->format[no]);
- if (errcode == FALSE) CL_ERROR;
+ if (cl_err == FALSE) CL_ERROR;
}
}
else if (i->flags[no] == GEGL_CL_BUFFER_WRITE)
@@ -329,8 +324,8 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
&i->op_cl_format [no],
i->roi[no][j].width,
i->roi[no][j].height,
- 0, NULL, &errcode);
- if (errcode != CL_SUCCESS) CL_ERROR;
+ 0, NULL, &cl_err);
+ if (cl_err != CL_SUCCESS) CL_ERROR;
}
}
else
@@ -370,7 +365,22 @@ gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator)
g_slice_free (GeglBufferClIterators, i);
}
+ *err = FALSE;
return result;
+
+error:
+
+ for (no=0; no<i->iterators;no++)
+ for (j=0; j < i->n; j++)
+ {
+ if (i->tex_aux[no][j]) gegl_clReleaseMemObject (i->tex_aux[no][j]);
+ if (i->tex[no][j]) gegl_clReleaseMemObject (i->tex[no][j]);
+ i->tex_aux[no][j] = NULL;
+ i->tex[no][j] = NULL;
+ }
+
+ *err = TRUE;
+ return FALSE;
}
GeglBufferClIterator *
diff --git a/gegl/buffer/gegl-buffer-cl-iterator.h b/gegl/buffer/gegl-buffer-cl-iterator.h
index 6bba2a6..5558f9f 100644
--- a/gegl/buffer/gegl-buffer-cl-iterator.h
+++ b/gegl/buffer/gegl-buffer-cl-iterator.h
@@ -27,7 +27,7 @@ gint gegl_buffer_cl_iterator_add (GeglBufferClIterator *iterator,
const Babl *format,
guint flags);
-gboolean gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator);
+gboolean gegl_buffer_cl_iterator_next (GeglBufferClIterator *iterator, gboolean *err);
GeglBufferClIterator *gegl_buffer_cl_iterator_new (GeglBuffer *buffer,
const GeglRectangle *roi,
diff --git a/gegl/operation/gegl-operation-point-filter.c b/gegl/operation/gegl-operation-point-filter.c
index 746ee31..c3eea28 100644
--- a/gegl/operation/gegl-operation-point-filter.c
+++ b/gegl/operation/gegl-operation-point-filter.c
@@ -72,9 +72,6 @@ gegl_operation_point_filter_init (GeglOperationPointFilter *self)
{
}
-//#define CL_ERROR {g_assert(0);}
-#define CL_ERROR {g_printf("[OpenCL] Error in %s:%d %s - %s\n", __FILE__, __LINE__, __func__, gegl_cl_errstring(errcode)); goto error;}
-
static gboolean
gegl_operation_point_filter_cl_process (GeglOperation *operation,
GeglBuffer *input,
@@ -87,7 +84,8 @@ gegl_operation_point_filter_cl_process (GeglOperation *operation,
GeglOperationPointFilterClass *point_filter_class = GEGL_OPERATION_POINT_FILTER_GET_CLASS (operation);
gint j;
- cl_int errcode = 0;
+ cl_int cl_err = 0;
+ gboolean err;
/* non-texturizable format! */
if (!gegl_cl_color_babl (in_format, NULL, NULL) ||
@@ -109,23 +107,25 @@ gegl_operation_point_filter_cl_process (GeglOperation *operation,
{
GeglBufferClIterator *i = gegl_buffer_cl_iterator_new (output, result, out_format, GEGL_CL_BUFFER_WRITE);
gint read = gegl_buffer_cl_iterator_add (i, input, result, in_format, GEGL_CL_BUFFER_READ);
- while (gegl_buffer_cl_iterator_next (i))
- for (j=0; j < i->n; j++)
- {
- errcode = point_filter_class->cl_process(operation, i->tex[read][j], i->tex[0][j],
- i->size[0][j], &i->roi[0][j]);
- if (errcode != CL_SUCCESS) CL_ERROR;
- }
+ while (gegl_buffer_cl_iterator_next (i, &err))
+ {
+ if (err) return FALSE;
+ for (j=0; j < i->n; j++)
+ {
+ cl_err = point_filter_class->cl_process(operation, i->tex[read][j], i->tex[0][j],
+ i->size[0][j], &i->roi[0][j]);
+ if (cl_err != CL_SUCCESS)
+ {
+ g_warning("[OpenCL] Error in %s [GeglOperationPointFilter] Kernel\n",
+ GEGL_OPERATION_CLASS (operation)->name);
+ return FALSE;
+ }
+ }
+ }
}
return TRUE;
-
-error:
-
- return FALSE;
}
-#undef CL_ERROR
-
static gboolean
gegl_operation_point_filter_process (GeglOperation *operation,
GeglBuffer *input,
diff --git a/gegl/operation/gegl-operation-point-filter.h b/gegl/operation/gegl-operation-point-filter.h
index dbcbd90..3e80856 100644
--- a/gegl/operation/gegl-operation-point-filter.h
+++ b/gegl/operation/gegl-operation-point-filter.h
@@ -59,7 +59,7 @@ struct _GeglOperationPointFilterClass
checkerboard op for
semantics */
- gboolean (* cl_process) (GeglOperation *self,
+ cl_int (* cl_process) (GeglOperation *self,
cl_mem in_tex,
cl_mem out_tex,
const size_t global_worksize[2],
diff --git a/operations/common/brightness-contrast.c b/operations/common/brightness-contrast.c
index 4ea8159..3501af5 100644
--- a/operations/common/brightness-contrast.c
+++ b/operations/common/brightness-contrast.c
@@ -126,7 +126,7 @@ static const char* kernel_source =
static gegl_cl_run_data *cl_data = NULL;
/* OpenCL processing function */
-static gboolean
+static cl_int
cl_process (GeglOperation *op,
cl_mem in_tex,
cl_mem out_tex,
@@ -142,7 +142,7 @@ cl_process (GeglOperation *op,
gfloat brightness = o->brightness;
gfloat contrast = o->contrast;
- cl_int errcode = 0;
+ cl_int cl_err = 0;
if (!cl_data)
{
@@ -152,23 +152,19 @@ cl_process (GeglOperation *op,
if (!cl_data) return 1;
- CL_SAFE_CALL(errcode = gegl_clSetKernelArg(cl_data->kernel[0], 0, sizeof(cl_mem), (void*)&in_tex));
- CL_SAFE_CALL(errcode = gegl_clSetKernelArg(cl_data->kernel[0], 1, sizeof(cl_mem), (void*)&out_tex));
- CL_SAFE_CALL(errcode = gegl_clSetKernelArg(cl_data->kernel[0], 2, sizeof(cl_float), (void*)&brightness));
- CL_SAFE_CALL(errcode = gegl_clSetKernelArg(cl_data->kernel[0], 3, sizeof(cl_float), (void*)&contrast));
-
- CL_SAFE_CALL(errcode = gegl_clEnqueueNDRangeKernel(gegl_cl_get_command_queue (),
- cl_data->kernel[0], 2,
- NULL, global_worksize, NULL,
- 0, NULL, NULL) );
+ cl_err |= gegl_clSetKernelArg(cl_data->kernel[0], 0, sizeof(cl_mem), (void*)&in_tex);
+ cl_err |= gegl_clSetKernelArg(cl_data->kernel[0], 1, sizeof(cl_mem), (void*)&out_tex);
+ cl_err |= gegl_clSetKernelArg(cl_data->kernel[0], 2, sizeof(cl_float), (void*)&brightness);
+ cl_err |= gegl_clSetKernelArg(cl_data->kernel[0], 3, sizeof(cl_float), (void*)&contrast);
+ if (cl_err != CL_SUCCESS) return cl_err;
- if (errcode != CL_SUCCESS)
- {
- g_warning("[OpenCL] Error in Brightness-Constrast Kernel\n");
- return errcode;
- }
+ cl_err = gegl_clEnqueueNDRangeKernel(gegl_cl_get_command_queue (),
+ cl_data->kernel[0], 2,
+ NULL, global_worksize, NULL,
+ 0, NULL, NULL);
+ if (cl_err != CL_SUCCESS) return cl_err;
- return errcode;
+ return cl_err;
}
/*
[
Date Prev][
Date Next] [
Thread Prev][
Thread Next]
[
Thread Index]
[
Date Index]
[
Author Index]
