|
@@ -6,7 +6,7 @@
|
|
|
#endif
|
|
|
|
|
|
#include <ufo/ufo-resource-manager.h>
|
|
|
-#include <ufo/ufo-filter.h>
|
|
|
+#include <ufo/ufo-filter-source.h>
|
|
|
#include <ufo/ufo-buffer.h>
|
|
|
#include "ufo-filter-lamino-ramp.h"
|
|
|
|
|
@@ -19,7 +19,6 @@
|
|
|
*/
|
|
|
|
|
|
struct _UfoFilterLaminoRampPrivate {
|
|
|
-
|
|
|
// filter extent
|
|
|
guint width; // is pow of 2
|
|
|
guint fill_width;
|
|
@@ -27,11 +26,12 @@ struct _UfoFilterLaminoRampPrivate {
|
|
|
// laminographic angle
|
|
|
float theta;
|
|
|
// pixel resolution (in um)
|
|
|
- float tau;
|
|
|
+ float tau;
|
|
|
cl_kernel kernel_cf;
|
|
|
+ gboolean done;
|
|
|
};
|
|
|
|
|
|
-G_DEFINE_TYPE(UfoFilterLaminoRamp, ufo_filter_lamino_ramp, UFO_TYPE_FILTER)
|
|
|
+G_DEFINE_TYPE(UfoFilterLaminoRamp, ufo_filter_lamino_ramp, UFO_TYPE_FILTER_SOURCE)
|
|
|
|
|
|
#define UFO_FILTER_LAMINO_RAMP_GET_PRIVATE(obj) (G_TYPE_INSTANCE_GET_PRIVATE((obj), UFO_TYPE_FILTER_LAMINO_RAMP, UfoFilterLaminoRampPrivate))
|
|
|
|
|
@@ -47,87 +47,82 @@ enum {
|
|
|
|
|
|
static GParamSpec *lamino_ramp_properties[N_PROPERTIES] = { NULL, };
|
|
|
|
|
|
+static int
|
|
|
+is_power_of_two (guint x)
|
|
|
+{
|
|
|
+ return ((x != 0) && !(x & (x - 1)));
|
|
|
+}
|
|
|
|
|
|
-static void ufo_filter_lamino_ramp_initialize(UfoFilter *filter)
|
|
|
+static void
|
|
|
+ufo_filter_lamino_ramp_initialize (UfoFilterSource *filter, guint **dim_sizes, GError **error)
|
|
|
{
|
|
|
-
|
|
|
- UfoFilterLaminoRamp *self = UFO_FILTER_LAMINO_RAMP(filter);
|
|
|
+ UfoFilterLaminoRampPrivate *priv = UFO_FILTER_LAMINO_RAMP_GET_PRIVATE(filter);
|
|
|
UfoResourceManager *manager = ufo_resource_manager();
|
|
|
- GError *error = NULL;
|
|
|
+ GError *tmp_error = NULL;
|
|
|
|
|
|
- self->priv->kernel_cf = ufo_resource_manager_get_kernel(manager, "lamino_ramp.cl", "lamino_ramp_create_filter", &error);
|
|
|
+ priv->kernel_cf = ufo_resource_manager_get_kernel(manager, "lamino_ramp.cl", "lamino_ramp_create_filter", &tmp_error);
|
|
|
|
|
|
- if (error != NULL) {
|
|
|
- g_warning("%s", error->message);
|
|
|
- g_error_free(error);
|
|
|
+ if (tmp_error != NULL) {
|
|
|
+ g_propagate_error (error, tmp_error);
|
|
|
+ return;
|
|
|
}
|
|
|
-}
|
|
|
|
|
|
-static int is_power_of_two (guint x)
|
|
|
-{
|
|
|
- return ((x != 0) && !(x & (x - 1)));
|
|
|
+ if (!is_power_of_two (priv->width)) {
|
|
|
+ g_set_error (error, UFO_FILTER_ERROR, UFO_FILTER_ERROR_INITIALIZATION,
|
|
|
+ "Filter width `%i` is not a power of two", priv->width);
|
|
|
+ return;
|
|
|
+ }
|
|
|
+
|
|
|
+ if (!is_power_of_two (priv->height)) {
|
|
|
+ g_set_error (error, UFO_FILTER_ERROR, UFO_FILTER_ERROR_INITIALIZATION,
|
|
|
+ "Filter height `%i` is not a power of two", priv->height);
|
|
|
+ return;
|
|
|
+ }
|
|
|
+
|
|
|
+ dim_sizes[0][0] = priv->width;
|
|
|
+ dim_sizes[0][1] = priv->height;
|
|
|
}
|
|
|
|
|
|
-static void ufo_filter_lamino_ramp_process(UfoFilter *filter)
|
|
|
+static gboolean
|
|
|
+ufo_filter_lamino_ramp_generate (UfoFilterSource *filter, UfoBuffer *output[], gpointer cmd_queue, GError **error)
|
|
|
{
|
|
|
- g_return_if_fail(UFO_IS_FILTER(filter));
|
|
|
- UfoFilterLaminoRamp *self = UFO_FILTER_LAMINO_RAMP(filter);
|
|
|
- UfoFilterLaminoRampPrivate *priv = self->priv = UFO_FILTER_LAMINO_RAMP_GET_PRIVATE(self);
|
|
|
-
|
|
|
- UfoChannel *output_channel = ufo_filter_get_output_channel(filter);
|
|
|
- cl_command_queue command_queue = (cl_command_queue) ufo_filter_get_command_queue(filter);
|
|
|
-
|
|
|
- UfoBuffer *output = NULL;
|
|
|
-
|
|
|
- guint width = priv->width; // check width for power of 2
|
|
|
- if(!is_power_of_two(width))
|
|
|
- g_warning("filter width is not the power of two");
|
|
|
+ g_return_val_if_fail (UFO_IS_FILTER (filter), FALSE);
|
|
|
+ UfoFilterLaminoRampPrivate *priv = UFO_FILTER_LAMINO_RAMP_GET_PRIVATE (filter);
|
|
|
|
|
|
- guint fwidth = priv->fill_width;
|
|
|
+ if (priv->done)
|
|
|
+ return FALSE;
|
|
|
|
|
|
+ guint width = priv->width;
|
|
|
+ guint fwidth = priv->fill_width;
|
|
|
guint height = priv->height;
|
|
|
- if(!is_power_of_two(height))
|
|
|
- g_warning("filter height is not the power of two");
|
|
|
-
|
|
|
- float theta = priv-> theta;
|
|
|
+ float theta = priv->theta;
|
|
|
float tau = priv->tau;
|
|
|
-
|
|
|
- guint dim_size[2] = {width, height};
|
|
|
- size_t global_work_size[2] = { (size_t) width, (size_t) height };
|
|
|
-
|
|
|
- ufo_channel_allocate_output_buffers(output_channel, 2, dim_size);
|
|
|
+ size_t global_work_size[2] = { (size_t) priv->width, (size_t) priv->height };
|
|
|
|
|
|
cl_kernel kernel = priv->kernel_cf;
|
|
|
+ cl_mem output_mem = (cl_mem) ufo_buffer_get_device_array(output[0], (cl_command_queue) cmd_queue);
|
|
|
|
|
|
- output = ufo_channel_get_output_buffer(output_channel);
|
|
|
- cl_mem output_mem = (cl_mem) ufo_buffer_get_device_array(output, command_queue);
|
|
|
-
|
|
|
-
|
|
|
- CHECK_OPENCL_ERROR(clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *) &output_mem));
|
|
|
+ CHECK_OPENCL_ERROR(clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *) &output_mem));
|
|
|
CHECK_OPENCL_ERROR(clSetKernelArg(kernel, 1, sizeof(int), &width));
|
|
|
CHECK_OPENCL_ERROR(clSetKernelArg(kernel, 2, sizeof(int), &fwidth));
|
|
|
CHECK_OPENCL_ERROR(clSetKernelArg(kernel, 3, sizeof(int), &height));
|
|
|
CHECK_OPENCL_ERROR(clSetKernelArg(kernel, 4, sizeof(float), &theta));
|
|
|
- CHECK_OPENCL_ERROR(clSetKernelArg(kernel, 5, sizeof(float), &tau));
|
|
|
-
|
|
|
- cl_event event;
|
|
|
- CHECK_OPENCL_ERROR(clEnqueueNDRangeKernel(command_queue, kernel,
|
|
|
- 2, NULL, global_work_size, NULL,
|
|
|
- 0, NULL, &event));
|
|
|
- ufo_buffer_attach_event(output, event);
|
|
|
-
|
|
|
- ufo_channel_finalize_output_buffer(output_channel, output);
|
|
|
- ufo_channel_finish(output_channel);
|
|
|
+ CHECK_OPENCL_ERROR(clSetKernelArg(kernel, 5, sizeof(float), &tau));
|
|
|
+
|
|
|
+ CHECK_OPENCL_ERROR(clEnqueueNDRangeKernel((cl_command_queue) cmd_queue, kernel,
|
|
|
+ 2, NULL, global_work_size, NULL,
|
|
|
+ 0, NULL, NULL));
|
|
|
+
|
|
|
+ priv->done = TRUE;
|
|
|
+
|
|
|
+ return TRUE;
|
|
|
}
|
|
|
|
|
|
-static void ufo_filter_lamino_ramp_set_property(GObject *object,
|
|
|
- guint property_id,
|
|
|
- const GValue *value,
|
|
|
- GParamSpec *pspec)
|
|
|
+static void
|
|
|
+ufo_filter_lamino_ramp_set_property (GObject *object, guint property_id, const GValue *value, GParamSpec *pspec)
|
|
|
{
|
|
|
UfoFilterLaminoRamp *self = UFO_FILTER_LAMINO_RAMP(object);
|
|
|
|
|
|
- /* Handle all properties accordingly */
|
|
|
switch (property_id) {
|
|
|
case PROP_WIDTH:
|
|
|
self->priv->width = g_value_get_uint(value);
|
|
@@ -135,14 +130,14 @@ static void ufo_filter_lamino_ramp_set_property(GObject *object,
|
|
|
case PROP_FILL_WIDTH:
|
|
|
self->priv->fill_width = g_value_get_uint(value);
|
|
|
break;
|
|
|
- case PROP_HEIGHT:
|
|
|
+ case PROP_HEIGHT:
|
|
|
self->priv->height = g_value_get_uint(value);
|
|
|
break;
|
|
|
- case PROP_THETA:
|
|
|
+ case PROP_THETA:
|
|
|
self->priv->theta = (float) g_value_get_double(value);
|
|
|
break;
|
|
|
- case PROP_TAU:
|
|
|
- self->priv->tau = (float) g_value_get_double(value);
|
|
|
+ case PROP_TAU:
|
|
|
+ self->priv->tau = (float) g_value_get_double(value);
|
|
|
break;
|
|
|
default:
|
|
|
G_OBJECT_WARN_INVALID_PROPERTY_ID(object, property_id, pspec);
|
|
@@ -150,25 +145,23 @@ static void ufo_filter_lamino_ramp_set_property(GObject *object,
|
|
|
}
|
|
|
}
|
|
|
|
|
|
-static void ufo_filter_lamino_ramp_get_property(GObject *object,
|
|
|
- guint property_id,
|
|
|
- GValue *value,
|
|
|
- GParamSpec *pspec)
|
|
|
+static void
|
|
|
+ufo_filter_lamino_ramp_get_property (GObject *object, guint property_id, GValue *value, GParamSpec *pspec)
|
|
|
{
|
|
|
UfoFilterLaminoRamp *self = UFO_FILTER_LAMINO_RAMP(object);
|
|
|
|
|
|
switch (property_id) {
|
|
|
- case PROP_WIDTH:
|
|
|
- g_value_set_uint(value, self->priv->width);
|
|
|
- break;
|
|
|
+ case PROP_WIDTH:
|
|
|
+ g_value_set_uint(value, self->priv->width);
|
|
|
+ break;
|
|
|
case PROP_FILL_WIDTH:
|
|
|
- g_value_set_uint(value, self->priv->fill_width);
|
|
|
- break;
|
|
|
+ g_value_set_uint(value, self->priv->fill_width);
|
|
|
+ break;
|
|
|
case PROP_HEIGHT:
|
|
|
- g_value_set_uint(value, self->priv->height);
|
|
|
- break;
|
|
|
- case PROP_THETA:
|
|
|
- g_value_set_double(value, self->priv->theta);
|
|
|
+ g_value_set_uint(value, self->priv->height);
|
|
|
+ break;
|
|
|
+ case PROP_THETA:
|
|
|
+ g_value_set_double(value, self->priv->theta);
|
|
|
break;
|
|
|
case PROP_TAU:
|
|
|
g_value_set_double(value, self->priv->tau);
|
|
@@ -179,52 +172,53 @@ static void ufo_filter_lamino_ramp_get_property(GObject *object,
|
|
|
}
|
|
|
}
|
|
|
|
|
|
-static void ufo_filter_lamino_ramp_class_init(UfoFilterLaminoRampClass *klass)
|
|
|
+static void
|
|
|
+ufo_filter_lamino_ramp_class_init(UfoFilterLaminoRampClass *klass)
|
|
|
{
|
|
|
GObjectClass *gobject_class = G_OBJECT_CLASS(klass);
|
|
|
- UfoFilterClass *filter_class = UFO_FILTER_CLASS(klass);
|
|
|
+ UfoFilterSourceClass *filter_class = UFO_FILTER_SOURCE_CLASS(klass);
|
|
|
|
|
|
gobject_class->set_property = ufo_filter_lamino_ramp_set_property;
|
|
|
gobject_class->get_property = ufo_filter_lamino_ramp_get_property;
|
|
|
filter_class->initialize = ufo_filter_lamino_ramp_initialize;
|
|
|
- filter_class->process = ufo_filter_lamino_ramp_process;
|
|
|
+ filter_class->generate = ufo_filter_lamino_ramp_generate;
|
|
|
|
|
|
lamino_ramp_properties[PROP_WIDTH] =
|
|
|
- g_param_spec_uint("width",
|
|
|
- "Width of the 2D image filter (power of 2)",
|
|
|
- "Width of the 2D image filter (power of 2)",
|
|
|
- 1, 32768, 1.0,
|
|
|
- G_PARAM_READWRITE);
|
|
|
-
|
|
|
+ g_param_spec_uint("width",
|
|
|
+ "Width of the 2D image filter (power of 2)",
|
|
|
+ "Width of the 2D image filter (power of 2)",
|
|
|
+ 1, 32768, 1.0,
|
|
|
+ G_PARAM_READWRITE);
|
|
|
+
|
|
|
lamino_ramp_properties[PROP_FILL_WIDTH] =
|
|
|
g_param_spec_uint("fwidth",
|
|
|
- "Filling width of the 2D image filter",
|
|
|
- "Filling width of the 2D image filter",
|
|
|
- 1, 32768, 1.0,
|
|
|
- G_PARAM_READWRITE);
|
|
|
-
|
|
|
- lamino_ramp_properties[PROP_HEIGHT] =
|
|
|
- g_param_spec_uint("height",
|
|
|
- "Height of the 2D image filter",
|
|
|
- "Height of the 2D image filter",
|
|
|
- 1, 16384, 1.0,
|
|
|
- G_PARAM_READWRITE);
|
|
|
-
|
|
|
- lamino_ramp_properties[PROP_THETA] =
|
|
|
- g_param_spec_double("theta",
|
|
|
- "Laminographic angle in radians",
|
|
|
- "Resolution (pixel size) in microns",
|
|
|
- -4.0 * G_PI, +4.0 * G_PI, 0.0,
|
|
|
- G_PARAM_READWRITE);
|
|
|
-
|
|
|
- lamino_ramp_properties[PROP_TAU] =
|
|
|
+ "Filling width of the 2D image filter",
|
|
|
+ "Filling width of the 2D image filter",
|
|
|
+ 1, 32768, 1.0,
|
|
|
+ G_PARAM_READWRITE);
|
|
|
+
|
|
|
+ lamino_ramp_properties[PROP_HEIGHT] =
|
|
|
+ g_param_spec_uint("height",
|
|
|
+ "Height of the 2D image filter",
|
|
|
+ "Height of the 2D image filter",
|
|
|
+ 1, 16384, 1.0,
|
|
|
+ G_PARAM_READWRITE);
|
|
|
+
|
|
|
+ lamino_ramp_properties[PROP_THETA] =
|
|
|
+ g_param_spec_double("theta",
|
|
|
+ "Laminographic angle in radians",
|
|
|
+ "Resolution (pixel size) in microns",
|
|
|
+ -4.0 * G_PI, +4.0 * G_PI, 0.0,
|
|
|
+ G_PARAM_READWRITE);
|
|
|
+
|
|
|
+ lamino_ramp_properties[PROP_TAU] =
|
|
|
g_param_spec_double("tau",
|
|
|
- "Resolution (pixel size) in microns",
|
|
|
- "Resolution (pixel size) in microns",
|
|
|
- 0.0, /* minimum */
|
|
|
- 100000.0, /* maximum */
|
|
|
- 10.0, /* default */
|
|
|
- G_PARAM_READWRITE);
|
|
|
+ "Resolution (pixel size) in microns",
|
|
|
+ "Resolution (pixel size) in microns",
|
|
|
+ 0.0, /* minimum */
|
|
|
+ 100000.0, /* maximum */
|
|
|
+ 10.0, /* default */
|
|
|
+ G_PARAM_READWRITE);
|
|
|
|
|
|
g_object_class_install_property(gobject_class, PROP_WIDTH, lamino_ramp_properties[PROP_WIDTH]);
|
|
|
g_object_class_install_property(gobject_class, PROP_FILL_WIDTH, lamino_ramp_properties[PROP_FILL_WIDTH]);
|
|
@@ -244,12 +238,13 @@ static void ufo_filter_lamino_ramp_init(UfoFilterLaminoRamp *self)
|
|
|
priv->height = 1;
|
|
|
priv->theta = 0.0;
|
|
|
priv->tau = 10.0;
|
|
|
+ priv->done = FALSE;
|
|
|
|
|
|
- // ufo_filter_register_input(UFO_FILTER(self), "input0", 2);
|
|
|
- ufo_filter_register_output(UFO_FILTER(self), "filter", 2);
|
|
|
+ ufo_filter_register_outputs (UFO_FILTER(self), 2, NULL);
|
|
|
}
|
|
|
|
|
|
-G_MODULE_EXPORT UfoFilter *ufo_filter_plugin_new(void)
|
|
|
+G_MODULE_EXPORT UfoFilter *
|
|
|
+ufo_filter_plugin_new(void)
|
|
|
{
|
|
|
return g_object_new(UFO_TYPE_FILTER_LAMINO_RAMP, NULL);
|
|
|
}
|