123456789101112131415161718192021222324252627282930313233343536373839404142434445464748495051525354555657585960616263646566676869707172737475767778798081828384858687888990919293949596979899100101102103104105106107108109110111112113114115116117118119120121122123124125126127128129130131132133134135136137138139140141142143144145146147148149150151152153154155156157158159160161162163164165166167168169170171172173174175176177178179180181182183184185186187188189190191192193194195196197198199200201202203204205206207208209210 |
- /**
- * SECTION:ufo-filter-task
- * @Short_description: Process arbitrary Filter kernels
- * @Title: filter
- *
- * This module is used to load an arbitrary #UfoLaminoConvTask:kernel from
- * #UfoLaminoConvTask:filename and execute it on each input. The kernel must have
- * only two global float array parameters, the first represents the input, the
- * second one the output. #UfoLaminoConvTask:num-dims must be changed, if the kernel
- * accesses either one or three dimensional index spaces.
- */
- #ifdef __APPLE__
- #include <OpenCL/cl.h>
- #else
- #include <CL/cl.h>
- #endif
- #include "ufo-lamino-conv-task.h"
- struct _UfoLaminoConvTaskPrivate {
- cl_kernel kernel;
- guint img_width;
- guint img_height;
- };
- static void ufo_task_interface_init (UfoTaskIface *iface);
- G_DEFINE_TYPE_WITH_CODE (UfoLaminoConvTask, ufo_lamino_conv_task, UFO_TYPE_TASK_NODE,
- G_IMPLEMENT_INTERFACE (UFO_TYPE_TASK,
- ufo_task_interface_init))
- #define UFO_LAMINO_CONV_TASK_GET_PRIVATE(obj) (G_TYPE_INSTANCE_GET_PRIVATE((obj), UFO_TYPE_LAMINO_CONV_TASK, UfoLaminoConvTaskPrivate))
- UfoNode *
- ufo_lamino_conv_task_new (void)
- {
- return UFO_NODE (g_object_new (UFO_TYPE_LAMINO_CONV_TASK, NULL));
- }
- static gboolean
- ufo_lamino_conv_task_process (UfoTask *task,
- UfoBuffer **inputs,
- UfoBuffer *output,
- UfoRequisition *requisition)
- {
- UfoLaminoConvTaskPrivate *priv;
- UfoGpuNode *node;
- cl_command_queue cmd_queue;
- cl_mem in1_mem;
- cl_mem in2_mem;
- cl_mem out_mem;
- priv = UFO_LAMINO_CONV_TASK (task)->priv;
- node = UFO_GPU_NODE (ufo_task_node_get_proc_node (UFO_TASK_NODE (task)));
- cmd_queue = ufo_gpu_node_get_cmd_queue (node);
- in1_mem = ufo_buffer_get_device_array (inputs[0], cmd_queue);
- in2_mem = ufo_buffer_get_device_array (inputs[1], cmd_queue);
- out_mem = ufo_buffer_get_device_array (output, cmd_queue);
- UFO_RESOURCES_CHECK_CLERR (clSetKernelArg (priv->kernel, 0, sizeof(cl_mem), (void *) &in1_mem));
- UFO_RESOURCES_CHECK_CLERR (clSetKernelArg (priv->kernel, 1, sizeof(cl_mem), (void *) &in2_mem));
- UFO_RESOURCES_CHECK_CLERR (clSetKernelArg (priv->kernel, 2, sizeof(cl_mem), (void *) &out_mem));
- UFO_RESOURCES_CHECK_CLERR (clEnqueueNDRangeKernel (cmd_queue, priv->kernel,
- 2, NULL, requisition->dims, NULL,
- 0, NULL, NULL));
- return TRUE;
- }
- static void
- ufo_lamino_conv_task_setup (UfoTask *task,
- UfoResources *resources,
- GError **error)
- {
- UfoLaminoConvTaskPrivate *priv;
- priv = UFO_LAMINO_CONV_TASK_GET_PRIVATE (task);
- priv->kernel = ufo_resources_get_kernel (resources, "lamino_ft_conv.cl", "lamino_c", error);
- if (priv->kernel) {
- UFO_RESOURCES_CHECK_CLERR (clRetainKernel (priv->kernel));
- }
- }
- static void
- ufo_lamino_conv_task_get_requisition (UfoTask *task,
- UfoBuffer **inputs,
- UfoRequisition *requisition)
- {
- ufo_buffer_get_requisition (inputs[0], requisition);
- }
- static guint
- ufo_lamino_conv_task_get_num_inputs (UfoTask *task)
- {
- return 2;
- }
- static guint
- ufo_lamino_conv_task_get_num_dimensions (UfoTask *task,
- guint input)
- {
- g_return_val_if_fail (input <= 1, 0);
- return 2;
- }
- static UfoTaskMode
- ufo_lamino_conv_task_get_mode (UfoTask *task)
- {
- return UFO_TASK_MODE_PROCESSOR | UFO_TASK_MODE_GPU;
- }
- static UfoNode *
- ufo_lamino_conv_task_copy_real (UfoNode *node,
- GError **error)
- {
- UfoLaminoConvTask *copy;
- copy = UFO_LAMINO_CONV_TASK (ufo_lamino_conv_task_new ());
- return UFO_NODE (copy);
- }
- static gboolean
- ufo_lamino_conv_task_equal_real (UfoNode *n1,
- UfoNode *n2)
- {
- g_return_val_if_fail (UFO_IS_LAMINO_CONV_TASK (n1) && UFO_IS_LAMINO_CONV_TASK (n2), FALSE);
- return TRUE;
- }
- static void
- ufo_lamino_conv_task_finalize (GObject *object)
- {
- UfoLaminoConvTaskPrivate *priv;
- priv = UFO_LAMINO_CONV_TASK_GET_PRIVATE (object);
- if (priv->kernel) {
- UFO_RESOURCES_CHECK_CLERR (clReleaseKernel (priv->kernel));
- priv->kernel = NULL;
- }
- G_OBJECT_CLASS (ufo_lamino_conv_task_parent_class)->finalize (object);
- }
- static void
- ufo_task_interface_init (UfoTaskIface *iface)
- {
- iface->setup = ufo_lamino_conv_task_setup;
- iface->get_requisition = ufo_lamino_conv_task_get_requisition;
- iface->get_num_inputs = ufo_lamino_conv_task_get_num_inputs;
- iface->get_num_dimensions = ufo_lamino_conv_task_get_num_dimensions;
- iface->get_mode = ufo_lamino_conv_task_get_mode;
- iface->process = ufo_lamino_conv_task_process;
- }
- static void
- ufo_lamino_conv_task_set_property (GObject *object,
- guint property_id,
- const GValue *value,
- GParamSpec *pspec)
- {
- switch (property_id) {
- default:
- G_OBJECT_WARN_INVALID_PROPERTY_ID (object, property_id, pspec);
- break;
- }
- }
- static void
- ufo_lamino_conv_task_get_property (GObject *object,
- guint property_id,
- GValue *value,
- GParamSpec *pspec)
- {
- switch (property_id) {
- default:
- G_OBJECT_WARN_INVALID_PROPERTY_ID(object, property_id, pspec);
- break;
- }
- }
- static void
- ufo_lamino_conv_task_class_init (UfoLaminoConvTaskClass *klass)
- {
- GObjectClass *oclass;
- UfoNodeClass *node_class;
-
- oclass = G_OBJECT_CLASS (klass);
- node_class = UFO_NODE_CLASS (klass);
- oclass->finalize = ufo_lamino_conv_task_finalize;
- oclass->set_property = ufo_lamino_conv_task_set_property;
- oclass->get_property = ufo_lamino_conv_task_get_property;
- node_class->copy = ufo_lamino_conv_task_copy_real;
- node_class->equal = ufo_lamino_conv_task_equal_real;
- g_type_class_add_private(klass, sizeof(UfoLaminoConvTaskPrivate));
- }
- static void
- ufo_lamino_conv_task_init (UfoLaminoConvTask *self)
- {
- UfoLaminoConvTaskPrivate *priv;
- self->priv = priv = UFO_LAMINO_CONV_TASK_GET_PRIVATE (self);
- priv->kernel = NULL;
- }
|