tests/datawizard/acquire_release: add cuda codelet and fix access_mode when calling starpu_data_acquire_cb

tests/Makefile.am

@@ -242,6 +242,12 @@ core_get_current_task_SOURCES =			\
 
 testbin_PROGRAMS +=				\
 	datawizard/acquire_release
+datawizard_acquire_release_SOURCES =		\
+	datawizard/acquire_release.c
+if STARPU_USE_CUDA
+datawizard_acquire_release_SOURCES +=		\
+	datawizard/acquire_release_cuda.cu
+endif
 
 testbin_PROGRAMS +=				\
 	datawizard/data_implicit_deps

tests/datawizard/acquire_release.c

@@ -18,6 +18,10 @@
 
 static unsigned ntasks = 10000;
 
+#ifdef STARPU_USE_CUDA
+extern void increment_cuda(void *descr[], __attribute__ ((unused)) void *_args);
+#endif
+
 void increment_cpu(void *descr[], __attribute__ ((unused)) void *_args)
 {
 	unsigned *tokenptr = (unsigned *)STARPU_VARIABLE_GET_PTR(descr[0]);
@@ -25,8 +29,11 @@ void increment_cpu(void *descr[], __attribute__ ((unused)) void *_args)
 }
 
 static starpu_codelet increment_cl = {
-        .where = STARPU_CPU,
+        .where = STARPU_CPU|STARPU_CUDA,
 	.cpu_func = increment_cpu,
+#ifdef STARPU_USE_CUDA
+	.cuda_func = increment_cuda,
+#endif
 	.nbuffers = 1
 };
 
@@ -66,7 +73,7 @@ int main(int argc, char **argv)
 
                 increment_token();
 
-                starpu_data_acquire_cb(token_handle, STARPU_R, callback, NULL);
+                starpu_data_acquire_cb(token_handle, STARPU_RW, callback, NULL);
 	}
 
 	starpu_data_unregister(token_handle);

tests/datawizard/acquire_release_cuda.cu

@@ -0,0 +1,31 @@
+/*
+ * StarPU
+ * Copyright (C) Université Bordeaux 1, CNRS 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 _increment_cuda_codelet(unsigned *val)
+{
+	val[0]++;
+}
+
+extern "C" void increment_cuda(void *descr[], STARPU_ATTRIBUTE_UNUSED void *cl_arg)
+{
+	unsigned *val = (unsigned *)STARPU_VARIABLE_GET_PTR(descr[0]);
+
+	_increment_cuda_codelet<<<1,1>>>(val);
+
+	cudaThreadSynchronize();
+}