Use streamlined API

src/ufo-lamino-bp-task.c

@@ -26,6 +26,7 @@ struct _UfoLaminoBpTaskPrivate {
     gint            proj_idx;
     CLParameters    params;
     gboolean        cleaned;
+    gboolean        produced;
 };
 
 static void ufo_task_interface_init (UfoTaskIface *iface);
@@ -111,7 +112,7 @@ ufo_lamino_bp_task_get_structure (UfoTask *task,
                                   UfoInputParam **input_params,
                                   UfoTaskMode *mode)
 {
-    *mode = UFO_TASK_MODE_REDUCE;
+    *mode = UFO_TASK_MODE_REDUCTOR;
     *n_inputs = 1;
     *input_params = g_new0 (UfoInputParam, 1);
     (*input_params)[0].n_dims = 2;
@@ -121,10 +122,10 @@ static gboolean
 ufo_lamino_bp_task_process (UfoGpuTask *task,
                             UfoBuffer **inputs,
                             UfoBuffer *output,
-                            UfoRequisition *requisition,
-                            UfoGpuNode *node)
+                            UfoRequisition *requisition)
 {
     UfoLaminoBpTaskPrivate *priv;
+    UfoGpuNode *node;
     cl_command_queue cmd_queue;
     cl_mem in_mem;
     cl_mem out_mem;
@@ -153,6 +154,7 @@ ufo_lamino_bp_task_process (UfoGpuTask *task,
     priv->params.mat_5 =  cg * ct;
 
     // send parameters to GPU
+    node = UFO_GPU_NODE (ufo_task_node_get_proc_node (UFO_TASK_NODE (task)));
     cmd_queue = ufo_gpu_node_get_cmd_queue (node);
 
     UFO_RESOURCES_CHECK_CLERR (clEnqueueWriteBuffer (cmd_queue,
@@ -189,17 +191,22 @@ ufo_lamino_bp_task_process (UfoGpuTask *task,
     return TRUE;
 }
 
-static void
-ufo_lamino_bp_task_reduce (UfoGpuTask *task,
-                           UfoBuffer *output,
-                           UfoRequisition *requisition,
-                           UfoGpuNode *node)
+static gboolean
+ufo_lamino_bp_task_generate (UfoGpuTask *task,
+                             UfoBuffer *output,
+                             UfoRequisition *requisition)
 {
     UfoLaminoBpTaskPrivate *priv;
+    UfoGpuNode *node;
     cl_command_queue cmd_queue;
     cl_mem out_mem;
 
     priv = UFO_LAMINO_BP_TASK_GET_PRIVATE (task);
+
+    if (priv->produced)
+        return FALSE;
+
+    node = UFO_GPU_NODE (ufo_task_node_get_proc_node (UFO_TASK_NODE (task)));
     cmd_queue = ufo_gpu_node_get_cmd_queue (node);
     out_mem = ufo_buffer_get_device_array (output, cmd_queue);
 
@@ -211,6 +218,8 @@ ufo_lamino_bp_task_reduce (UfoGpuTask *task,
     UFO_RESOURCES_CHECK_CLERR (clEnqueueNDRangeKernel (cmd_queue, priv->norm_kernel,
                                                        3, NULL, requisition->dims, NULL,
                                                        0, NULL, NULL));
+    priv->produced = TRUE;
+    return TRUE;
 }
 
 static UfoNode *
@@ -355,7 +364,7 @@ static void
 ufo_gpu_task_interface_init (UfoGpuTaskIface *iface)
 {
     iface->process = ufo_lamino_bp_task_process;
-    iface->reduce = ufo_lamino_bp_task_reduce;
+    iface->generate = ufo_lamino_bp_task_generate;
 }
 
 static void
@@ -464,4 +473,5 @@ ufo_lamino_bp_task_init(UfoLaminoBpTask *self)
     self->priv = priv = UFO_LAMINO_BP_TASK_GET_PRIVATE(self);
     priv->param_mem = NULL;
     priv->cleaned = FALSE;
+    priv->produced = FALSE;
 }

src/ufo-lamino-conv-task.c

@@ -44,16 +44,17 @@ static gboolean
 ufo_lamino_conv_task_process (UfoGpuTask *task,
                               UfoBuffer **inputs,
                               UfoBuffer *output,
-                              UfoRequisition *requisition,
-                              UfoGpuNode *node)
+                              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);
@@ -103,7 +104,7 @@ ufo_lamino_conv_task_get_structure (UfoTask *task,
     UfoLaminoConvTaskPrivate *priv;
 
     priv = UFO_LAMINO_CONV_TASK_GET_PRIVATE (task);
-    *mode = UFO_TASK_MODE_SINGLE;
+    *mode = UFO_TASK_MODE_PROCESSOR;
     *n_inputs = 2;
     *in_params = g_new0 (UfoInputParam, 2);
     (*in_params)[0].n_dims = 2;

src/ufo-lamino-ramp-task.c

@@ -120,17 +120,16 @@ ufo_lamino_ramp_task_get_structure (UfoTask *task,
                                     UfoTaskMode *mode)
 {
     *n_inputs = 0;
-    *mode = UFO_TASK_MODE_SINGLE;
+    *mode = UFO_TASK_MODE_GENERATOR;
 }
 
 static gboolean
-ufo_lamino_ramp_task_process (UfoGpuTask *task,
-                              UfoBuffer **inputs,
-                              UfoBuffer *output,
-                              UfoRequisition *requisition,
-                              UfoGpuNode *node)
+ufo_lamino_ramp_task_generate (UfoGpuTask *task,
+                               UfoBuffer *output,
+                               UfoRequisition *requisition)
 {
     UfoLaminoRampTaskPrivate *priv;
+    UfoGpuNode *node;
     cl_command_queue cmd_queue;
     cl_mem out_mem;
 
@@ -139,6 +138,7 @@ ufo_lamino_ramp_task_process (UfoGpuTask *task,
     if (priv->done)
         return FALSE;
 
+    node = UFO_GPU_NODE (ufo_task_node_get_proc_node (UFO_TASK_NODE (task)));
     cmd_queue = ufo_gpu_node_get_cmd_queue (node);
     out_mem = ufo_buffer_get_device_array (output, cmd_queue);
 
@@ -241,7 +241,7 @@ ufo_task_interface_init (UfoTaskIface *iface)
 static void
 ufo_gpu_task_interface_init (UfoGpuTaskIface *iface)
 {
-    iface->process = ufo_lamino_ramp_task_process;
+    iface->generate = ufo_lamino_ramp_task_generate;
 }
 
 static void

src/ufo-padding-2d-task.c

@@ -81,10 +81,10 @@ static gboolean
 ufo_padding_2d_task_process (UfoGpuTask *task,
                              UfoBuffer **inputs,
                              UfoBuffer *output,
-                             UfoRequisition *requisition,
-                             UfoGpuNode *node)
+                             UfoRequisition *requisition)
 {
     UfoPadding2DTaskPrivate *priv;
+    UfoGpuNode *node;
     cl_command_queue cmd_queue;
     cl_mem in_mem;
     cl_mem out_mem;
@@ -111,6 +111,7 @@ ufo_padding_2d_task_process (UfoGpuTask *task,
         pval = sum / (gfloat) psz;
     }
 
+    node = UFO_GPU_NODE (ufo_task_node_get_proc_node (UFO_TASK_NODE (task)));
     cmd_queue = ufo_gpu_node_get_cmd_queue (node);
     in_mem = ufo_buffer_get_device_array (inputs[0], cmd_queue);
     out_mem = ufo_buffer_get_device_array (output, cmd_queue);
@@ -207,7 +208,7 @@ ufo_padding_2d_task_get_structure (UfoTask *task,
     UfoPadding2DTaskPrivate *priv;
 
     priv = UFO_PADDING_2D_TASK_GET_PRIVATE (task);
-    *mode = UFO_TASK_MODE_SINGLE;
+    *mode = UFO_TASK_MODE_PROCESSOR;
     *n_inputs = 1;
     *in_params = g_new0 (UfoInputParam, 1);
     (*in_params)[0].n_dims = 2;

src/ufo-scale-task.c

@@ -49,17 +49,18 @@ ufo_scale_task_new (void)
 
 static gboolean
 ufo_scale_task_process (UfoGpuTask *task,
-                         UfoBuffer **inputs,
-                         UfoBuffer *output,
-                         UfoRequisition *requisition,
-                         UfoGpuNode *node)
+                        UfoBuffer **inputs,
+                        UfoBuffer *output,
+                        UfoRequisition *requisition)
 {
     UfoScaleTaskPrivate *priv;
+    UfoGpuNode *node;
     cl_command_queue cmd_queue;
     cl_mem in_mem;
     cl_mem out_mem;
 
     priv = UFO_SCALE_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);
     in_mem = ufo_buffer_get_device_array (inputs[0], cmd_queue);
     out_mem = ufo_buffer_get_device_array (output, cmd_queue);
@@ -85,7 +86,7 @@ ufo_scale_task_get_structure (UfoTask *task,
     UfoScaleTaskPrivate *priv;
 
     priv = UFO_SCALE_TASK_GET_PRIVATE (task);
-    *mode = UFO_TASK_MODE_SINGLE;
+    *mode = UFO_TASK_MODE_PROCESSOR;
     *n_inputs = 1;
     *in_params = g_new0 (UfoInputParam, 1);
     (*in_params)[0].n_dims = 2;