SOCL: fix use of deprecated API

 * use task->opencl_funcs instead of task->opencl_func
 * use task->handles instead of task->buffers
 * use codelet->modes instead of task->buffers[..]->mode

socl/src/cl_enqueuecopybuffer.c

@@ -63,6 +63,7 @@ static struct starpu_codelet codelet_copybuffer = {
    .model = NULL,
    .cpu_funcs = { &soclEnqueueCopyBuffer_cpu_task, NULL },
    .opencl_funcs = { &soclEnqueueCopyBuffer_opencl_task, NULL },
+   .modes = {STARPU_R, STARPU_RW},
    .nbuffers = 2
 };
 
@@ -79,10 +80,8 @@ cl_int command_copy_buffer_submit(command_copy_buffer cmd) {
 
 	task = task_create(CL_COMMAND_COPY_BUFFER);
 
-	task->buffers[0].handle = src_buffer->handle;
-	task->buffers[0].mode = STARPU_R;
-	task->buffers[1].handle = dst_buffer->handle;
-	task->buffers[1].mode = STARPU_RW;
+	task->handles[0] = src_buffer->handle;
+	task->handles[1] = dst_buffer->handle;
 	task->cl = &codelet_copybuffer;
 
 	arg = (struct arg_copybuffer*)malloc(sizeof(struct arg_copybuffer));

socl/src/cl_enqueuendrangekernel.c

@@ -131,7 +131,7 @@ cl_int command_ndrange_kernel_submit(command_ndrange_kernel cmd) {
 			cl_mem buf = *(cl_mem*)cmd->args[i];
 
 			gc_entity_store(&cmd->buffers[cmd->num_buffers], buf);
-			task->buffers[cmd->num_buffers].handle = buf->handle;
+			task->handles[cmd->num_buffers] = buf->handle;
 
 			/* Determine best StarPU buffer access mode */
 			int mode;
@@ -149,7 +149,7 @@ cl_int command_ndrange_kernel_submit(command_ndrange_kernel cmd) {
 				mode = STARPU_RW;
 				buf->scratch = 0;
 			}
-			task->buffers[cmd->num_buffers].mode = mode; 
+			codelet->modes[cmd->num_buffers] = mode; 
 
 			cmd->num_buffers += 1;
 		}

socl/src/cl_enqueuereadbuffer.c

@@ -63,6 +63,7 @@ static struct starpu_codelet codelet_readbuffer = {
    .model = NULL,
    .cpu_funcs = { &soclEnqueueReadBuffer_cpu_task, NULL },
    .opencl_funcs = { &soclEnqueueReadBuffer_opencl_task, NULL },
+   .modes = {STARPU_R},
    .nbuffers = 1
 };
 
@@ -78,8 +79,7 @@ cl_int command_read_buffer_submit(command_read_buffer cmd) {
 
 	task = task_create(CL_COMMAND_READ_BUFFER);
 
-	task->buffers[0].handle = buffer->handle;
-	task->buffers[0].mode = STARPU_R;
+	task->handles[0] = buffer->handle;
 	task->cl = &codelet_readbuffer;
 
 	arg = (struct arg_readbuffer*)malloc(sizeof(struct arg_readbuffer));

socl/src/cl_enqueuewritebuffer.c

@@ -65,6 +65,16 @@ static struct starpu_codelet codelet_writebuffer = {
    .model = NULL,
    .cpu_funcs = { &soclEnqueueWriteBuffer_cpu_task, NULL },
    .opencl_funcs = { &soclEnqueueWriteBuffer_opencl_task, NULL },
+   .modes = {STARPU_W},
+   .nbuffers = 1
+};
+
+static struct starpu_codelet codelet_writebuffer_partial = {
+   .where = STARPU_OPENCL,
+   .model = NULL,
+   .cpu_funcs = { &soclEnqueueWriteBuffer_cpu_task, NULL },
+   .opencl_funcs = { &soclEnqueueWriteBuffer_opencl_task, NULL },
+   .modes = {STARPU_RW},
    .nbuffers = 1
 };
 
@@ -80,13 +90,12 @@ cl_int command_write_buffer_submit(command_write_buffer cmd) {
 
 	task = task_create(CL_COMMAND_WRITE_BUFFER);
 
-	task->buffers[0].handle = buffer->handle;
+	task->handles[0] = buffer->handle;
 	//If only a subpart of the buffer is written, RW access mode is required
 	if (cb != buffer->size)
-		task->buffers[0].mode = STARPU_RW;
+		task->cl = &codelet_writebuffer_partial;
 	else 
-		task->buffers[0].mode = STARPU_W;
-	task->cl = &codelet_writebuffer;
+		task->cl = &codelet_writebuffer;
 
 	arg = (struct arg_writebuffer*)malloc(sizeof(struct arg_writebuffer));
 	arg->offset = offset;

socl/src/command.c

@@ -111,7 +111,8 @@ command_ndrange_kernel command_ndrange_kernel_create (
 	struct starpu_codelet * codelet = cmd->codelet;
 	codelet->where = STARPU_OPENCL;
 	codelet->power_model = NULL;
-	codelet->opencl_func = &soclEnqueueNDRangeKernel_task;
+	codelet->opencl_funcs[0] = &soclEnqueueNDRangeKernel_task;
+	codelet->opencl_funcs[1] = NULL;
 	codelet->model = NULL;
 
    	/* Kernel is mutable, so we duplicate its parameters... */