clover: Implement the OpenCL ICDclover-icd

25 files changed, 423 insertions, 21 deletions

diff --git a/configure.ac b/configure.ac
index 5bc8c94a2ec..b4178bc87df 100644
--- a/configure.ac
+++ b/configure.ac
@@ -612,6 +612,12 @@ AC_ARG_ENABLE([opencl],
           @<:@default=no@:>@])],
    [],
    [enable_opencl=no])
+AC_ARG_ENABLE([opencl_icd],
+   [AS_HELP_STRING([--enable-opencl-icd],
+          [Build an OpenCL library that can be loaded by an ICD implementation
+           @<:@default=auto@:>@])],
+    [enable_opencl_icd="$enableval"],
+    [enable_opencl_icd="$enable_opencl"])
 AC_ARG_ENABLE([xlib_glx],
     [AS_HELP_STRING([--enable-xlib-glx],
         [make GLX library Xlib-based instead of DRI-based @<:@default=disabled@:>@])],
@@ -1375,8 +1381,23 @@ if test "x$enable_opencl" = xyes; then
     GALLIUM_STATE_TRACKERS_DIRS="$GALLIUM_STATE_TRACKERS_DIRS clover"
     GALLIUM_TARGET_DIRS="$GALLIUM_TARGET_DIRS opencl"
     enable_gallium_loader=yes
+
+    if test "x$enable_opencl_icd" = xyes; then
+        PKG_CHECK_MODULES([OCL_ICD], [ocl-icd],
+                  [], [AC_MSG_ERROR([ocl-icd not found.
+                       If you want to use an ICD loader (recommended), please
+                       install it using a distro provided package, or from source
+                       (https://forge.imag.fr/projects/ocl-icd/).  If you do not want to use an ICD
+                       loader, then  configure mesa with --disable-opencl-icd.)])])
+        OPENCL_LIBNAME="MesaOpenCL"
+    else
+        OPENCL_LIBNAME="OpenCL"
+    fi
+
 fi
 AM_CONDITIONAL(HAVE_CLOVER, test "x$enable_opencl" = xyes)
+AM_CONDITIONAL(HAVE_CLOVER_ICD, test "x$enable_opencl_icd" = xyes)
+AC_SUBST([OPENCL_LIBNAME])
 
 dnl
 dnl Gallium configuration
diff --git a/src/gallium/state_trackers/clover/Makefile.am b/src/gallium/state_trackers/clover/Makefile.am
index b4c197ab139..ca8275e6441 100644
--- a/src/gallium/state_trackers/clover/Makefile.am
+++ b/src/gallium/state_trackers/clover/Makefile.am
@@ -32,6 +32,13 @@ libclllvm_la_SOURCES = \
 libclover_la_CXXFLAGS = \
 	-std=c++0x
 
+if HAVE_CLOVER_ICD
+
+libclover_la_CXXFLAGS += -DUSE_ICD
+
+endif
+
+
 libclover_la_LIBADD = \
 	libcltgsi.la libclllvm.la
 
@@ -74,7 +81,8 @@ libclover_la_SOURCES = \
 	api/sampler.cpp \
 	api/event.cpp \
 	api/program.cpp \
-	api/kernel.cpp
+	api/kernel.cpp \
+	api/icd.cpp
 
 cldir = $(includedir)/CL
 cl_HEADERS = \
diff --git a/src/gallium/state_trackers/clover/api/context.cpp b/src/gallium/state_trackers/clover/api/context.cpp
index 371744120ee..a24459ff2ea 100644
--- a/src/gallium/state_trackers/clover/api/context.cpp
+++ b/src/gallium/state_trackers/clover/api/context.cpp
@@ -20,6 +20,7 @@
 // OTHER DEALINGS IN THE SOFTWARE.
 //
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/context.hpp"
 
diff --git a/src/gallium/state_trackers/clover/api/device.cpp b/src/gallium/state_trackers/clover/api/device.cpp
index e132656032a..0e4a21fbfa7 100644
--- a/src/gallium/state_trackers/clover/api/device.cpp
+++ b/src/gallium/state_trackers/clover/api/device.cpp
@@ -20,6 +20,7 @@
 // OTHER DEALINGS IN THE SOFTWARE.
 //
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/platform.hpp"
 #include "core/device.hpp"
diff --git a/src/gallium/state_trackers/clover/api/event.cpp b/src/gallium/state_trackers/clover/api/event.cpp
index 39a647b5ee7..0a9a1d6ccb3 100644
--- a/src/gallium/state_trackers/clover/api/event.cpp
+++ b/src/gallium/state_trackers/clover/api/event.cpp
@@ -20,6 +20,7 @@
 // OTHER DEALINGS IN THE SOFTWARE.
 //
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/event.hpp"
 
diff --git a/src/gallium/state_trackers/clover/api/icd.cpp b/src/gallium/state_trackers/clover/api/icd.cpp
new file mode 100644
index 00000000000..898df224a47
--- /dev/null
+++ b/src/gallium/state_trackers/clover/api/icd.cpp
@@ -0,0 +1,145 @@
+//
+// Copyright 2013 Advanced Micro Devices Inc.
+//
+// Permission is hereby granted, free of charge, to any person obtaining a
+// copy of this software and associated documentation files (the "Software"),
+// to deal in the Software without restriction, including without limitation
+// the rights to use, copy, modify, merge, publish, distribute, sublicense,
+// and/or sell copies of the Software, and to permit persons to whom the
+// Software is furnished to do so, subject to the following conditions:
+//
+// The above copyright notice and this permission notice shall be included in
+// all copies or substantial portions of the Software.
+//
+// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
+// THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR
+// OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE,
+// ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
+// OTHER DEALINGS IN THE SOFTWARE.
+//
+//
+// Author: Tom Stellard <thomas.stellard@amd.com>
+//
+
+#include "api/icd.hpp"
+
+
+const cl_icd_dispatch clover_icd_dispatch =
+#ifndef USE_ICD
+  0
+#else
+{
+  &CLOVER_API(clGetPlatformIDs),
+  &CLOVER_API(clGetPlatformInfo),
+  &CLOVER_API(clGetDeviceIDs),
+  &CLOVER_API(clGetDeviceInfo),
+  &CLOVER_API(clCreateContext),
+  &CLOVER_API(clCreateContextFromType),
+  &CLOVER_API(clRetainContext),
+  &CLOVER_API(clReleaseContext),
+  &CLOVER_API(clGetContextInfo),
+  &CLOVER_API(clCreateCommandQueue),
+  &CLOVER_API(clRetainCommandQueue),
+  &CLOVER_API(clReleaseCommandQueue),
+  &CLOVER_API(clGetCommandQueueInfo),
+  NULL, //  &CLOVER_API(clSetCommandQueueProperty),
+  &CLOVER_API(clCreateBuffer),
+  NULL, //  &CLOVER_API(clCreateImage2D),
+  NULL, //  &CLOVER_API(clCreateImage3D),
+  &CLOVER_API(clRetainMemObject),
+  &CLOVER_API(clReleaseMemObject),
+  &CLOVER_API(clGetSupportedImageFormats),
+  &CLOVER_API(clGetMemObjectInfo),
+  &CLOVER_API(clGetImageInfo),
+  &CLOVER_API(clCreateSampler),
+  &CLOVER_API(clRetainSampler),
+  &CLOVER_API(clReleaseSampler),
+  &CLOVER_API(clGetSamplerInfo),
+  &CLOVER_API(clCreateProgramWithSource),
+  &CLOVER_API(clCreateProgramWithBinary),
+  &CLOVER_API(clRetainProgram),
+  &CLOVER_API(clReleaseProgram),
+  &CLOVER_API(clBuildProgram),
+  NULL, //  &CLOVER_API(clUnloadCompiler),
+  &CLOVER_API(clGetProgramInfo),
+  &CLOVER_API(clGetProgramBuildInfo),
+  &CLOVER_API(clCreateKernel),
+  &CLOVER_API(clCreateKernelsInProgram),
+  &CLOVER_API(clRetainKernel),
+  &CLOVER_API(clReleaseKernel),
+  &CLOVER_API(clSetKernelArg),
+  &CLOVER_API(clGetKernelInfo),
+  &CLOVER_API(clGetKernelWorkGroupInfo),
+  &CLOVER_API(clWaitForEvents),
+  &CLOVER_API(clGetEventInfo),
+  &CLOVER_API(clRetainEvent),
+  &CLOVER_API(clReleaseEvent),
+  &CLOVER_API(clGetEventProfilingInfo),
+  &CLOVER_API(clFlush),
+  &CLOVER_API(clFinish),
+  &CLOVER_API(clEnqueueReadBuffer),
+  &CLOVER_API(clEnqueueWriteBuffer),
+  &CLOVER_API(clEnqueueCopyBuffer),
+  &CLOVER_API(clEnqueueReadImage),
+  &CLOVER_API(clEnqueueWriteImage),
+  &CLOVER_API(clEnqueueCopyImage),
+  &CLOVER_API(clEnqueueCopyImageToBuffer),
+  &CLOVER_API(clEnqueueCopyBufferToImage),
+  &CLOVER_API(clEnqueueMapBuffer),
+  &CLOVER_API(clEnqueueMapImage),
+  &CLOVER_API(clEnqueueUnmapMemObject),
+  &CLOVER_API(clEnqueueNDRangeKernel),
+  &CLOVER_API(clEnqueueTask),
+  &CLOVER_API(clEnqueueNativeKernel),
+  NULL, //  &CLOVER_API(clEnqueueMarker),
+  NULL, //  &CLOVER_API(clEnqueueWaitForEvents),
+  NULL, //  &CLOVER_API(clEnqueueBarrier),
+  clGetExtensionFunctionAddress,
+  NULL, //  &CLOVER_API(clCreateFromGLBuffer),      
+  NULL, //  &CLOVER_API(clCreateFromGLTexture2D),
+  NULL, //  &CLOVER_API(clCreateFromGLTexture3D),
+  NULL, //  &CLOVER_API(clCreateFromGLRenderbuffer), 
+  NULL, //  &CLOVER_API(clGetGLObjectInfo),         
+  NULL, //  &CLOVER_API(clGetGLTextureInfo),        
+  NULL, //  &CLOVER_API(clEnqueueAcquireGLObjects), 
+  NULL, //  &CLOVER_API(clEnqueueReleaseGLObjects), 
+  NULL, //  &CLOVER_API(clGetGLContextInfoKHR),     
+  NULL,
+  NULL,
+  NULL,
+  NULL,
+  NULL,
+  NULL,
+  &CLOVER_API(clSetEventCallback),
+  &CLOVER_API(clCreateSubBuffer),
+  &CLOVER_API(clSetMemObjectDestructorCallback),
+  &CLOVER_API(clCreateUserEvent),
+  &CLOVER_API(clSetUserEventStatus),
+  &CLOVER_API(clEnqueueReadBufferRect),
+  &CLOVER_API(clEnqueueWriteBufferRect),
+  &CLOVER_API(clEnqueueCopyBufferRect),
+  NULL, //  &CLOVER_API(clCreateSubDevicesEXT),     
+  NULL, //  &CLOVER_API(clRetainDeviceEXT),         
+  NULL, //  &CLOVER_API(clReleaseDeviceEXT),        
+  NULL,
+  NULL, //  &CLOVER_API(clCreateSubDevices),        
+  NULL, //  &CLOVER_API(clRetainDevice),            
+  NULL, //  &CLOVER_API(clReleaseDevice),           
+  NULL, //  &CLOVER_API(clCreateImage),
+  NULL, //  &CLOVER_API(clCreateProgramWithBuiltInKernels), 
+  NULL, //  &CLOVER_API(clCompileProgram),          
+  NULL, //  &CLOVER_API(clLinkProgram),             
+  NULL, //  &CLOVER_API(clUnloadPlatformCompiler),  
+  NULL, //  &CLOVER_API(clGetKernelArgInfo),        
+  NULL, //  &CLOVER_API(clEnqueueFillBuffer),       
+  NULL, //  &CLOVER_API(clEnqueueFillImage),        
+  NULL, //  &CLOVER_API(clEnqueueMigrateMemObjects), 
+  NULL, //  &CLOVER_API(clEnqueueMarkerWithWaitList), 
+  NULL, //  &CLOVER_API(clEnqueueBarrierWithWaitList), 
+  NULL, //  &CLOVER_API(clGetExtensionFunctionAddressForPlatform), 
+  NULL, //  &CLOVER_API(clCreateFromGLTexture),     
+}
+#endif // USE_ICD
+;
diff --git a/src/gallium/state_trackers/clover/api/icd.hpp b/src/gallium/state_trackers/clover/api/icd.hpp
new file mode 100644
index 00000000000..6a1a0125056
--- /dev/null
+++ b/src/gallium/state_trackers/clover/api/icd.hpp
@@ -0,0 +1,169 @@
+//
+// Copyright 2013 Advanced Micro Devices Inc.
+//
+// Permission is hereby granted, free of charge, to any person obtaining a
+// copy of this software and associated documentation files (the "Software"),
+// to deal in the Software without restriction, including without limitation
+// the rights to use, copy, modify, merge, publish, distribute, sublicense,
+// and/or sell copies of the Software, and to permit persons to whom the
+// Software is furnished to do so, subject to the following conditions:
+//
+// The above copyright notice and this permission notice shall be included in
+// all copies or substantial portions of the Software.
+//
+// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
+// THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR
+// OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE,
+// ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
+// OTHER DEALINGS IN THE SOFTWARE.
+//
+//
+// Author: Tom Stellard <thomas.stellard@amd.com>
+//
+
+#ifndef __CL_ICD_HPP__
+#define __CL_ICD_HPP__
+
+#ifdef USE_ICD
+typedef struct _cl_icd_dispatch cl_icd_dispatch;
+#else
+typedef unsigned cl_icd_dispatch;
+#endif
+
+extern const cl_icd_dispatch clover_icd_dispatch;
+
+namespace clover {
+
+struct icd_object {
+   const cl_icd_dispatch *dispatch;
+   icd_object() : dispatch(&clover_icd_dispatch) { }
+};
+
+}
+
+#ifndef USE_ICD
+
+#define CLOVER_API(function) function
+
+#else
+
+#define CLOVER_API(function) CLOVER##function
+
+#define clGetPlatformIDs CLOVER_API(clGetPlatformIDs)
+#define clGetPlatformInfo CLOVER_API(clGetPlatformInfo)
+#define clGetDeviceIDs CLOVER_API(clGetDeviceIDs)
+#define clGetDeviceInfo CLOVER_API(clGetDeviceInfo)
+#define clCreateContext CLOVER_API(clCreateContext)
+#define clCreateContextFromType CLOVER_API(clCreateContextFromType)
+#define clRetainContext CLOVER_API(clRetainContext)
+#define clReleaseContext CLOVER_API(clReleaseContext)
+#define clGetContextInfo CLOVER_API(clGetContextInfo)
+#define clCreateCommandQueue CLOVER_API(clCreateCommandQueue)
+#define clRetainCommandQueue CLOVER_API(clRetainCommandQueue)
+#define clReleaseCommandQueue CLOVER_API(clReleaseCommandQueue)
+#define clGetCommandQueueInfo CLOVER_API(clGetCommandQueueInfo)
+#define clSetCommandQueueProperty CLOVER_API(clSetCommandQueueProperty)
+#define clCreateBuffer CLOVER_API(clCreateBuffer)
+#define clCreateImage2D CLOVER_API(clCreateImage2D)
+#define clCreateImage3D CLOVER_API(clCreateImage3D)
+#define clRetainMemObject CLOVER_API(clRetainMemObject)
+#define clReleaseMemObject CLOVER_API(clReleaseMemObject)
+#define clGetSupportedImageFormats CLOVER_API(clGetSupportedImageFormats)
+#define clGetMemObjectInfo CLOVER_API(clGetMemObjectInfo)
+#define clGetImageInfo CLOVER_API(clGetImageInfo)
+#define clCreateSampler CLOVER_API(clCreateSampler)
+#define clRetainSampler CLOVER_API(clRetainSampler)
+#define clReleaseSampler CLOVER_API(clReleaseSampler)
+#define clGetSamplerInfo CLOVER_API(clGetSamplerInfo)
+#define clCreateProgramWithSource CLOVER_API(clCreateProgramWithSource)
+#define clCreateProgramWithBinary CLOVER_API(clCreateProgramWithBinary)
+#define clRetainProgram CLOVER_API(clRetainProgram)
+#define clReleaseProgram CLOVER_API(clReleaseProgram)
+#define clBuildProgram CLOVER_API(clBuildProgram)
+#define clUnloadCompiler CLOVER_API(clUnloadCompiler)
+#define clGetProgramInfo CLOVER_API(clGetProgramInfo)
+#define clGetProgramBuildInfo CLOVER_API(clGetProgramBuildInfo)
+#define clCreateKernel CLOVER_API(clCreateKernel)
+#define clCreateKernelsInProgram CLOVER_API(clCreateKernelsInProgram)
+#define clRetainKernel CLOVER_API(clRetainKernel)
+#define clReleaseKernel CLOVER_API(clReleaseKernel)
+#define clSetKernelArg CLOVER_API(clSetKernelArg)
+#define clGetKernelInfo CLOVER_API(clGetKernelInfo)
+#define clGetKernelWorkGroupInfo CLOVER_API(clGetKernelWorkGroupInfo)
+#define clWaitForEvents CLOVER_API(clWaitForEvents)
+#define clGetEventInfo CLOVER_API(clGetEventInfo)
+#define clRetainEvent CLOVER_API(clRetainEvent)
+#define clReleaseEvent CLOVER_API(clReleaseEvent)
+#define clGetEventProfilingInfo CLOVER_API(clGetEventProfilingInfo)
+#define clFlush CLOVER_API(clFlush)
+#define clFinish CLOVER_API(clFinish)
+#define clEnqueueReadBuffer CLOVER_API(clEnqueueReadBuffer)
+#define clEnqueueWriteBuffer CLOVER_API(clEnqueueWriteBuffer)
+#define clEnqueueCopyBuffer CLOVER_API(clEnqueueCopyBuffer)
+#define clEnqueueReadImage CLOVER_API(clEnqueueReadImage)
+#define clEnqueueWriteImage CLOVER_API(clEnqueueWriteImage)
+#define clEnqueueCopyImage CLOVER_API(clEnqueueCopyImage)
+#define clEnqueueCopyImageToBuffer CLOVER_API(clEnqueueCopyImageToBuffer)
+#define clEnqueueCopyBufferToImage CLOVER_API(clEnqueueCopyBufferToImage)
+#define clEnqueueMapBuffer CLOVER_API(clEnqueueMapBuffer)
+#define clEnqueueMapImage CLOVER_API(clEnqueueMapImage)
+#define clEnqueueUnmapMemObject CLOVER_API(clEnqueueUnmapMemObject)
+#define clEnqueueNDRangeKernel CLOVER_API(clEnqueueNDRangeKernel)
+#define clEnqueueTask CLOVER_API(clEnqueueTask)
+#define clEnqueueNativeKernel CLOVER_API(clEnqueueNativeKernel)
+#define clEnqueueMarker CLOVER_API(clEnqueueMarker)
+#define clEnqueueWaitForEvents CLOVER_API(clEnqueueWaitForEvents)
+#define clEnqueueBarrier CLOVER_API(clEnqueueBarrier)
+//#define clGetExtensionFunctionAddress CLOVER_API(clGetExtensionFunctionAddress)
+#define clCreateFromGLBuffer CLOVER_API(clCreateFromGLBuffer)      
+#define clCreateFromGLTexture2D CLOVER_API(clCreateFromGLTexture2D)
+#define clCreateFromGLTexture3D CLOVER_API(clCreateFromGLTexture3D)
+#define clCreateFromGLRenderbuffer CLOVER_API(clCreateFromGLRenderbuffer) 
+#define clGetGLObjectInfo CLOVER_API(clGetGLObjectInfo)         
+#define clGetGLTextureInfo CLOVER_API(clGetGLTextureInfo)        
+#define clEnqueueAcquireGLObjects CLOVER_API(clEnqueueAcquireGLObjects) 
+#define clEnqueueReleaseGLObjects CLOVER_API(clEnqueueReleaseGLObjects) 
+#define clGetGLContextInfoKHR CLOVER_API(clGetGLContextInfoKHR)     
+
+
+
+
+
+
+#define clSetEventCallback CLOVER_API(clSetEventCallback)
+#define clCreateSubBuffer CLOVER_API(clCreateSubBuffer)
+#define clSetMemObjectDestructorCallback CLOVER_API(clSetMemObjectDestructorCallback)
+#define clCreateUserEvent CLOVER_API(clCreateUserEvent)
+#define clSetUserEventStatus CLOVER_API(clSetUserEventStatus)
+#define clEnqueueReadBufferRect CLOVER_API(clEnqueueReadBufferRect)
+#define clEnqueueWriteBufferRect CLOVER_API(clEnqueueWriteBufferRect)
+#define clEnqueueCopyBufferRect CLOVER_API(clEnqueueCopyBufferRect)
+#define clCreateSubDevicesEXT CLOVER_API(clCreateSubDevicesEXT)     
+#define clRetainDeviceEXT CLOVER_API(clRetainDeviceEXT)         
+#define clReleaseDeviceEXT CLOVER_API(clReleaseDeviceEXT)        
+
+#define clCreateSubDevices CLOVER_API(clCreateSubDevices)        
+#define clRetainDevice CLOVER_API(clRetainDevice)            
+#define clReleaseDevice CLOVER_API(clReleaseDevice)           
+#define clCreateImage CLOVER_API(clCreateImage)
+#define clCreateProgramWithBuiltInKernels CLOVER_API(clCreateProgramWithBuiltInKernels) 
+#define clCompileProgram CLOVER_API(clCompileProgram)          
+#define clLinkProgram CLOVER_API(clLinkProgram)             
+#define clUnloadPlatformCompiler CLOVER_API(clUnloadPlatformCompiler)  
+#define clGetKernelArgInfo CLOVER_API(clGetKernelArgInfo)        
+#define clEnqueueFillBuffer CLOVER_API(clEnqueueFillBuffer)       
+#define clEnqueueFillImage CLOVER_API(clEnqueueFillImage)        
+#define clEnqueueMigrateMemObjects CLOVER_API(clEnqueueMigrateMemObjects) 
+#define clEnqueueMarkerWithWaitList CLOVER_API(clEnqueueMarkerWithWaitList) 
+#define clEnqueueBarrierWithWaitList CLOVER_API(clEnqueueBarrierWithWaitList) 
+//#define clGetExtensionFunctionAddressForPlatform CLOVER_API(clGetExtensionFunctionAddressForPlatform) 
+#define clCreateFromGLTexture CLOVER_API(clCreateFromGLTexture)     
+
+#include <CL/cl.h>
+#include <ocl_icd.h>
+
+#endif
+
+#endif // __CL_ICD_HPP__
diff --git a/src/gallium/state_trackers/clover/api/kernel.cpp b/src/gallium/state_trackers/clover/api/kernel.cpp
index 13113a22b8e..8e30496fcc3 100644
--- a/src/gallium/state_trackers/clover/api/kernel.cpp
+++ b/src/gallium/state_trackers/clover/api/kernel.cpp
@@ -20,6 +20,7 @@
 // OTHER DEALINGS IN THE SOFTWARE.
 //
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/kernel.hpp"
 #include "core/event.hpp"
diff --git a/src/gallium/state_trackers/clover/api/memory.cpp b/src/gallium/state_trackers/clover/api/memory.cpp
index 63b3619f820..2e2f96b49b4 100644
--- a/src/gallium/state_trackers/clover/api/memory.cpp
+++ b/src/gallium/state_trackers/clover/api/memory.cpp
@@ -20,6 +20,7 @@
 // OTHER DEALINGS IN THE SOFTWARE.
 //
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/memory.hpp"
 #include "core/format.hpp"
diff --git a/src/gallium/state_trackers/clover/api/platform.cpp b/src/gallium/state_trackers/clover/api/platform.cpp
index bfac67bc55a..ec8454d5eb2 100644
--- a/src/gallium/state_trackers/clover/api/platform.cpp
+++ b/src/gallium/state_trackers/clover/api/platform.cpp
@@ -20,6 +20,7 @@
 // OTHER DEALINGS IN THE SOFTWARE.
 //
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/platform.hpp"
 
@@ -43,8 +44,9 @@ clGetPlatformIDs(cl_uint num_entries, cl_platform_id *platforms,
 }
 
 PUBLIC cl_int
-clGetPlatformInfo(cl_platform_id platform, cl_platform_info param_name,
-                  size_t size, void *buf, size_t *size_ret) {
+CLOVER_API(clGetPlatformInfo)(cl_platform_id platform,
+                              cl_platform_info param_name,
+                              size_t size, void *buf, size_t *size_ret) {
    if (platform != &__platform)
       return CL_INVALID_PLATFORM;
 
@@ -59,13 +61,45 @@ clGetPlatformInfo(cl_platform_id platform, cl_platform_info param_name,
    case CL_PLATFORM_NAME:
       return string_property(buf, size, size_ret, "Default");
 
+#ifdef USE_ICD
+   case CL_PLATFORM_ICD_SUFFIX_KHR:
+#endif
    case CL_PLATFORM_VENDOR:
       return string_property(buf, size, size_ret, "Mesa");
 
    case CL_PLATFORM_EXTENSIONS:
-      return string_property(buf, size, size_ret, "");
+      return string_property(buf, size, size_ret, "cl_khr_icd");
 
    default:
       return CL_INVALID_VALUE;
    }
 }
+
+#ifdef USE_ICD
+
+#undef clGetPlatformInfo
+
+PUBLIC cl_int
+clGetPlatformInfo(cl_platform_id platform, cl_platform_info param_name,
+                              size_t size, void *buf, size_t *size_ret) {
+   return CLOVER_API(clGetPlatformInfo)(platform, param_name, size, buf,
+                                        size_ret);
+}
+
+PUBLIC cl_int
+clIcdGetPlatformIDsKHR(cl_uint num_entries, cl_platform_id *platforms,
+                       cl_uint *num_platforms) {
+   return clGetPlatformIDs(num_entries, platforms, num_platforms);
+}
+
+PUBLIC void*
+clGetExtensionFunctionAddress(const char *funcname) {
+   if (strcmp(funcname, "clIcdGetPlatformIDsKHR") == 0) {
+      return (void*)&clIcdGetPlatformIDsKHR;
+   } else if (strcmp(funcname, "clGetPlatformInfo") == 0) {
+      return (void*)&clGetPlatformInfo;
+   }
+   return NULL;
+}
+
+#endif
diff --git a/src/gallium/state_trackers/clover/api/program.cpp b/src/gallium/state_trackers/clover/api/program.cpp
index 6e5ca08d76b..9399b3f2e5c 100644
--- a/src/gallium/state_trackers/clover/api/program.cpp
+++ b/src/gallium/state_trackers/clover/api/program.cpp
@@ -20,6 +20,7 @@
 // OTHER DEALINGS IN THE SOFTWARE.
 //
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/program.hpp"
 
diff --git a/src/gallium/state_trackers/clover/api/queue.cpp b/src/gallium/state_trackers/clover/api/queue.cpp
index 9c7d7e6120d..66aeff3dd28 100644
--- a/src/gallium/state_trackers/clover/api/queue.cpp
+++ b/src/gallium/state_trackers/clover/api/queue.cpp
@@ -20,6 +20,7 @@
 // OTHER DEALINGS IN THE SOFTWARE.
 //
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/queue.hpp"
 
diff --git a/src/gallium/state_trackers/clover/api/sampler.cpp b/src/gallium/state_trackers/clover/api/sampler.cpp
index 4f5ea2fd898..b698c2bb280 100644
--- a/src/gallium/state_trackers/clover/api/sampler.cpp
+++ b/src/gallium/state_trackers/clover/api/sampler.cpp
@@ -20,6 +20,7 @@
 // OTHER DEALINGS IN THE SOFTWARE.
 //
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/sampler.hpp"
 
diff --git a/src/gallium/state_trackers/clover/api/transfer.cpp b/src/gallium/state_trackers/clover/api/transfer.cpp
index 1615d6fdfdd..a911323a570 100644
--- a/src/gallium/state_trackers/clover/api/transfer.cpp
+++ b/src/gallium/state_trackers/clover/api/transfer.cpp
@@ -22,6 +22,7 @@
 
 #include <cstring>
 
+#include "api/icd.hpp"
 #include "api/util.hpp"
 #include "core/event.hpp"
 #include "core/resource.hpp"
diff --git a/src/gallium/state_trackers/clover/core/context.hpp b/src/gallium/state_trackers/clover/core/context.hpp
index 9d19b15aa06..7364a2f0d01 100644
--- a/src/gallium/state_trackers/clover/core/context.hpp
+++ b/src/gallium/state_trackers/clover/core/context.hpp
@@ -25,12 +25,13 @@
 
 #include "core/base.hpp"
 #include "core/device.hpp"
+#include "api/icd.hpp"
 
 namespace clover {
    typedef struct _cl_context context;
 }
 
-struct _cl_context : public clover::ref_counter {
+struct _cl_context : public clover::icd_object, public clover::ref_counter {
 public:
    _cl_context(const std::vector<cl_context_properties> &props,
                const std::vector<clover::device *> &devs);
diff --git a/src/gallium/state_trackers/clover/core/device.hpp b/src/gallium/state_trackers/clover/core/device.hpp
index 465af9f197d..ea2be82ecd7 100644
--- a/src/gallium/state_trackers/clover/core/device.hpp
+++ b/src/gallium/state_trackers/clover/core/device.hpp
@@ -29,6 +29,7 @@
 #include "core/base.hpp"
 #include "core/format.hpp"
 #include "pipe-loader/pipe_loader.h"
+#include "api/icd.hpp"
 
 namespace clover {
    typedef struct _cl_device_id device;
@@ -37,7 +38,7 @@ namespace clover {
    class hard_event;
 }
 
-struct _cl_device_id {
+struct _cl_device_id : public clover::icd_object {
 public:
    _cl_device_id(clover::platform &platform, pipe_loader_device *ldev);
    _cl_device_id(_cl_device_id &&dev);
diff --git a/src/gallium/state_trackers/clover/core/event.hpp b/src/gallium/state_trackers/clover/core/event.hpp
index eb8195383ab..df159bcfc55 100644
--- a/src/gallium/state_trackers/clover/core/event.hpp
+++ b/src/gallium/state_trackers/clover/core/event.hpp
@@ -27,6 +27,7 @@
 
 #include "core/base.hpp"
 #include "core/queue.hpp"
+#include "api/icd.hpp"
 
 namespace clover {
    typedef struct _cl_event event;
@@ -49,7 +50,7 @@ namespace clover {
 /// the status() method, and it can be waited for completion using the
 /// wait() method.
 ///
-struct _cl_event : public clover::ref_counter {
+struct _cl_event : public clover::icd_object, public clover::ref_counter {
 public:
    typedef std::function<void (clover::event &)> action;
 
diff --git a/src/gallium/state_trackers/clover/core/kernel.hpp b/src/gallium/state_trackers/clover/core/kernel.hpp
index fd64f3eb625..4ffd52e862c 100644
--- a/src/gallium/state_trackers/clover/core/kernel.hpp
+++ b/src/gallium/state_trackers/clover/core/kernel.hpp
@@ -30,13 +30,14 @@
 #include "core/memory.hpp"
 #include "core/sampler.hpp"
 #include "pipe/p_state.h"
+#include "api/icd.hpp"
 
 namespace clover {
    typedef struct _cl_kernel kernel;
    class argument;
 }
 
-struct _cl_kernel : public clover::ref_counter {
+struct _cl_kernel : public clover::icd_object, public clover::ref_counter {
 private:
    ///
    /// Class containing all the state required to execute a compute
diff --git a/src/gallium/state_trackers/clover/core/memory.hpp b/src/gallium/state_trackers/clover/core/memory.hpp
index c71a03c8757..18a8975e7ee 100644
--- a/src/gallium/state_trackers/clover/core/memory.hpp
+++ b/src/gallium/state_trackers/clover/core/memory.hpp
@@ -37,7 +37,7 @@ namespace clover {
    class sub_resource;
 }
 
-struct _cl_mem : public clover::ref_counter {
+struct _cl_mem : public clover::icd_object, public clover::ref_counter {
 protected:
    _cl_mem(clover::context &ctx, cl_mem_flags flags,
            size_t size, void *host_ptr);
diff --git a/src/gallium/state_trackers/clover/core/platform.hpp b/src/gallium/state_trackers/clover/core/platform.hpp
index 831a742a4c4..9e8f2e9770c 100644
--- a/src/gallium/state_trackers/clover/core/platform.hpp
+++ b/src/gallium/state_trackers/clover/core/platform.hpp
@@ -27,12 +27,13 @@
 
 #include "core/base.hpp"
 #include "core/device.hpp"
+#include "api/icd.hpp"
 
 namespace clover {
    typedef struct _cl_platform_id platform;
 }
 
-struct _cl_platform_id {
+struct _cl_platform_id : public clover::icd_object {
 public:
    typedef std::vector<clover::device>::iterator iterator;
 
diff --git a/src/gallium/state_trackers/clover/core/program.hpp b/src/gallium/state_trackers/clover/core/program.hpp
index a1452df213e..c83b5fcdc1a 100644
--- a/src/gallium/state_trackers/clover/core/program.hpp
+++ b/src/gallium/state_trackers/clover/core/program.hpp
@@ -28,12 +28,13 @@
 #include "core/base.hpp"
 #include "core/context.hpp"
 #include "core/module.hpp"
+#include "api/icd.hpp"
 
 namespace clover {
    typedef struct _cl_program program;
 }
 
-struct _cl_program : public clover::ref_counter {
+struct _cl_program : public clover::icd_object, public clover::ref_counter {
 public:
    _cl_program(clover::context &ctx,
                const std::string &source);
diff --git a/src/gallium/state_trackers/clover/core/queue.hpp b/src/gallium/state_trackers/clover/core/queue.hpp
index eee85279223..db04a54df41 100644
--- a/src/gallium/state_trackers/clover/core/queue.hpp
+++ b/src/gallium/state_trackers/clover/core/queue.hpp
@@ -26,6 +26,7 @@
 #include "core/base.hpp"
 #include "core/context.hpp"
 #include "pipe/p_context.h"
+#include "api/icd.hpp"
 
 namespace clover {
    typedef struct _cl_command_queue command_queue;
@@ -34,7 +35,8 @@ namespace clover {
    class hard_event;
 }
 
-struct _cl_command_queue : public clover::ref_counter {
+struct _cl_command_queue : public clover::icd_object,
+                           public clover::ref_counter {
 public:
    _cl_command_queue(clover::context &ctx, clover::device &dev,
                      cl_command_queue_properties props);
diff --git a/src/gallium/state_trackers/clover/core/sampler.hpp b/src/gallium/state_trackers/clover/core/sampler.hpp
index 850240e0a12..d1867d9a8a8 100644
--- a/src/gallium/state_trackers/clover/core/sampler.hpp
+++ b/src/gallium/state_trackers/clover/core/sampler.hpp
@@ -25,12 +25,13 @@
 
 #include "core/base.hpp"
 #include "core/queue.hpp"
+#include "api/icd.hpp"
 
 namespace clover {
    typedef struct _cl_sampler sampler;
 }
 
-struct _cl_sampler : public clover::ref_counter {
+struct _cl_sampler : public clover::icd_object, public clover::ref_counter {
 public:
    _cl_sampler(clover::context &ctx, bool norm_mode,
                cl_addressing_mode addr_mode, cl_filter_mode filter_mode);
diff --git a/src/gallium/targets/opencl/Makefile.am b/src/gallium/targets/opencl/Makefile.am
index 46bb29fffac..dcba8bb5903 100644
--- a/src/gallium/targets/opencl/Makefile.am
+++ b/src/gallium/targets/opencl/Makefile.am
@@ -1,12 +1,11 @@
 AUTOMAKE_OPTIONS = subdir-objects
 
-lib_LTLIBRARIES = libOpenCL.la
+lib_LTLIBRARIES = lib@OPENCL_LIBNAME@.la
 
-libOpenCL_la_LDFLAGS = \
+lib@OPENCL_LIBNAME@_la_LDFLAGS = \
 	$(LLVM_LDFLAGS) \
 	-version-number 1:0
-
-libOpenCL_la_LIBADD = \
+lib@OPENCL_LIBNAME@_la_LIBADD = \
 	$(top_builddir)/src/gallium/auxiliary/pipe-loader/libpipe_loader.la \
 	$(top_builddir)/src/gallium/winsys/sw/null/libws_null.la \
 	$(top_builddir)/src/gallium/state_trackers/clover/libclover.la \
@@ -29,13 +28,20 @@ libOpenCL_la_LIBADD = \
 	$(LLVM_LIBS)
 
 
-libOpenCL_la_SOURCES =
+lib@OPENCL_LIBNAME@_la_SOURCES =
 
 # Force usage of a C++ linker
-nodist_EXTRA_libOpenCL_la_SOURCES = dummy.cpp
+nodist_EXTRA_lib@OPENCL_LIBNAME@_la_SOURCES = dummy.cpp
+
+if HAVE_CLOVER_ICD
+
+icddir = /etc/OpenCL/vendors/
+icd_DATA = Mesa.icd
+
+endif
 
 # Provide compatibility with scripts for the old Mesa build system for
 # a while by putting a link to the driver into /lib of the build tree.
-all-local: libOpenCL.la
+all-local: lib@OPENCL_LIBNAME@.la
 	$(MKDIR_P) $(top_builddir)/$(LIB_DIR)
-	ln -f .libs/libOpenCL.so* $(top_builddir)/$(LIB_DIR)/
+	ln -f .libs/lib@OPENCL_LIBNAME@.so* $(top_builddir)/$(LIB_DIR)/
diff --git a/src/gallium/targets/opencl/Mesa.icd b/src/gallium/targets/opencl/Mesa.icd
new file mode 100644
index 00000000000..6a6a8706d7c
--- /dev/null
+++ b/src/gallium/targets/opencl/Mesa.icd
@@ -0,0 +1 @@
+libMesaOpenCL.so