doc: update hybrid example

doc/starpu.texi

@@ -1967,6 +1967,8 @@ __kernel void vector_mult_opencl(__global float* val, int nx, float factor)
 #include <starpu.h>
 #include <starpu_opencl.h>
 
+extern struct starpu_opencl_codelet codelet;
+
 void scal_opencl_func(void *buffers[], void *_args)
 @{
     float *factor = (float *)_args;
@@ -1983,8 +1985,7 @@ void scal_opencl_func(void *buffers[], void *_args)
     id = starpu_worker_get_id();
     devid = starpu_worker_get_devid(id);
 
-    err = starpu_opencl_load_kernel(&kernel, &queue,
-                    "examples/basic_examples/vector_scal_opencl_codelet.cl",
+    err = starpu_opencl_load_kernel(&kernel, &queue, &codelet,
                     "vector_mult_opencl", devid);   /* @b{Name of the codelet defined above} */
     if (err != CL_SUCCESS) STARPU_OPENCL_REPORT_ERROR(err);
 

doc/vector_scal_c.texi

@@ -26,8 +26,9 @@
 #include <stdio.h>
 #include <stdint.h>
 #include <starpu.h>
+#include <starpu_opencl.h>
 
-#define    NX	2048
+#define    NX    2048
 
 /* This kernel takes a buffer and scales it by a constant factor */
 static void scal_cpu_func(void *buffers[], void *cl_arg)
@@ -58,7 +59,7 @@ static void scal_cpu_func(void *buffers[], void *cl_arg)
 
     /* scale the vector */
     for (i = 0; i < n; i++)
-    	val[i] *= *factor;
+        val[i] *= *factor;
 @}
 
 extern void scal_cuda_func(void *buffers[], void *_args);
@@ -67,12 +68,12 @@ extern void scal_opencl_func(void *buffers[], void *_args);
 static starpu_codelet cl = @{
     .where = STARPU_CPU
 #ifdef STARPU_USE_CUDA
-    	| STARPU_CUDA
+        | STARPU_CUDA
 #endif
 #ifdef STARPU_USE_OPENCL
-    	| STARPU_OPENCL
+        | STARPU_OPENCL
 #endif
-    	,
+        ,
     /* CPU implementation of the codelet */
     .cpu_func = scal_cpu_func,
 #ifdef STARPU_USE_CUDA
@@ -86,6 +87,10 @@ static starpu_codelet cl = @{
     .nbuffers = 1
 @};
 
+#ifdef STARPU_USE_OPENCL
+struct starpu_opencl_codelet codelet;
+#endif
+
 int main(int argc, char **argv)
 @{
     /* We consider a vector of float that is initialized just as any of C
@@ -93,7 +98,7 @@ int main(int argc, char **argv)
     float vector[NX];
     unsigned i;
     for (i = 0; i < NX; i++)
-    	vector[i] = 1.0f;
+        vector[i] = 1.0f;
 
     fprintf(stderr, "BEFORE : First element was %f\n", vector[0]);
 
@@ -101,7 +106,8 @@ int main(int argc, char **argv)
     starpu_init(NULL);
 
 #ifdef STARPU_USE_OPENCL
-        _starpu_opencl_compile_source_to_opencl("examples/basic_examples/vector_scal_opencl_codelet.cl");
+        starpu_opencl_load_opencl_from_file("examples/basic_examples/vector_scal_opencl_codelet.cl",
+                                            &codelet);
 #endif
 
     /* Tell StaPU to associate the "vector" vector with the "vector_handle"
@@ -146,6 +152,10 @@ int main(int argc, char **argv)
       * monitoring it */
     starpu_data_unregister(vector_handle);
 
+#ifdef STARPU_USE_OPENCL
+        starpu_opencl_unload_opencl(&codelet);
+#endif
+
     /* terminate StarPU, no task can be submitted after */
     starpu_shutdown();
 

doc/vector_scal_opencl.texi

@@ -21,6 +21,8 @@
 #include <starpu.h>
 #include <starpu_opencl.h>
 
+extern struct starpu_opencl_codelet codelet;
+
 void scal_opencl_func(void *buffers[], void *_args)
 @{
     float *factor = (float *)_args;
@@ -37,9 +39,7 @@ void scal_opencl_func(void *buffers[], void *_args)
     id = starpu_worker_get_id();
     devid = starpu_worker_get_devid(id);
 
-    err = starpu_opencl_load_kernel(&kernel, &queue,
-                    "examples/basic_examples/vector_scal_opencl_codelet.cl",
-                    "vector_mult_opencl", devid);
+    err = starpu_opencl_load_kernel(&kernel, &queue, &codelet, "vector_mult_opencl", devid);
     if (err != CL_SUCCESS) STARPU_OPENCL_REPORT_ERROR(err);
 
     err = 0;
@@ -49,13 +49,13 @@ void scal_opencl_func(void *buffers[], void *_args)
     if (err) STARPU_OPENCL_REPORT_ERROR(err);
 
     @{
-        size_t global=1;
-        size_t local=1;
+        size_t global=n;
+        size_t local=n;
         err = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &global, &local, 0, NULL, NULL);
         if (err != CL_SUCCESS) STARPU_OPENCL_REPORT_ERROR(err);
     @}
 
     clFinish(queue);
 
-    starpu_opencl_release(kernel);
+    starpu_opencl_release_kernel(kernel);
 @}