Provide the CUDA implementation of the codelet in the mpi_like(_async) tests.

tests/Makefile.am

@@ -234,6 +234,14 @@ datawizard_mpi_like_SOURCES =			\
 datawizard_mpi_like_async_SOURCES =		\
 	datawizard/mpi_like_async.c
 
+if STARPU_USE_CUDA
+datawizard_mpi_like_SOURCES +=			\
+	datawizard/cuda_codelet_unsigned_inc.cu
+
+datawizard_mpi_like_async_SOURCES +=		\
+	datawizard/cuda_codelet_unsigned_inc.cu
+endif
+
 errorcheck_starpu_init_noworker_SOURCES =	\
 	errorcheck/starpu_init_noworker.c
 

tests/datawizard/cuda_codelet_unsigned_inc.cu

@@ -0,0 +1,31 @@
+/*
+ * StarPU
+ * Copyright (C) INRIA 2008-2010 (see AUTHORS file)
+ *
+ * This program 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.
+ *
+ * This program 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>
+
+static __global__ void _cuda_unsigned_inc(unsigned *val)
+{
+	val[0]++;
+}
+
+extern "C" void cuda_codelet_unsigned_inc(void *descr[], __attribute__ ((unused)) void *cl_arg)
+{
+	unsigned *val = (unsigned *)STARPU_GET_VARIABLE_PTR(descr[0]);
+
+	_cuda_unsigned_inc<<<1,1>>>(val);
+
+	cudaThreadSynchronize();
+}

tests/datawizard/mpi_like.c

@@ -42,6 +42,10 @@ static struct thread_data problem_data[NTHREADS];
 /* We implement some ring transfer, every thread will try to receive a piece of
  * data from its neighbour and increment it before transmitting it to its
  * successor. */
+#ifdef STARPU_USE_CUDA
+void cuda_codelet_unsigned_inc(void *descr[], __attribute__ ((unused)) void *cl_arg);
+#endif
+
 static void increment_handle_cpu_kernel(void *descr[], void *cl_arg __attribute__((unused)))
 {
 	unsigned *val = (unsigned *)STARPU_GET_VARIABLE_PTR(descr[0]);
@@ -49,8 +53,11 @@ static void increment_handle_cpu_kernel(void *descr[], void *cl_arg __attribute_
 }
 
 static starpu_codelet increment_handle_cl = {
-	.where = STARPU_CPU,
+	.where = STARPU_CPU|STARPU_CUDA,
 	.cpu_func = increment_handle_cpu_kernel,
+#ifdef STARPU_USE_CUDA
+	.cuda_func = cuda_codelet_unsigned_inc,
+#endif
 	.nbuffers = 1
 };
 

tests/datawizard/mpi_like_async.c

@@ -42,6 +42,10 @@ static struct thread_data problem_data[NTHREADS];
  * data from its neighbour and increment it before transmitting it to its
  * successor. */
 
+#ifdef STARPU_USE_CUDA
+void cuda_codelet_unsigned_inc(void *descr[], __attribute__ ((unused)) void *cl_arg);
+#endif
+
 static void increment_handle_cpu_kernel(void *descr[], void *cl_arg __attribute__((unused)))
 {
 	unsigned *val = (unsigned *)STARPU_GET_VARIABLE_PTR(descr[0]);
@@ -49,8 +53,11 @@ static void increment_handle_cpu_kernel(void *descr[], void *cl_arg __attribute_
 }
 
 static starpu_codelet increment_handle_cl = {
-	.where = STARPU_CPU,
+	.where = STARPU_CPU|STARPU_CUDA,
 	.cpu_func = increment_handle_cpu_kernel,
+#ifdef STARPU_USE_CUDA
+	.cuda_func = cuda_codelet_unsigned_inc,
+#endif
 	.nbuffers = 1
 };