add a example on how to define a new data interface

doc/chapters/advanced-api.texi

@@ -146,8 +146,132 @@ future CRC computation. This is used for computing data size footprint.
 Returns the next available id for a newly created data interface.
 @end deftypefun
 
-TODO
-See @code{src/datawizard/interfaces/vector_interface.c} for now.
+Let's define a new data interface to manage complex numbers.
+
+@cartouche
+@smallexample
+/* interface for complex numbers */
+struct starpu_complex_interface
+@{
+        double *real;
+        double *imaginary;
+        int nx;
+@};
+@end smallexample
+@end cartouche
+
+Registering such a data to StarPU is easily done using the function
+@code{starpu_data_register} (@pxref{Basic Data Library API}). The last
+parameter of the function, @code{interface_complex_ops}, will be
+described below.
+
+@cartouche
+@smallexample
+void starpu_complex_data_register(starpu_data_handle_t *handle,
+     uint32_t home_node, double *real, double *imaginary, int nx)
+@{
+        struct starpu_complex_interface complex =
+        @{
+                .real = real,
+                .imaginary = imaginary,
+                .nx = nx
+        @};
+
+        if (interface_complex_ops.interfaceid == -1)
+        @{
+                interface_complex_ops.interfaceid = starpu_data_interface_get_next_id();
+        @}
+
+        starpu_data_register(handleptr, home_node, &complex, &interface_complex_ops);
+@}
+@end smallexample
+@end cartouche
+
+Different operations need to be defined for a data interface through
+the type @code{struct starpu_data_interface_ops} (@pxref{Data
+Interface API}). We only define here the basic operations needed to
+run simple applications. The source code for the different functions
+can be found in the file
+@code{examples/interface/complex_interface.c}.
+
+@cartouche
+@smallexample
+static struct starpu_data_interface_ops interface_complex_ops =
+@{
+        .register_data_handle = complex_register_data_handle,
+        .allocate_data_on_node = complex_allocate_data_on_node,
+        .copy_methods = &complex_copy_methods,
+        .get_size = complex_get_size,
+        .footprint = complex_footprint,
+        .interfaceid = -1,
+@};
+@end smallexample
+@end cartouche
+
+Functions need to be defined to access the different fields of the
+complex interface from a StarPU data handle.
+
+@cartouche
+@smallexample
+double *starpu_complex_get_real(starpu_data_handle_t handle)
+@{
+        struct starpu_complex_interface *complex_interface =
+          (struct starpu_complex_interface *) starpu_data_get_interface_on_node(handle, 0);
+        return complex_interface->real;
+@}
+
+double *starpu_complex_get_imaginary(starpu_data_handle_t handle);
+int starpu_complex_get_nx(starpu_data_handle_t handle);
+@end smallexample
+@end cartouche
+
+Similar functions need to be defined to access the different fields of the
+complex interface from a @code{void *} pointer to be used within codelet
+implemetations.
+
+@cartouche
+@smallexample
+#define STARPU_COMPLEX_GET_REAL(interface)	\
+        (((struct starpu_complex_interface *)(interface))->real)
+#define STARPU_COMPLEX_GET_IMAGINARY(interface)	\
+        (((struct starpu_complex_interface *)(interface))->imaginary)
+#define STARPU_COMPLEX_GET_NX(interface)	\
+        (((struct starpu_complex_interface *)(interface))->nx)
+@end smallexample
+@end cartouche
+
+Complex data interfaces can then be registered to StarPU.
+
+@cartouche
+@smallexample
+double real = 45.0;
+double imaginary = 12.0;
+starpu_complex_data_register(&handle1, 0, &real, &imaginary, 1);
+starpu_insert_task(&cl_display, STARPU_R, handle1, 0);
+@end smallexample
+@end cartouche
+
+and used by codelets.
+
+@cartouche
+@smallexample
+void display_complex_codelet(void *descr[], __attribute__ ((unused)) void *_args)
+@{
+        int nx = STARPU_COMPLEX_GET_NX(descr[0]);
+        double *real = STARPU_COMPLEX_GET_REAL(descr[0]);
+        double *imaginary = STARPU_COMPLEX_GET_IMAGINARY(descr[0]);
+        int i;
+
+        for(i=0 ; i<nx ; i++)
+        @{
+                fprintf(stderr, "Complex[%d] = %3.2f + %3.2f i\n", i, real[i], imaginary[i]);
+        @}
+@}
+@end smallexample
+@end cartouche
+
+The whole code for this complex data interface is available in the
+directory @code{examples/interface/}.
 
 @node Multiformat Data Interface
 @section Multiformat Data Interface

examples/Makefile.am

@@ -128,7 +128,8 @@ noinst_HEADERS = 				\
 	spmv/dw_block_spmv.h                    \
 	basic_examples/multiformat_types.h      \
 	filters/custom_mf/custom_interface.h    \
-	filters/custom_mf/custom_types.h
+	filters/custom_mf/custom_types.h	\
+	interface/complex_interface.h
 
 #####################################
 # What to install and what to check #
@@ -171,6 +172,7 @@ examplebin_PROGRAMS +=				\
 	spmv/spmv				\
 	callback/callback			\
 	incrementer/incrementer			\
+	interface/complex			\
 	matvecmult/matvecmult			\
 	profiling/profiling			\
 	scheduler/dummy_sched			\
@@ -233,6 +235,7 @@ STARPU_EXAMPLES +=				\
 	spmv/spmv				\
 	callback/callback			\
 	incrementer/incrementer			\
+	interface/complex			\
 	matvecmult/matvecmult			\
 	profiling/profiling			\
 	scheduler/dummy_sched			\
@@ -692,6 +695,18 @@ nobase_STARPU_OPENCL_DATA_DATA += \
 	incrementer/incrementer_kernels_opencl_kernel.cl
 endif
 
+#####################
+# interface example #
+#####################
+
+interface_complex_SOURCES	=	\
+	interface/complex.c		\
+	interface/complex_interface.c
+if STARPU_USE_CUDA
+interface_complex_SOURCES	+=	\
+	interface/complex_kernels.cu
+endif
+
 ######################
 # matVecMult example #
 ######################

examples/interface/complex.c

@@ -0,0 +1,152 @@
+/* StarPU --- Runtime system for heterogeneous multicore architectures.
+ *
+ * Copyright (C) 2012  Centre National de la Recherche Scientifique
+ *
+ * StarPU is free software; you can redistribute it and/or modify
+ * it under the terms of the GNU Lesser General Public License as published by
+ * the Free Software Foundation; either version 2.1 of the License, or (at
+ * your option) any later version.
+ *
+ * StarPU is distributed in the hope that it will be useful, but
+ * WITHOUT ANY WARRANTY; without even the implied warranty of
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.
+ *
+ * See the GNU Lesser General Public License in COPYING.LGPL for more details.
+ */
+
+#include <starpu.h>
+#include "complex_interface.h"
+
+#ifdef STARPU_USE_CUDA
+extern void copy_complex_codelet_cuda(void *descr[], __attribute__ ((unused)) void *_args);
+#endif
+
+void compare_complex_codelet(void *descr[], __attribute__ ((unused)) void *_args)
+{
+	int nx1 = STARPU_COMPLEX_GET_NX(descr[0]);
+	double *real1 = STARPU_COMPLEX_GET_REAL(descr[0]);
+	double *imaginary1 = STARPU_COMPLEX_GET_IMAGINARY(descr[0]);
+
+	int nx2 = STARPU_COMPLEX_GET_NX(descr[1]);
+	double *real2 = STARPU_COMPLEX_GET_REAL(descr[1]);
+	double *imaginary2 = STARPU_COMPLEX_GET_IMAGINARY(descr[1]);
+
+	int compare = (nx1 == nx2);
+	if (nx1 == nx2)
+	{
+		int i;
+		for(i=0 ; i<nx1 ; i++)
+		{
+			if (real1[i] != real2[i] || imaginary1[i] != imaginary2[i])
+			{
+				compare = 0;
+				break;
+			}
+		}
+	}
+	fprintf(stderr, "Complex numbers are%s similar\n", compare==0 ? " NOT" : "");
+}
+
+void display_complex_codelet(void *descr[], __attribute__ ((unused)) void *_args)
+{
+	int nx = STARPU_COMPLEX_GET_NX(descr[0]);
+	double *real = STARPU_COMPLEX_GET_REAL(descr[0]);
+	double *imaginary = STARPU_COMPLEX_GET_IMAGINARY(descr[0]);
+	int i;
+
+	for(i=0 ; i<nx ; i++)
+	{
+		fprintf(stderr, "Complex[%d] = %3.2f + %3.2f i\n", i, real[i], imaginary[i]);
+	}
+}
+
+struct starpu_codelet cl_display =
+{
+	.cpu_funcs = {display_complex_codelet, NULL},
+	.nbuffers = 1,
+	.modes = {STARPU_R}
+};
+
+struct starpu_codelet cl_copy =
+{
+#ifdef STARPU_USE_CUDA
+	.cuda_funcs = {copy_complex_codelet_cuda, NULL},
+#endif
+	.nbuffers = 2,
+	.modes = {STARPU_R, STARPU_W}
+};
+
+struct starpu_codelet cl_compare =
+{
+	.cpu_funcs = {compare_complex_codelet, NULL},
+	.nbuffers = 2,
+	.modes = {STARPU_R, STARPU_R}
+};
+
+int main(int argc, char **argv)
+{
+	int ret = 0;
+	starpu_data_handle_t handle1;
+	starpu_data_handle_t handle2;
+
+	double real = 45.0;
+	double imaginary = 12.0;
+	double copy_real = 78.0;
+	double copy_imaginary = 78.0;
+
+	ret = starpu_init(NULL);
+	if (ret == -ENODEV) return 77;
+	STARPU_CHECK_RETURN_VALUE(ret, "starpu_init");
+
+	starpu_complex_data_register(&handle1, 0, &real, &imaginary, 1);
+	starpu_complex_data_register(&handle2, 0, &copy_real, &copy_imaginary, 1);
+
+	ret = starpu_insert_task(&cl_display, STARPU_R, handle1, 0);
+	if (ret == -ENODEV) goto enodev;
+	STARPU_CHECK_RETURN_VALUE(ret, "starpu_insert_task");
+
+	ret = starpu_insert_task(&cl_display, STARPU_R, handle2, 0);
+	if (ret == -ENODEV) goto enodev;
+	STARPU_CHECK_RETURN_VALUE(ret, "starpu_insert_task");
+
+	ret = starpu_insert_task(&cl_compare,
+				 STARPU_R, handle1,
+				 STARPU_R, handle2,
+				 0);
+	if (ret == -ENODEV) goto enodev;
+	STARPU_CHECK_RETURN_VALUE(ret, "starpu_insert_task");
+
+	ret = starpu_insert_task(&cl_copy,
+				 STARPU_R, handle1,
+				 STARPU_W, handle2,
+				 0);
+	if (ret == -ENODEV) goto enodev;
+	STARPU_CHECK_RETURN_VALUE(ret, "starpu_insert_task");
+
+	ret = starpu_insert_task(&cl_display, STARPU_R, handle1, 0);
+	if (ret == -ENODEV) goto enodev;
+	STARPU_CHECK_RETURN_VALUE(ret, "starpu_insert_task");
+
+	ret = starpu_insert_task(&cl_display, STARPU_R, handle2, 0);
+	if (ret == -ENODEV) goto enodev;
+	STARPU_CHECK_RETURN_VALUE(ret, "starpu_insert_task");
+
+
+	ret = starpu_insert_task(&cl_compare,
+				 STARPU_R, handle1,
+				 STARPU_R, handle2,
+				 0);
+	if (ret == -ENODEV) goto enodev;
+	STARPU_CHECK_RETURN_VALUE(ret, "starpu_insert_task");
+
+	starpu_task_wait_for_all();
+
+	starpu_shutdown();
+	return 0;
+
+enodev:
+	starpu_data_unregister(handle1);
+	starpu_data_unregister(handle2);
+	starpu_shutdown();
+	return 77;
+}

examples/interface/complex_interface.c

@@ -0,0 +1,207 @@
+/* StarPU --- Runtime system for heterogeneous multicore architectures.
+ *
+ * Copyright (C) 2012  Centre National de la Recherche Scientifique
+ *
+ * StarPU is free software; you can redistribute it and/or modify
+ * it under the terms of the GNU Lesser General Public License as published by
+ * the Free Software Foundation; either version 2.1 of the License, or (at
+ * your option) any later version.
+ *
+ * StarPU is distributed in the hope that it will be useful, but
+ * WITHOUT ANY WARRANTY; without even the implied warranty of
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.
+ *
+ * See the GNU Lesser General Public License in COPYING.LGPL for more details.
+ */
+
+#include <starpu.h>
+#include <starpu_cuda.h>
+#include <starpu_hash.h>
+
+#include "complex_interface.h"
+
+double *starpu_complex_get_real(starpu_data_handle_t handle)
+{
+	struct starpu_complex_interface *complex_interface =
+		(struct starpu_complex_interface *) starpu_data_get_interface_on_node(handle, 0);
+
+	return complex_interface->real;
+}
+
+double *starpu_complex_get_imaginary(starpu_data_handle_t handle)
+{
+	struct starpu_complex_interface *complex_interface =
+		(struct starpu_complex_interface *) starpu_data_get_interface_on_node(handle, 0);
+
+	return complex_interface->imaginary;
+}
+
+int starpu_complex_get_nx(starpu_data_handle_t handle)
+{
+	struct starpu_complex_interface *complex_interface =
+		(struct starpu_complex_interface *) starpu_data_get_interface_on_node(handle, 0);
+
+	return complex_interface->nx;
+}
+
+static void complex_register_data_handle(starpu_data_handle_t handle, uint32_t home_node, void *data_interface)
+{
+	struct starpu_complex_interface *complex_interface = (struct starpu_complex_interface *) data_interface;
+
+	unsigned node;
+	for (node = 0; node < STARPU_MAXNODES; node++)
+	{
+		struct starpu_complex_interface *local_interface = (struct starpu_complex_interface *)
+			starpu_data_get_interface_on_node(handle, node);
+
+		local_interface->real = complex_interface->real;
+		local_interface->imaginary = complex_interface->imaginary;
+		local_interface->nx = complex_interface->nx;
+	}
+}
+
+static starpu_ssize_t complex_allocate_data_on_node(void *data_interface, uint32_t node)
+{
+	struct starpu_complex_interface *complex_interface = (struct starpu_complex_interface *) data_interface;
+
+	unsigned fail = 0;
+	double *addr_real = 0;
+	double *addr_imaginary = 0;
+	ssize_t requested_memory = complex_interface->nx * sizeof(complex_interface->real[0]);
+
+	enum starpu_node_kind kind = starpu_node_get_kind(node);
+
+	switch(kind)
+	{
+		case STARPU_CPU_RAM:
+			addr_real = malloc(requested_memory);
+			addr_imaginary = malloc(requested_memory);
+			if (!addr_real || !addr_imaginary)
+				fail = 1;
+			break;
+#ifdef STARPU_USE_CUDA
+		case STARPU_CUDA_RAM:
+		{
+			cudaError_t status;
+			status = cudaMalloc((void **)&addr_real, requested_memory);
+			if (!addr_real || (status != cudaSuccess))
+			{
+				if (STARPU_UNLIKELY(status != cudaErrorMemoryAllocation))
+					STARPU_CUDA_REPORT_ERROR(status);
+
+				fail = 1;
+			}
+			else
+			{
+				status = cudaMalloc((void **)&addr_imaginary, requested_memory);
+				if (!addr_imaginary || (status != cudaSuccess))
+				{
+					if (STARPU_UNLIKELY(status != cudaErrorMemoryAllocation))
+						STARPU_CUDA_REPORT_ERROR(status);
+
+					fail = 1;
+				}
+			}
+
+			break;
+		}
+#endif
+#ifdef STARPU_USE_OPENCL
+	        case STARPU_OPENCL_RAM:
+		{
+			STARPU_ASSERT(0);
+		}
+#endif
+		default:
+			STARPU_ASSERT(0);
+	}
+
+	if (fail)
+		return -ENOMEM;
+
+	/* update the data properly in consequence */
+	complex_interface->real = addr_real;
+	complex_interface->imaginary = addr_imaginary;
+
+	return 2*requested_memory;
+}
+
+static size_t complex_get_size(starpu_data_handle_t handle)
+{
+	size_t size;
+	struct starpu_complex_interface *complex_interface = (struct starpu_complex_interface *) starpu_data_get_interface_on_node(handle, 0);
+
+	size = complex_interface->nx * 2 * sizeof(double);
+	return size;
+}
+
+static uint32_t complex_footprint(starpu_data_handle_t handle)
+{
+	return starpu_crc32_be(starpu_complex_get_nx(handle), 0);
+}
+
+#ifdef STARPU_USE_CUDA
+static int copy_cuda_common(void *src_interface, unsigned src_node, void *dst_interface, unsigned dst_node, enum cudaMemcpyKind kind)
+{
+	struct starpu_complex_interface *src_complex = src_interface;
+	struct starpu_complex_interface *dst_complex = dst_interface;
+
+	cudaError_t cures;
+
+	cures = cudaMemcpy((void *)dst_complex->real, (void *)src_complex->real, src_complex->nx*sizeof(src_complex->real[0]), kind);
+	if (STARPU_UNLIKELY(cures))
+		STARPU_CUDA_REPORT_ERROR(cures);
+
+	cures = cudaMemcpy((char *)dst_complex->imaginary, (char *)src_complex->imaginary, src_complex->nx*sizeof(src_complex->imaginary[0]), kind);
+	if (STARPU_UNLIKELY(cures))
+		STARPU_CUDA_REPORT_ERROR(cures);
+
+	return 0;
+}
+
+static int copy_ram_to_cuda(void *src_interface, unsigned src_node, void *dst_interface, unsigned dst_node)
+{
+	return copy_cuda_common(src_interface, src_node, dst_interface, dst_node, cudaMemcpyHostToDevice);
+}
+
+static int copy_cuda_to_ram(void *src_interface, unsigned src_node, void *dst_interface, unsigned dst_node)
+{
+	return copy_cuda_common(src_interface, src_node, dst_interface, dst_node, cudaMemcpyDeviceToHost);
+}
+#endif
+
+
+static const struct starpu_data_copy_methods complex_copy_methods =
+{
+#ifdef STARPU_USE_CUDA
+	.ram_to_cuda = copy_ram_to_cuda,
+	.cuda_to_ram = copy_cuda_to_ram,
+#endif
+};
+
+static struct starpu_data_interface_ops interface_complex_ops =
+{
+	.register_data_handle = complex_register_data_handle,
+	.allocate_data_on_node = complex_allocate_data_on_node,
+	.copy_methods = &complex_copy_methods,
+	.get_size = complex_get_size,
+	.footprint = complex_footprint,
+	.interfaceid = -1,
+};
+
+void starpu_complex_data_register(starpu_data_handle_t *handleptr, uint32_t home_node, double *real, double *imaginary, int nx)
+{
+	struct starpu_complex_interface complex =
+	{
+		.real = real,
+		.imaginary = imaginary,
+		.nx = nx
+	};
+
+	if (interface_complex_ops.interfaceid == -1)
+	{
+		interface_complex_ops.interfaceid = starpu_data_interface_get_next_id();
+	}
+
+	starpu_data_register(handleptr, home_node, &complex, &interface_complex_ops);
+}

examples/interface/complex_interface.h

@@ -0,0 +1,36 @@
+/* StarPU --- Runtime system for heterogeneous multicore architectures.
+ *
+ * Copyright (C) 2012  Centre National de la Recherche Scientifique
+ *
+ * StarPU is free software; you can redistribute it and/or modify
+ * it under the terms of the GNU Lesser General Public License as published by
+ * the Free Software Foundation; either version 2.1 of the License, or (at
+ * your option) any later version.
+ *
+ * StarPU is distributed in the hope that it will be useful, but
+ * WITHOUT ANY WARRANTY; without even the implied warranty of
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.
+ *
+ * See the GNU Lesser General Public License in COPYING.LGPL for more details.
+ */
+
+#include <starpu.h>
+
+/* interface for complex numbers */
+struct starpu_complex_interface
+{
+	double *real;
+	double *imaginary;
+	int nx;
+};
+
+void starpu_complex_data_register(starpu_data_handle_t *handle, uint32_t home_node, double *real, double *imaginary, int nx);
+
+double *starpu_complex_get_real(starpu_data_handle_t handle);
+double *starpu_complex_get_imaginary(starpu_data_handle_t handle);
+int starpu_complex_get_nx(starpu_data_handle_t handle);
+
+#define STARPU_COMPLEX_GET_REAL(interface)	(((struct starpu_complex_interface *)(interface))->real)
+#define STARPU_COMPLEX_GET_IMAGINARY(interface)	(((struct starpu_complex_interface *)(interface))->imaginary)
+#define STARPU_COMPLEX_GET_NX(interface)	(((struct starpu_complex_interface *)(interface))->nx)
+

examples/interface/complex_kernels.cu

@@ -0,0 +1,50 @@
+/* StarPU --- Runtime system for heterogeneous multicore architectures.
+ *
+ * Copyright (C) 2012  Centre National de la Recherche Scientifique
+ *
+ * StarPU is free software; you can redistribute it and/or modify
+ * it under the terms of the GNU Lesser General Public License as published by
+ * the Free Software Foundation; either version 2.1 of the License, or (at
+ * your option) any later version.
+ *
+ * StarPU is distributed in the hope that it will be useful, but
+ * WITHOUT ANY WARRANTY; without even the implied warranty of
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.
+ *
+ * See the GNU Lesser General Public License in COPYING.LGPL for more details.
+ */
+
+#include <starpu.h>
+#include <starpu_cuda.h>
+#include "complex_interface.h"
+
+static __global__ void complex_copy_cuda(double *o_real, double *o_imaginary, double *i_real, double *i_imaginary, unsigned n)
+{
+        unsigned i =  blockIdx.x*blockDim.x + threadIdx.x;
+
+	if (i < n)
+	{
+		o_real[i] = i_real[i];
+		o_imaginary[i] = i_imaginary[i];
+	}
+}
+
+extern "C" void copy_complex_codelet_cuda(void *descr[], __attribute__ ((unused)) void *_args)
+{
+	(void)_args;
+
+	int nx = STARPU_COMPLEX_GET_NX(descr[0]);
+
+	double *i_real = STARPU_COMPLEX_GET_REAL(descr[0]);
+	double *i_imaginary = STARPU_COMPLEX_GET_IMAGINARY(descr[0]);
+
+	double *o_real = STARPU_COMPLEX_GET_REAL(descr[1]);
+	double *o_imaginary = STARPU_COMPLEX_GET_IMAGINARY(descr[1]);
+
+	unsigned threads_per_block = 64;
+	unsigned nblocks = (nx + threads_per_block-1) / threads_per_block;
+
+        complex_copy_cuda<<<nblocks, threads_per_block, 0, starpu_cuda_get_local_stream()>>>(o_real, o_imaginary, i_real, i_imaginary, nx);
+
+	cudaStreamSynchronize(starpu_cuda_get_local_stream());
+}