Browse Source

examples/stencil: set local_work_size to NULL value so that the OpenCL
implementation will determine how to be break the global work-items
into appropriate work-group instances.

Nathalie Furmento 11 years ago
parent
commit
7399a909ca
2 changed files with 2 additions and 6 deletions
  1. 1 3
      examples/stencil/life_opencl.c
  2. 1 3
      examples/stencil/shadow_opencl.c

+ 1 - 3
examples/stencil/life_opencl.c

@@ -93,10 +93,8 @@ opencl_life_update_host(int bz, const TYPE *old, TYPE *newp, int nx, int ny, int
 		threads_per_dim_y /= 2;
 #if 0
 	unsigned threads_per_dim_z = 4;
-	size_t dimBlock[] = {threads_per_dim_x, threads_per_dim_y, threads_per_dim_z};
 	size_t dimGrid[] = {nx / threads_per_dim_x, ny / threads_per_dim_y, nz / threads_per_dim_z};
 #else
-	size_t dimBlock[] = {threads_per_dim_x, threads_per_dim_y, 1};
 	size_t dimGrid[] = {((nx + threads_per_dim_x-1) / threads_per_dim_x)*threads_per_dim_x, ((ny + threads_per_dim_y-1) / threads_per_dim_y)*threads_per_dim_y, 1};
 #endif
 
@@ -119,7 +117,7 @@ opencl_life_update_host(int bz, const TYPE *old, TYPE *newp, int nx, int ny, int
   clSetKernelArg(kernel, 8, sizeof(iter), &iter);
 
   cl_event ev;
-  clEnqueueNDRangeKernel(cq, kernel, 3, NULL, dimGrid, dimBlock, 0, NULL, &ev);
+  clEnqueueNDRangeKernel(cq, kernel, 3, NULL, dimGrid, NULL, 0, NULL, &ev);
   clWaitForEvents(1, &ev);
   starpu_opencl_collect_stats(ev);
   clReleaseEvent(ev);

+ 1 - 3
examples/stencil/shadow_opencl.c

@@ -90,10 +90,8 @@ opencl_shadow_host(int bz, TYPE *ptr, int nx, int ny, int nz, int ldy, int ldz,
 		threads_per_dim_y /= 2;
 #if 0
 	unsigned threads_per_dim_z = 4;
-	size_t dimBlock[] = {threads_per_dim_x, threads_per_dim_y, threads_per_dim_z};
 	size_t dimGrid[] = {nx / threads_per_dim_x, ny / threads_per_dim_y, nz / threads_per_dim_z};
 #else
-	size_t dimBlock[] = {threads_per_dim_x, threads_per_dim_y, 1};
 	size_t dimGrid[] = {((nx + threads_per_dim_x-1) / threads_per_dim_x)*threads_per_dim_x, ((ny + threads_per_dim_y-1) / threads_per_dim_y)*threads_per_dim_y, 1};
 #endif
 
@@ -115,7 +113,7 @@ opencl_shadow_host(int bz, TYPE *ptr, int nx, int ny, int nz, int ldy, int ldz,
         clSetKernelArg(kernel, 7, sizeof(i), &i);
 
         cl_event ev;
-        cl_int err = clEnqueueNDRangeKernel(cq, kernel, 3, NULL, dimGrid, dimBlock, 0, NULL, &ev);
+        cl_int err = clEnqueueNDRangeKernel(cq, kernel, 3, NULL, dimGrid, NULL, 0, NULL, &ev);
         if (err != CL_SUCCESS)
                 STARPU_OPENCL_REPORT_ERROR(err);
         clWaitForEvents(1, &ev);