lamino-filters/src/ufo-filter-lamino-ramp.c

#include <gmodule.h>
#ifdef __APPLE__
#include <OpenCL/cl.h>
#else
#include <CL/cl.h>
#endif

#include <ufo/ufo-resource-manager.h>
#include <ufo/ufo-filter-source.h>
#include <ufo/ufo-buffer.h>
#include "ufo-filter-lamino-ramp.h"

/**
 * SECTION:ufo-filter-lamino-ramp
 * @Short_description:
 * @Title: laminoramp
 *
 * Detailed description.
 */

struct _UfoFilterLaminoRampPrivate {
    // filter extent
    guint width; // is pow of 2
    guint fill_width;
    guint height;
    // laminographic angle
    float theta;
    // pixel resolution (in um)
    float tau;
    cl_kernel kernel_cf;
    gboolean done;
};

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))

enum {
    PROP_0,
    PROP_WIDTH,
    PROP_FILL_WIDTH,
    PROP_HEIGHT,
    PROP_THETA,
    PROP_TAU,
    N_PROPERTIES
};

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 (UfoFilterSource *filter, guint **dim_sizes, GError **error)
{
    UfoFilterLaminoRampPrivate *priv = UFO_FILTER_LAMINO_RAMP_GET_PRIVATE(filter);
    UfoResourceManager *manager = ufo_filter_get_resource_manager(UFO_FILTER(filter));
    GError *tmp_error = NULL;

    priv->kernel_cf = ufo_resource_manager_get_kernel(manager, "lamino_ramp.cl", "lamino_ramp_create_filter", &tmp_error);

    if (tmp_error != NULL) {
        g_propagate_error (error, tmp_error);
        return;
    }

    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 gboolean
ufo_filter_lamino_ramp_generate (UfoFilterSource *filter, UfoBuffer *output[], GError **error)
{
    g_return_val_if_fail (UFO_IS_FILTER (filter), FALSE);
    UfoFilterLaminoRampPrivate *priv = UFO_FILTER_LAMINO_RAMP_GET_PRIVATE (filter);
    cl_command_queue cmd_queue = ufo_filter_get_command_queue (UFO_FILTER (filter));

    if (priv->done)
        return FALSE;

    guint width  = priv->width;
    guint fwidth = priv->fill_width;
    guint height = priv->height;
    float theta = priv->theta;
    float tau   = priv->tau;
    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], cmd_queue);

    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));

    CHECK_OPENCL_ERROR(clEnqueueNDRangeKernel(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)
{
    UfoFilterLaminoRamp *self = UFO_FILTER_LAMINO_RAMP(object);

    switch (property_id) {
        case PROP_WIDTH:
            self->priv->width = g_value_get_uint(value);
            break;
        case PROP_FILL_WIDTH:
            self->priv->fill_width = g_value_get_uint(value);
            break;
        case PROP_HEIGHT:
            self->priv->height = g_value_get_uint(value);
            break;
        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);
            break;
        default:
            G_OBJECT_WARN_INVALID_PROPERTY_ID(object, property_id, pspec);
            break;
    }
}

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_FILL_WIDTH:
            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);
            break;
        case PROP_TAU:
            g_value_set_double(value, self->priv->tau);
            break;
        default:
            G_OBJECT_WARN_INVALID_PROPERTY_ID(object, property_id, pspec);
            break;
    }
}

static void
ufo_filter_lamino_ramp_class_init(UfoFilterLaminoRampClass *klass)
{
    GObjectClass *gobject_class = G_OBJECT_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->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);

    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] =
        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);

    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]);
    g_object_class_install_property(gobject_class, PROP_HEIGHT, lamino_ramp_properties[PROP_HEIGHT]);
    g_object_class_install_property(gobject_class, PROP_THETA, lamino_ramp_properties[PROP_THETA]);
    g_object_class_install_property(gobject_class, PROP_TAU, lamino_ramp_properties[PROP_TAU]);

    g_type_class_add_private(gobject_class, sizeof(UfoFilterLaminoRampPrivate));
}

static void ufo_filter_lamino_ramp_init(UfoFilterLaminoRamp *self)
{
    UfoFilterLaminoRampPrivate *priv = self->priv = UFO_FILTER_LAMINO_RAMP_GET_PRIVATE(self);
    UfoOutputParameter output_params[] = {{2}};

    priv->width = 4;
    priv->fill_width=2;
    priv->height = 1;
    priv->theta = 0.0;
    priv->tau = 10.0;
    priv->done = FALSE;

    ufo_filter_register_outputs (UFO_FILTER (self), 1, output_params);
}

G_MODULE_EXPORT UfoFilter *
ufo_filter_plugin_new(void)
{
    return g_object_new(UFO_TYPE_FILTER_LAMINO_RAMP, NULL);
}