Make OpenCL dummyKernel to require a input parameter

This is a tentative fix for OpenCL error with mdrun on MacOS.
MacOS seems to require that kernel functions have at least one parameter.
Otherwise clBuildProgram fails with error -11.

Here dummyKernel takes an empty pointer in the definition.
We then pass an empty pointer to dummyKernel with clSetKernelArg.

Fixes #2865

Change-Id: Ib7c08eeeb2ec6d8a43bfe703cf1b273819a45a29

diff --git a/src/gromacs/gpu_utils/gpu_utils_ocl.cpp b/src/gromacs/gpu_utils/gpu_utils_ocl.cpp
index 3efe5d40dbc350589ade552eaaff6de078ddc753..893ec0a9a85063432602a445fb920b3d32bc9103 100644 (file)
--- a/src/gromacs/gpu_utils/gpu_utils_ocl.cpp
+++ b/src/gromacs/gpu_utils/gpu_utils_ocl.cpp
@@ -169,7 +169,8 @@ static bool isDeviceSane(const gmx_device_info_t *devInfo,
         return false;
     }
 
-    const char *lines[] = { "__kernel void dummyKernel(){}" };
+    // Some compilers such as Apple's require kernel functions to have at least one argument
+    const char *lines[] = { "__kernel void dummyKernel(__global void* input){}" };
     ClProgram   program(clCreateProgramWithSource(context, 1, lines, nullptr, &status));
     if (status != CL_SUCCESS)
     {
@@ -190,6 +191,8 @@ static bool isDeviceSane(const gmx_device_info_t *devInfo,
         return false;
     }
 
+    clSetKernelArg(kernel, 0, sizeof(void*), nullptr);
+
     const size_t localWorkSize = 1, globalWorkSize = 1;
     if ((status =
              clEnqueueNDRangeKernel(commandQueue, kernel, 1, nullptr,