From b92bb8b4043aca86e7bbe4fd07c8478e30472135 Mon Sep 17 00:00:00 2001 From: mgiunton Date: Wed, 27 May 2026 14:26:02 +0100 Subject: [PATCH 1/6] adapt intercept layer to load natively via OPENCL_LAYERS --- CMakeLists.txt | 1 + intercept/CMakeLists.txt | 3 + intercept/src/clIntercept.def | 12 +++ intercept/src/dispatch.cpp | 153 ++++++++++++++++++++++++++++++++++ intercept/src/dispatch.h | 6 ++ intercept/src/intercept.cpp | 2 + intercept/src/intercept.h | 7 ++ intercept/src/main.cpp | 105 +++++++++++++++++++++++ 8 files changed, 289 insertions(+) diff --git a/CMakeLists.txt b/CMakeLists.txt index 8f9e2039..2b0156ff 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -39,6 +39,7 @@ option(ENABLE_CLIPROF "Enable cliprof Support and Build the Executable") option(ENABLE_ITT "Enable ITT (Instrumentation Tracing Technology) API Support") option(ENABLE_MDAPI "Enable MDAPI Support" ON) option(ENABLE_HIGH_RESOLUTION_CLOCK "Use the high_resolution_clock for timing instead of the steady_clock") +option(ENABLE_OPENCL_NATIVE_LAYER "Build as a native OpenCL layer for the Khronos ICD loader" OFF) if(WIN32) option(ENABLE_CLICONFIG "Build the cliconfig Configuration Utility" ON) endif() diff --git a/intercept/CMakeLists.txt b/intercept/CMakeLists.txt index 11e81e7e..db32ef23 100644 --- a/intercept/CMakeLists.txt +++ b/intercept/CMakeLists.txt @@ -99,6 +99,9 @@ target_link_libraries(OpenCL ${CMAKE_DL_LIBS}) # git_version.cpp - version information for CLIntercept log # git_version.rc2 - DLL version information (Windows only) target_compile_definitions(OpenCL PRIVATE CLINTERCEPT_CMAKE) +if (ENABLE_OPENCL_NATIVE_LAYER) + target_compile_definitions(OpenCL PRIVATE OPENCL_NATIVE_LAYER) +endif () configure_file(src/git_version.cpp.in "${CMAKE_CURRENT_BINARY_DIR}/git_version.cpp" @ONLY) configure_file(src/git_version.rc.in "${CMAKE_CURRENT_BINARY_DIR}/git_version.rc2" @ONLY) diff --git a/intercept/src/clIntercept.def b/intercept/src/clIntercept.def index 5fbc4e6e..4b8c8eab 100644 --- a/intercept/src/clIntercept.def +++ b/intercept/src/clIntercept.def @@ -4,6 +4,18 @@ LIBRARY opencl.dll EXPORTS +; ========================================================== +; OpenCL Native Layer Entry Points +; +; Required exclusively by the Khronos ICD Loader to +; negotiate and initialize the layer (via OPENCL_LAYERS). +; ========================================================== + clGetLayerInfo + clInitLayer + +; ========================================================== +; Standard OpenCL APIs +; ========================================================== clBuildProgram clCloneKernel clCompileProgram diff --git a/intercept/src/dispatch.cpp b/intercept/src/dispatch.cpp index 61a36170..994069fd 100644 --- a/intercept/src/dispatch.cpp +++ b/intercept/src/dispatch.cpp @@ -11988,6 +11988,159 @@ CL_API_ENTRY cl_int CL_API_CALL clGetMutableCommandInfoKHR( NULL_FUNCTION_POINTER_RETURN_ERROR(CL_INVALID_MUTABLE_COMMAND_KHR); } +#if defined(OPENCL_NATIVE_LAYER) +void initDispatchTableForNativeLayer( + cl_icd_dispatch& dispatch ) +{ + INIT_NATIVE_LAYER_FUNCTION( clGetPlatformIDs ); + INIT_NATIVE_LAYER_FUNCTION( clGetPlatformInfo ); + INIT_NATIVE_LAYER_FUNCTION( clGetDeviceIDs ); + INIT_NATIVE_LAYER_FUNCTION( clGetDeviceInfo ); + INIT_NATIVE_LAYER_FUNCTION( clCreateContext ); + INIT_NATIVE_LAYER_FUNCTION( clCreateContextFromType ); + INIT_NATIVE_LAYER_FUNCTION( clRetainContext ); + INIT_NATIVE_LAYER_FUNCTION( clReleaseContext ); + INIT_NATIVE_LAYER_FUNCTION( clGetContextInfo ); + INIT_NATIVE_LAYER_FUNCTION( clCreateCommandQueue ); + INIT_NATIVE_LAYER_FUNCTION( clRetainCommandQueue ); + INIT_NATIVE_LAYER_FUNCTION( clReleaseCommandQueue ); + INIT_NATIVE_LAYER_FUNCTION( clGetCommandQueueInfo ); + INIT_NATIVE_LAYER_FUNCTION( clSetCommandQueueProperty ); + INIT_NATIVE_LAYER_FUNCTION( clCreateBuffer ); + INIT_NATIVE_LAYER_FUNCTION( clCreateImage2D ); + INIT_NATIVE_LAYER_FUNCTION( clCreateImage3D ); + INIT_NATIVE_LAYER_FUNCTION( clRetainMemObject ); + INIT_NATIVE_LAYER_FUNCTION( clReleaseMemObject ); + INIT_NATIVE_LAYER_FUNCTION( clGetSupportedImageFormats ); + INIT_NATIVE_LAYER_FUNCTION( clGetMemObjectInfo ); + INIT_NATIVE_LAYER_FUNCTION( clGetImageInfo ); + INIT_NATIVE_LAYER_FUNCTION( clCreateSampler ); + INIT_NATIVE_LAYER_FUNCTION( clRetainSampler ); + INIT_NATIVE_LAYER_FUNCTION( clReleaseSampler ); + INIT_NATIVE_LAYER_FUNCTION( clGetSamplerInfo ); + INIT_NATIVE_LAYER_FUNCTION( clCreateProgramWithSource ); + INIT_NATIVE_LAYER_FUNCTION( clCreateProgramWithBinary ); + INIT_NATIVE_LAYER_FUNCTION( clRetainProgram ); + INIT_NATIVE_LAYER_FUNCTION( clReleaseProgram ); + INIT_NATIVE_LAYER_FUNCTION( clBuildProgram ); + INIT_NATIVE_LAYER_FUNCTION( clUnloadCompiler ); + INIT_NATIVE_LAYER_FUNCTION( clGetProgramInfo ); + INIT_NATIVE_LAYER_FUNCTION( clGetProgramBuildInfo ); + INIT_NATIVE_LAYER_FUNCTION( clCreateKernel ); + INIT_NATIVE_LAYER_FUNCTION( clCreateKernelsInProgram ); + INIT_NATIVE_LAYER_FUNCTION( clRetainKernel ); + INIT_NATIVE_LAYER_FUNCTION( clReleaseKernel ); + INIT_NATIVE_LAYER_FUNCTION( clSetKernelArg ); + INIT_NATIVE_LAYER_FUNCTION( clGetKernelInfo ); + INIT_NATIVE_LAYER_FUNCTION( clGetKernelWorkGroupInfo ); + INIT_NATIVE_LAYER_FUNCTION( clWaitForEvents ); + INIT_NATIVE_LAYER_FUNCTION( clGetEventInfo ); + INIT_NATIVE_LAYER_FUNCTION( clRetainEvent ); + INIT_NATIVE_LAYER_FUNCTION( clReleaseEvent ); + INIT_NATIVE_LAYER_FUNCTION( clGetEventProfilingInfo ); + INIT_NATIVE_LAYER_FUNCTION( clFlush ); + INIT_NATIVE_LAYER_FUNCTION( clFinish ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueReadBuffer ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueWriteBuffer ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyBuffer ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueReadImage ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueWriteImage ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyImage ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyImageToBuffer ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyBufferToImage ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueMapBuffer ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueMapImage ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueUnmapMemObject ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueNDRangeKernel ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueTask ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueNativeKernel ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueMarker ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueWaitForEvents ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueBarrier ); + INIT_NATIVE_LAYER_FUNCTION( clGetExtensionFunctionAddress ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLBuffer ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLTexture2D ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLTexture3D ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLRenderbuffer ); + INIT_NATIVE_LAYER_FUNCTION( clGetGLObjectInfo ); + INIT_NATIVE_LAYER_FUNCTION( clGetGLTextureInfo ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueAcquireGLObjects ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueReleaseGLObjects ); + INIT_NATIVE_LAYER_FUNCTION( clGetGLContextInfoKHR ); +#if defined(_WIN32) + INIT_NATIVE_LAYER_FUNCTION( clGetDeviceIDsFromD3D10KHR ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D10BufferKHR ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D10Texture2DKHR ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D10Texture3DKHR ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueAcquireD3D10ObjectsKHR ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueReleaseD3D10ObjectsKHR ); +#endif + INIT_NATIVE_LAYER_FUNCTION( clSetEventCallback ); + INIT_NATIVE_LAYER_FUNCTION( clCreateSubBuffer ); + INIT_NATIVE_LAYER_FUNCTION( clSetMemObjectDestructorCallback ); + INIT_NATIVE_LAYER_FUNCTION( clCreateUserEvent ); + INIT_NATIVE_LAYER_FUNCTION( clSetUserEventStatus ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueReadBufferRect ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueWriteBufferRect ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyBufferRect ); + INIT_NATIVE_LAYER_FUNCTION( clCreateEventFromGLsyncKHR ); + INIT_NATIVE_LAYER_FUNCTION( clCreateSubDevices ); + INIT_NATIVE_LAYER_FUNCTION( clRetainDevice ); + INIT_NATIVE_LAYER_FUNCTION( clReleaseDevice ); + INIT_NATIVE_LAYER_FUNCTION( clCreateImage ); + INIT_NATIVE_LAYER_FUNCTION( clCreateProgramWithBuiltInKernels ); + INIT_NATIVE_LAYER_FUNCTION( clCompileProgram ); + INIT_NATIVE_LAYER_FUNCTION( clLinkProgram ); + INIT_NATIVE_LAYER_FUNCTION( clUnloadPlatformCompiler ); + INIT_NATIVE_LAYER_FUNCTION( clGetKernelArgInfo ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueFillBuffer ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueFillImage ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueMigrateMemObjects ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueMarkerWithWaitList ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueBarrierWithWaitList ); + INIT_NATIVE_LAYER_FUNCTION( clGetExtensionFunctionAddressForPlatform ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLTexture ); +#if defined(_WIN32) + INIT_NATIVE_LAYER_FUNCTION( clGetDeviceIDsFromD3D11KHR ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D11BufferKHR ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D11Texture2DKHR ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D11Texture3DKHR ); + INIT_NATIVE_LAYER_FUNCTION( clCreateFromDX9MediaSurfaceKHR ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueAcquireD3D11ObjectsKHR ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueReleaseD3D11ObjectsKHR ); + INIT_NATIVE_LAYER_FUNCTION( clGetDeviceIDsFromDX9MediaAdapterKHR ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueAcquireDX9MediaSurfacesKHR ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueReleaseDX9MediaSurfacesKHR ); +#endif + INIT_NATIVE_LAYER_FUNCTION( clCreateCommandQueueWithProperties ); + INIT_NATIVE_LAYER_FUNCTION( clCreatePipe ); + INIT_NATIVE_LAYER_FUNCTION( clGetPipeInfo ); + INIT_NATIVE_LAYER_FUNCTION( clSVMAlloc ); + INIT_NATIVE_LAYER_FUNCTION( clSVMFree ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMFree ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMMemcpy ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMMemFill ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMMap ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMUnmap ); + INIT_NATIVE_LAYER_FUNCTION( clCreateSamplerWithProperties ); + INIT_NATIVE_LAYER_FUNCTION( clSetKernelArgSVMPointer ); + INIT_NATIVE_LAYER_FUNCTION( clSetKernelExecInfo ); + INIT_NATIVE_LAYER_FUNCTION( clGetKernelSubGroupInfoKHR ); + INIT_NATIVE_LAYER_FUNCTION( clCloneKernel ); + INIT_NATIVE_LAYER_FUNCTION( clCreateProgramWithIL ); + INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMMigrateMem ); + INIT_NATIVE_LAYER_FUNCTION( clGetDeviceAndHostTimer ); + INIT_NATIVE_LAYER_FUNCTION( clGetHostTimer ); + INIT_NATIVE_LAYER_FUNCTION( clGetKernelSubGroupInfo ); + INIT_NATIVE_LAYER_FUNCTION( clSetDefaultDeviceCommandQueue ); + INIT_NATIVE_LAYER_FUNCTION( clSetProgramReleaseCallback ); + INIT_NATIVE_LAYER_FUNCTION( clSetProgramSpecializationConstant ); + INIT_NATIVE_LAYER_FUNCTION( clCreateBufferWithProperties ); + INIT_NATIVE_LAYER_FUNCTION( clCreateImageWithProperties ); + INIT_NATIVE_LAYER_FUNCTION( clSetContextDestructorCallback ); +} +#endif + #if defined(__APPLE__) #include "OS/OS_mac_interpose.h" #endif diff --git a/intercept/src/dispatch.h b/intercept/src/dispatch.h index b18eda66..1ddca82c 100644 --- a/intercept/src/dispatch.h +++ b/intercept/src/dispatch.h @@ -10,6 +10,12 @@ #include "CL/cl_icd.h" +#define INIT_NATIVE_LAYER_FUNCTION(funcname) \ + dispatch.funcname = CLIRN(funcname) + +void initDispatchTableForNativeLayer( + cl_icd_dispatch& dispatch ); + // Dispatch table for extension APIs: struct CLdispatchX diff --git a/intercept/src/intercept.cpp b/intercept/src/intercept.cpp index 82b9a144..ec64d96d 100644 --- a/intercept/src/intercept.cpp +++ b/intercept/src/intercept.cpp @@ -518,10 +518,12 @@ bool CLIntercept::init() log( "Read OpenCL file name from user parameters: " + libName + "\n" ); log( "Trying to load dispatch from: " + libName + "\n" ); +#if !defined(OPENCL_NATIVE_LAYER) if( initDispatch( libName ) ) { log( "... success!\n" ); } +#endif } else { diff --git a/intercept/src/intercept.h b/intercept/src/intercept.h index a54f1160..827164a5 100644 --- a/intercept/src/intercept.h +++ b/intercept/src/intercept.h @@ -841,6 +841,7 @@ class CLIntercept #endif const cl_icd_dispatch& dispatch() const; + cl_icd_dispatch& dispatch(); const CLdispatchX& dispatchX( cl_accelerator_intel accelerator ) const; const CLdispatchX& dispatchX( cl_command_queue queue ) const; @@ -1592,6 +1593,12 @@ inline const cl_icd_dispatch& CLIntercept::dispatch() const return m_Dispatch; } + +inline cl_icd_dispatch& CLIntercept::dispatch() +{ + return m_Dispatch; +} + /////////////////////////////////////////////////////////////////////////////// // inline const CLdispatchX& CLIntercept::dispatchX( cl_accelerator_intel accelerator ) const diff --git a/intercept/src/main.cpp b/intercept/src/main.cpp index 24aeed1b..7395a7e6 100644 --- a/intercept/src/main.cpp +++ b/intercept/src/main.cpp @@ -16,14 +16,19 @@ CLIntercept* g_pIntercept = NULL; #include +static HINSTANCE g_hInstance = NULL; + BOOL APIENTRY DllMain(HINSTANCE hInstance, DWORD dwReasonForCall, LPVOID lpReserved) { switch(dwReasonForCall) { case DLL_PROCESS_ATTACH: + g_hInstance = hInstance; +#if !defined(OPENCL_NATIVE_LAYER) if( CLIntercept::Create( hInstance, g_pIntercept ) == false ) { return FALSE; } +#endif break; case DLL_PROCESS_DETACH: @@ -51,6 +56,8 @@ void __attribute__((destructor)) CLIntercept_Unload(void); void CLIntercept_Load(void) { +#if !defined(OPENCL_NATIVE_LAYER) + #ifdef __ANDROID__ __android_log_print( ANDROID_LOG_INFO, "clIntercept", ">>Load.pid=%d\n", getpid() ); #endif @@ -58,6 +65,8 @@ void CLIntercept_Load(void) #ifdef __ANDROID__ __android_log_print( ANDROID_LOG_INFO, "clIntercept", "<(&g_hInstance), + &hModule ) ) + { + hInstance = static_cast(hModule); + } +#else + void* hInstance = NULL; +#endif + if( CLIntercept::Create( hInstance, g_pIntercept ) == false ) { + return CL_INVALID_VALUE; + } + + auto& interceptDispatch = g_pIntercept->dispatch(); + interceptDispatch = *target_dispatch; + + layer_dispatch = *target_dispatch; + initDispatchTableForNativeLayer( layer_dispatch ); + + *layer_dispatch_ret = &layer_dispatch; + *num_entries_out = num_entries; +#endif + return CL_SUCCESS; +} From a8bd2032b58341f161575e5e3a58a2b1a55e00c6 Mon Sep 17 00:00:00 2001 From: mgiunton Date: Mon, 17 Aug 2026 15:33:14 +0100 Subject: [PATCH 2/6] Remove ENABLE_OPENCL_NATIVE_LAYER flag to use both methods at runtime --- CMakeLists.txt | 1 - intercept/CMakeLists.txt | 3 - intercept/src/dispatch.cpp | 153 ---------------------------------- intercept/src/dispatch.h | 6 -- intercept/src/intercept.cpp | 160 +++++++++++++++++++++++++++++++++++- intercept/src/intercept.h | 9 +- intercept/src/main.cpp | 43 +--------- 7 files changed, 161 insertions(+), 214 deletions(-) diff --git a/CMakeLists.txt b/CMakeLists.txt index 2b0156ff..8f9e2039 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -39,7 +39,6 @@ option(ENABLE_CLIPROF "Enable cliprof Support and Build the Executable") option(ENABLE_ITT "Enable ITT (Instrumentation Tracing Technology) API Support") option(ENABLE_MDAPI "Enable MDAPI Support" ON) option(ENABLE_HIGH_RESOLUTION_CLOCK "Use the high_resolution_clock for timing instead of the steady_clock") -option(ENABLE_OPENCL_NATIVE_LAYER "Build as a native OpenCL layer for the Khronos ICD loader" OFF) if(WIN32) option(ENABLE_CLICONFIG "Build the cliconfig Configuration Utility" ON) endif() diff --git a/intercept/CMakeLists.txt b/intercept/CMakeLists.txt index db32ef23..11e81e7e 100644 --- a/intercept/CMakeLists.txt +++ b/intercept/CMakeLists.txt @@ -99,9 +99,6 @@ target_link_libraries(OpenCL ${CMAKE_DL_LIBS}) # git_version.cpp - version information for CLIntercept log # git_version.rc2 - DLL version information (Windows only) target_compile_definitions(OpenCL PRIVATE CLINTERCEPT_CMAKE) -if (ENABLE_OPENCL_NATIVE_LAYER) - target_compile_definitions(OpenCL PRIVATE OPENCL_NATIVE_LAYER) -endif () configure_file(src/git_version.cpp.in "${CMAKE_CURRENT_BINARY_DIR}/git_version.cpp" @ONLY) configure_file(src/git_version.rc.in "${CMAKE_CURRENT_BINARY_DIR}/git_version.rc2" @ONLY) diff --git a/intercept/src/dispatch.cpp b/intercept/src/dispatch.cpp index 994069fd..61a36170 100644 --- a/intercept/src/dispatch.cpp +++ b/intercept/src/dispatch.cpp @@ -11988,159 +11988,6 @@ CL_API_ENTRY cl_int CL_API_CALL clGetMutableCommandInfoKHR( NULL_FUNCTION_POINTER_RETURN_ERROR(CL_INVALID_MUTABLE_COMMAND_KHR); } -#if defined(OPENCL_NATIVE_LAYER) -void initDispatchTableForNativeLayer( - cl_icd_dispatch& dispatch ) -{ - INIT_NATIVE_LAYER_FUNCTION( clGetPlatformIDs ); - INIT_NATIVE_LAYER_FUNCTION( clGetPlatformInfo ); - INIT_NATIVE_LAYER_FUNCTION( clGetDeviceIDs ); - INIT_NATIVE_LAYER_FUNCTION( clGetDeviceInfo ); - INIT_NATIVE_LAYER_FUNCTION( clCreateContext ); - INIT_NATIVE_LAYER_FUNCTION( clCreateContextFromType ); - INIT_NATIVE_LAYER_FUNCTION( clRetainContext ); - INIT_NATIVE_LAYER_FUNCTION( clReleaseContext ); - INIT_NATIVE_LAYER_FUNCTION( clGetContextInfo ); - INIT_NATIVE_LAYER_FUNCTION( clCreateCommandQueue ); - INIT_NATIVE_LAYER_FUNCTION( clRetainCommandQueue ); - INIT_NATIVE_LAYER_FUNCTION( clReleaseCommandQueue ); - INIT_NATIVE_LAYER_FUNCTION( clGetCommandQueueInfo ); - INIT_NATIVE_LAYER_FUNCTION( clSetCommandQueueProperty ); - INIT_NATIVE_LAYER_FUNCTION( clCreateBuffer ); - INIT_NATIVE_LAYER_FUNCTION( clCreateImage2D ); - INIT_NATIVE_LAYER_FUNCTION( clCreateImage3D ); - INIT_NATIVE_LAYER_FUNCTION( clRetainMemObject ); - INIT_NATIVE_LAYER_FUNCTION( clReleaseMemObject ); - INIT_NATIVE_LAYER_FUNCTION( clGetSupportedImageFormats ); - INIT_NATIVE_LAYER_FUNCTION( clGetMemObjectInfo ); - INIT_NATIVE_LAYER_FUNCTION( clGetImageInfo ); - INIT_NATIVE_LAYER_FUNCTION( clCreateSampler ); - INIT_NATIVE_LAYER_FUNCTION( clRetainSampler ); - INIT_NATIVE_LAYER_FUNCTION( clReleaseSampler ); - INIT_NATIVE_LAYER_FUNCTION( clGetSamplerInfo ); - INIT_NATIVE_LAYER_FUNCTION( clCreateProgramWithSource ); - INIT_NATIVE_LAYER_FUNCTION( clCreateProgramWithBinary ); - INIT_NATIVE_LAYER_FUNCTION( clRetainProgram ); - INIT_NATIVE_LAYER_FUNCTION( clReleaseProgram ); - INIT_NATIVE_LAYER_FUNCTION( clBuildProgram ); - INIT_NATIVE_LAYER_FUNCTION( clUnloadCompiler ); - INIT_NATIVE_LAYER_FUNCTION( clGetProgramInfo ); - INIT_NATIVE_LAYER_FUNCTION( clGetProgramBuildInfo ); - INIT_NATIVE_LAYER_FUNCTION( clCreateKernel ); - INIT_NATIVE_LAYER_FUNCTION( clCreateKernelsInProgram ); - INIT_NATIVE_LAYER_FUNCTION( clRetainKernel ); - INIT_NATIVE_LAYER_FUNCTION( clReleaseKernel ); - INIT_NATIVE_LAYER_FUNCTION( clSetKernelArg ); - INIT_NATIVE_LAYER_FUNCTION( clGetKernelInfo ); - INIT_NATIVE_LAYER_FUNCTION( clGetKernelWorkGroupInfo ); - INIT_NATIVE_LAYER_FUNCTION( clWaitForEvents ); - INIT_NATIVE_LAYER_FUNCTION( clGetEventInfo ); - INIT_NATIVE_LAYER_FUNCTION( clRetainEvent ); - INIT_NATIVE_LAYER_FUNCTION( clReleaseEvent ); - INIT_NATIVE_LAYER_FUNCTION( clGetEventProfilingInfo ); - INIT_NATIVE_LAYER_FUNCTION( clFlush ); - INIT_NATIVE_LAYER_FUNCTION( clFinish ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueReadBuffer ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueWriteBuffer ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyBuffer ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueReadImage ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueWriteImage ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyImage ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyImageToBuffer ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyBufferToImage ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueMapBuffer ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueMapImage ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueUnmapMemObject ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueNDRangeKernel ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueTask ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueNativeKernel ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueMarker ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueWaitForEvents ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueBarrier ); - INIT_NATIVE_LAYER_FUNCTION( clGetExtensionFunctionAddress ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLBuffer ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLTexture2D ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLTexture3D ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLRenderbuffer ); - INIT_NATIVE_LAYER_FUNCTION( clGetGLObjectInfo ); - INIT_NATIVE_LAYER_FUNCTION( clGetGLTextureInfo ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueAcquireGLObjects ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueReleaseGLObjects ); - INIT_NATIVE_LAYER_FUNCTION( clGetGLContextInfoKHR ); -#if defined(_WIN32) - INIT_NATIVE_LAYER_FUNCTION( clGetDeviceIDsFromD3D10KHR ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D10BufferKHR ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D10Texture2DKHR ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D10Texture3DKHR ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueAcquireD3D10ObjectsKHR ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueReleaseD3D10ObjectsKHR ); -#endif - INIT_NATIVE_LAYER_FUNCTION( clSetEventCallback ); - INIT_NATIVE_LAYER_FUNCTION( clCreateSubBuffer ); - INIT_NATIVE_LAYER_FUNCTION( clSetMemObjectDestructorCallback ); - INIT_NATIVE_LAYER_FUNCTION( clCreateUserEvent ); - INIT_NATIVE_LAYER_FUNCTION( clSetUserEventStatus ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueReadBufferRect ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueWriteBufferRect ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueCopyBufferRect ); - INIT_NATIVE_LAYER_FUNCTION( clCreateEventFromGLsyncKHR ); - INIT_NATIVE_LAYER_FUNCTION( clCreateSubDevices ); - INIT_NATIVE_LAYER_FUNCTION( clRetainDevice ); - INIT_NATIVE_LAYER_FUNCTION( clReleaseDevice ); - INIT_NATIVE_LAYER_FUNCTION( clCreateImage ); - INIT_NATIVE_LAYER_FUNCTION( clCreateProgramWithBuiltInKernels ); - INIT_NATIVE_LAYER_FUNCTION( clCompileProgram ); - INIT_NATIVE_LAYER_FUNCTION( clLinkProgram ); - INIT_NATIVE_LAYER_FUNCTION( clUnloadPlatformCompiler ); - INIT_NATIVE_LAYER_FUNCTION( clGetKernelArgInfo ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueFillBuffer ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueFillImage ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueMigrateMemObjects ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueMarkerWithWaitList ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueBarrierWithWaitList ); - INIT_NATIVE_LAYER_FUNCTION( clGetExtensionFunctionAddressForPlatform ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromGLTexture ); -#if defined(_WIN32) - INIT_NATIVE_LAYER_FUNCTION( clGetDeviceIDsFromD3D11KHR ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D11BufferKHR ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D11Texture2DKHR ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromD3D11Texture3DKHR ); - INIT_NATIVE_LAYER_FUNCTION( clCreateFromDX9MediaSurfaceKHR ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueAcquireD3D11ObjectsKHR ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueReleaseD3D11ObjectsKHR ); - INIT_NATIVE_LAYER_FUNCTION( clGetDeviceIDsFromDX9MediaAdapterKHR ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueAcquireDX9MediaSurfacesKHR ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueReleaseDX9MediaSurfacesKHR ); -#endif - INIT_NATIVE_LAYER_FUNCTION( clCreateCommandQueueWithProperties ); - INIT_NATIVE_LAYER_FUNCTION( clCreatePipe ); - INIT_NATIVE_LAYER_FUNCTION( clGetPipeInfo ); - INIT_NATIVE_LAYER_FUNCTION( clSVMAlloc ); - INIT_NATIVE_LAYER_FUNCTION( clSVMFree ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMFree ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMMemcpy ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMMemFill ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMMap ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMUnmap ); - INIT_NATIVE_LAYER_FUNCTION( clCreateSamplerWithProperties ); - INIT_NATIVE_LAYER_FUNCTION( clSetKernelArgSVMPointer ); - INIT_NATIVE_LAYER_FUNCTION( clSetKernelExecInfo ); - INIT_NATIVE_LAYER_FUNCTION( clGetKernelSubGroupInfoKHR ); - INIT_NATIVE_LAYER_FUNCTION( clCloneKernel ); - INIT_NATIVE_LAYER_FUNCTION( clCreateProgramWithIL ); - INIT_NATIVE_LAYER_FUNCTION( clEnqueueSVMMigrateMem ); - INIT_NATIVE_LAYER_FUNCTION( clGetDeviceAndHostTimer ); - INIT_NATIVE_LAYER_FUNCTION( clGetHostTimer ); - INIT_NATIVE_LAYER_FUNCTION( clGetKernelSubGroupInfo ); - INIT_NATIVE_LAYER_FUNCTION( clSetDefaultDeviceCommandQueue ); - INIT_NATIVE_LAYER_FUNCTION( clSetProgramReleaseCallback ); - INIT_NATIVE_LAYER_FUNCTION( clSetProgramSpecializationConstant ); - INIT_NATIVE_LAYER_FUNCTION( clCreateBufferWithProperties ); - INIT_NATIVE_LAYER_FUNCTION( clCreateImageWithProperties ); - INIT_NATIVE_LAYER_FUNCTION( clSetContextDestructorCallback ); -} -#endif - #if defined(__APPLE__) #include "OS/OS_mac_interpose.h" #endif diff --git a/intercept/src/dispatch.h b/intercept/src/dispatch.h index 1ddca82c..b18eda66 100644 --- a/intercept/src/dispatch.h +++ b/intercept/src/dispatch.h @@ -10,12 +10,6 @@ #include "CL/cl_icd.h" -#define INIT_NATIVE_LAYER_FUNCTION(funcname) \ - dispatch.funcname = CLIRN(funcname) - -void initDispatchTableForNativeLayer( - cl_icd_dispatch& dispatch ); - // Dispatch table for extension APIs: struct CLdispatchX diff --git a/intercept/src/intercept.cpp b/intercept/src/intercept.cpp index ec64d96d..23da6615 100644 --- a/intercept/src/intercept.cpp +++ b/intercept/src/intercept.cpp @@ -518,12 +518,10 @@ bool CLIntercept::init() log( "Read OpenCL file name from user parameters: " + libName + "\n" ); log( "Trying to load dispatch from: " + libName + "\n" ); -#if !defined(OPENCL_NATIVE_LAYER) if( initDispatch( libName ) ) { log( "... success!\n" ); } -#endif } else { @@ -14150,6 +14148,164 @@ bool CLIntercept::initDispatch( void ) #error Unknown OS! #endif +/////////////////////////////////////////////////////////////////////////////// +// +#define INIT_NATIVE_LAYER_FUNC(funcname) \ + layer_dispatch . funcname = CLIRN(funcname) + +void CLIntercept::initLayer(cl_icd_dispatch& layer_dispatch, const cl_icd_dispatch* target_dispatch) +{ + m_Dispatch = *target_dispatch; + layer_dispatch = *target_dispatch; + + INIT_NATIVE_LAYER_FUNC( clGetPlatformIDs ); + INIT_NATIVE_LAYER_FUNC( clGetPlatformInfo ); + INIT_NATIVE_LAYER_FUNC( clGetDeviceIDs ); + INIT_NATIVE_LAYER_FUNC( clGetDeviceInfo ); + INIT_NATIVE_LAYER_FUNC( clCreateContext ); + INIT_NATIVE_LAYER_FUNC( clCreateContextFromType ); + INIT_NATIVE_LAYER_FUNC( clRetainContext ); + INIT_NATIVE_LAYER_FUNC( clReleaseContext ); + INIT_NATIVE_LAYER_FUNC( clGetContextInfo ); + INIT_NATIVE_LAYER_FUNC( clCreateCommandQueue ); + INIT_NATIVE_LAYER_FUNC( clRetainCommandQueue ); + INIT_NATIVE_LAYER_FUNC( clReleaseCommandQueue ); + INIT_NATIVE_LAYER_FUNC( clGetCommandQueueInfo ); + INIT_NATIVE_LAYER_FUNC( clSetCommandQueueProperty ); + INIT_NATIVE_LAYER_FUNC( clCreateBuffer ); + INIT_NATIVE_LAYER_FUNC( clCreateImage2D ); + INIT_NATIVE_LAYER_FUNC( clCreateImage3D ); + INIT_NATIVE_LAYER_FUNC( clRetainMemObject ); + INIT_NATIVE_LAYER_FUNC( clReleaseMemObject ); + INIT_NATIVE_LAYER_FUNC( clGetSupportedImageFormats ); + INIT_NATIVE_LAYER_FUNC( clGetMemObjectInfo ); + INIT_NATIVE_LAYER_FUNC( clGetImageInfo ); + INIT_NATIVE_LAYER_FUNC( clCreateSampler ); + INIT_NATIVE_LAYER_FUNC( clRetainSampler ); + INIT_NATIVE_LAYER_FUNC( clReleaseSampler ); + INIT_NATIVE_LAYER_FUNC( clGetSamplerInfo ); + INIT_NATIVE_LAYER_FUNC( clCreateProgramWithSource ); + INIT_NATIVE_LAYER_FUNC( clCreateProgramWithBinary ); + INIT_NATIVE_LAYER_FUNC( clRetainProgram ); + INIT_NATIVE_LAYER_FUNC( clReleaseProgram ); + INIT_NATIVE_LAYER_FUNC( clBuildProgram ); + INIT_NATIVE_LAYER_FUNC( clUnloadCompiler ); + INIT_NATIVE_LAYER_FUNC( clGetProgramInfo ); + INIT_NATIVE_LAYER_FUNC( clGetProgramBuildInfo ); + INIT_NATIVE_LAYER_FUNC( clCreateKernel ); + INIT_NATIVE_LAYER_FUNC( clCreateKernelsInProgram ); + INIT_NATIVE_LAYER_FUNC( clRetainKernel ); + INIT_NATIVE_LAYER_FUNC( clReleaseKernel ); + INIT_NATIVE_LAYER_FUNC( clSetKernelArg ); + INIT_NATIVE_LAYER_FUNC( clGetKernelInfo ); + INIT_NATIVE_LAYER_FUNC( clGetKernelWorkGroupInfo ); + INIT_NATIVE_LAYER_FUNC( clWaitForEvents ); + INIT_NATIVE_LAYER_FUNC( clGetEventInfo ); + INIT_NATIVE_LAYER_FUNC( clRetainEvent ); + INIT_NATIVE_LAYER_FUNC( clReleaseEvent ); + INIT_NATIVE_LAYER_FUNC( clGetEventProfilingInfo ); + INIT_NATIVE_LAYER_FUNC( clFlush ); + INIT_NATIVE_LAYER_FUNC( clFinish ); + INIT_NATIVE_LAYER_FUNC( clEnqueueReadBuffer ); + INIT_NATIVE_LAYER_FUNC( clEnqueueWriteBuffer ); + INIT_NATIVE_LAYER_FUNC( clEnqueueCopyBuffer ); + INIT_NATIVE_LAYER_FUNC( clEnqueueReadImage ); + INIT_NATIVE_LAYER_FUNC( clEnqueueWriteImage ); + INIT_NATIVE_LAYER_FUNC( clEnqueueCopyImage ); + INIT_NATIVE_LAYER_FUNC( clEnqueueCopyImageToBuffer ); + INIT_NATIVE_LAYER_FUNC( clEnqueueCopyBufferToImage ); + INIT_NATIVE_LAYER_FUNC( clEnqueueMapBuffer ); + INIT_NATIVE_LAYER_FUNC( clEnqueueMapImage ); + INIT_NATIVE_LAYER_FUNC( clEnqueueUnmapMemObject ); + INIT_NATIVE_LAYER_FUNC( clEnqueueNDRangeKernel ); + INIT_NATIVE_LAYER_FUNC( clEnqueueTask ); + INIT_NATIVE_LAYER_FUNC( clEnqueueNativeKernel ); + INIT_NATIVE_LAYER_FUNC( clEnqueueMarker ); + INIT_NATIVE_LAYER_FUNC( clEnqueueWaitForEvents ); + INIT_NATIVE_LAYER_FUNC( clEnqueueBarrier ); + INIT_NATIVE_LAYER_FUNC( clGetExtensionFunctionAddress ); + INIT_NATIVE_LAYER_FUNC( clCreateFromGLBuffer ); + INIT_NATIVE_LAYER_FUNC( clCreateFromGLTexture2D ); + INIT_NATIVE_LAYER_FUNC( clCreateFromGLTexture3D ); + INIT_NATIVE_LAYER_FUNC( clCreateFromGLRenderbuffer ); + INIT_NATIVE_LAYER_FUNC( clGetGLObjectInfo ); + INIT_NATIVE_LAYER_FUNC( clGetGLTextureInfo ); + INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireGLObjects ); + INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseGLObjects ); + INIT_NATIVE_LAYER_FUNC( clGetGLContextInfoKHR ); +#if defined(_WIN32) + INIT_NATIVE_LAYER_FUNC( clGetDeviceIDsFromD3D10KHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromD3D10BufferKHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromD3D10Texture2DKHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromD3D10Texture3DKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireD3D10ObjectsKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseD3D10ObjectsKHR ); +#endif + INIT_NATIVE_LAYER_FUNC( clSetEventCallback ); + INIT_NATIVE_LAYER_FUNC( clCreateSubBuffer ); + INIT_NATIVE_LAYER_FUNC( clSetMemObjectDestructorCallback ); + INIT_NATIVE_LAYER_FUNC( clCreateUserEvent ); + INIT_NATIVE_LAYER_FUNC( clSetUserEventStatus ); + INIT_NATIVE_LAYER_FUNC( clEnqueueReadBufferRect ); + INIT_NATIVE_LAYER_FUNC( clEnqueueWriteBufferRect ); + INIT_NATIVE_LAYER_FUNC( clEnqueueCopyBufferRect ); + INIT_NATIVE_LAYER_FUNC( clCreateEventFromGLsyncKHR ); + INIT_NATIVE_LAYER_FUNC( clCreateSubDevices ); + INIT_NATIVE_LAYER_FUNC( clRetainDevice ); + INIT_NATIVE_LAYER_FUNC( clReleaseDevice ); + INIT_NATIVE_LAYER_FUNC( clCreateImage ); + INIT_NATIVE_LAYER_FUNC( clCreateProgramWithBuiltInKernels ); + INIT_NATIVE_LAYER_FUNC( clCompileProgram ); + INIT_NATIVE_LAYER_FUNC( clLinkProgram ); + INIT_NATIVE_LAYER_FUNC( clUnloadPlatformCompiler ); + INIT_NATIVE_LAYER_FUNC( clGetKernelArgInfo ); + INIT_NATIVE_LAYER_FUNC( clEnqueueFillBuffer ); + INIT_NATIVE_LAYER_FUNC( clEnqueueFillImage ); + INIT_NATIVE_LAYER_FUNC( clEnqueueMigrateMemObjects ); + INIT_NATIVE_LAYER_FUNC( clEnqueueMarkerWithWaitList ); + INIT_NATIVE_LAYER_FUNC( clEnqueueBarrierWithWaitList ); + INIT_NATIVE_LAYER_FUNC( clGetExtensionFunctionAddressForPlatform ); + INIT_NATIVE_LAYER_FUNC( clCreateFromGLTexture ); +#if defined(_WIN32) + INIT_NATIVE_LAYER_FUNC( clGetDeviceIDsFromD3D11KHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromD3D11BufferKHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromD3D11Texture2DKHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromD3D11Texture3DKHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromDX9MediaSurfaceKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireD3D11ObjectsKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseD3D11ObjectsKHR ); + INIT_NATIVE_LAYER_FUNC( clGetDeviceIDsFromDX9MediaAdapterKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireDX9MediaSurfacesKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseDX9MediaSurfacesKHR ); +#endif + INIT_NATIVE_LAYER_FUNC( clCreateCommandQueueWithProperties ); + INIT_NATIVE_LAYER_FUNC( clCreatePipe ); + INIT_NATIVE_LAYER_FUNC( clGetPipeInfo ); + INIT_NATIVE_LAYER_FUNC( clSVMAlloc ); + INIT_NATIVE_LAYER_FUNC( clSVMFree ); + INIT_NATIVE_LAYER_FUNC( clEnqueueSVMFree ); + INIT_NATIVE_LAYER_FUNC( clEnqueueSVMMemcpy ); + INIT_NATIVE_LAYER_FUNC( clEnqueueSVMMemFill ); + INIT_NATIVE_LAYER_FUNC( clEnqueueSVMMap ); + INIT_NATIVE_LAYER_FUNC( clEnqueueSVMUnmap ); + INIT_NATIVE_LAYER_FUNC( clCreateSamplerWithProperties ); + INIT_NATIVE_LAYER_FUNC( clSetKernelArgSVMPointer ); + INIT_NATIVE_LAYER_FUNC( clSetKernelExecInfo ); + INIT_NATIVE_LAYER_FUNC( clGetKernelSubGroupInfoKHR ); + INIT_NATIVE_LAYER_FUNC( clCloneKernel ); + INIT_NATIVE_LAYER_FUNC( clCreateProgramWithIL ); + INIT_NATIVE_LAYER_FUNC( clEnqueueSVMMigrateMem ); + INIT_NATIVE_LAYER_FUNC( clGetDeviceAndHostTimer ); + INIT_NATIVE_LAYER_FUNC( clGetHostTimer ); + INIT_NATIVE_LAYER_FUNC( clGetKernelSubGroupInfo ); + INIT_NATIVE_LAYER_FUNC( clSetDefaultDeviceCommandQueue ); + INIT_NATIVE_LAYER_FUNC( clSetProgramReleaseCallback ); + INIT_NATIVE_LAYER_FUNC( clSetProgramSpecializationConstant ); + INIT_NATIVE_LAYER_FUNC( clCreateBufferWithProperties ); + INIT_NATIVE_LAYER_FUNC( clCreateImageWithProperties ); + INIT_NATIVE_LAYER_FUNC( clSetContextDestructorCallback ); +} + /////////////////////////////////////////////////////////////////////////////// // void CLIntercept::dumpMemoryToFile( diff --git a/intercept/src/intercept.h b/intercept/src/intercept.h index 827164a5..288dce5b 100644 --- a/intercept/src/intercept.h +++ b/intercept/src/intercept.h @@ -61,6 +61,8 @@ class CLIntercept static bool Create( void* pGlobalData, CLIntercept*& pIntercept ); static void Delete( CLIntercept*& pIntercept ); + void initLayer(cl_icd_dispatch& layer_dispatch, const cl_icd_dispatch* target_dispatch); + void report(); void callLoggingEnter( @@ -841,7 +843,6 @@ class CLIntercept #endif const cl_icd_dispatch& dispatch() const; - cl_icd_dispatch& dispatch(); const CLdispatchX& dispatchX( cl_accelerator_intel accelerator ) const; const CLdispatchX& dispatchX( cl_command_queue queue ) const; @@ -1593,12 +1594,6 @@ inline const cl_icd_dispatch& CLIntercept::dispatch() const return m_Dispatch; } - -inline cl_icd_dispatch& CLIntercept::dispatch() -{ - return m_Dispatch; -} - /////////////////////////////////////////////////////////////////////////////// // inline const CLdispatchX& CLIntercept::dispatchX( cl_accelerator_intel accelerator ) const diff --git a/intercept/src/main.cpp b/intercept/src/main.cpp index 7395a7e6..7b197967 100644 --- a/intercept/src/main.cpp +++ b/intercept/src/main.cpp @@ -16,19 +16,14 @@ CLIntercept* g_pIntercept = NULL; #include -static HINSTANCE g_hInstance = NULL; - BOOL APIENTRY DllMain(HINSTANCE hInstance, DWORD dwReasonForCall, LPVOID lpReserved) { switch(dwReasonForCall) { case DLL_PROCESS_ATTACH: - g_hInstance = hInstance; -#if !defined(OPENCL_NATIVE_LAYER) if( CLIntercept::Create( hInstance, g_pIntercept ) == false ) { return FALSE; } -#endif break; case DLL_PROCESS_DETACH: @@ -56,8 +51,6 @@ void __attribute__((destructor)) CLIntercept_Unload(void); void CLIntercept_Load(void) { -#if !defined(OPENCL_NATIVE_LAYER) - #ifdef __ANDROID__ __android_log_print( ANDROID_LOG_INFO, "clIntercept", ">>Load.pid=%d\n", getpid() ); #endif @@ -65,8 +58,6 @@ void CLIntercept_Load(void) #ifdef __ANDROID__ __android_log_print( ANDROID_LOG_INFO, "clIntercept", "<(&g_hInstance), - &hModule ) ) - { - hInstance = static_cast(hModule); - } -#else - void* hInstance = NULL; -#endif - if( CLIntercept::Create( hInstance, g_pIntercept ) == false ) { - return CL_INVALID_VALUE; - } - - auto& interceptDispatch = g_pIntercept->dispatch(); - interceptDispatch = *target_dispatch; - - layer_dispatch = *target_dispatch; - initDispatchTableForNativeLayer( layer_dispatch ); - + g_pIntercept->initLayer( layer_dispatch, target_dispatch ); *layer_dispatch_ret = &layer_dispatch; *num_entries_out = num_entries; -#endif return CL_SUCCESS; } From e8a3c9f2b8753daf38d89cd4882edd1ce96e9e86 Mon Sep 17 00:00:00 2001 From: mgiunton Date: Wed, 26 Aug 2026 13:20:06 +0100 Subject: [PATCH 3/6] Fix CI build failures --- intercept/src/intercept.cpp | 7 +++++-- intercept/src/main.cpp | 4 ++-- 2 files changed, 7 insertions(+), 4 deletions(-) diff --git a/intercept/src/intercept.cpp b/intercept/src/intercept.cpp index 23da6615..33522080 100644 --- a/intercept/src/intercept.cpp +++ b/intercept/src/intercept.cpp @@ -14150,8 +14150,11 @@ bool CLIntercept::initDispatch( void ) /////////////////////////////////////////////////////////////////////////////// // -#define INIT_NATIVE_LAYER_FUNC(funcname) \ - layer_dispatch . funcname = CLIRN(funcname) +#define INIT_NATIVE_LAYER_FUNC(funcname) \ +{ \ + void** pfunc = (void**)( &layer_dispatch . funcname ); \ + *pfunc = CLIRN(funcname); \ +} void CLIntercept::initLayer(cl_icd_dispatch& layer_dispatch, const cl_icd_dispatch* target_dispatch) { diff --git a/intercept/src/main.cpp b/intercept/src/main.cpp index 7b197967..00215759 100644 --- a/intercept/src/main.cpp +++ b/intercept/src/main.cpp @@ -94,7 +94,7 @@ extern "C" CL_API_ENTRY cl_int CL_API_CALL clGetLayerInfo( if (param_value_size < sizeof(version)) return CL_INVALID_VALUE; - std::memcpy(param_value, &version, sizeof(version)); + CLI_MEMCPY(param_value, param_value_size, &version, sizeof(version)); } return CL_SUCCESS; } @@ -107,7 +107,7 @@ extern "C" CL_API_ENTRY cl_int CL_API_CALL clGetLayerInfo( if (param_value_size < sizeof(name)) return CL_INVALID_VALUE; - std::memcpy(param_value, name, sizeof(name)); + CLI_MEMCPY(param_value, param_value_size, name, sizeof(name)); } return CL_SUCCESS; } From 70b953c0232c782c6b18c5cad393c1245b0700fa Mon Sep 17 00:00:00 2001 From: mgiunton Date: Thu, 27 Aug 2026 17:06:28 +0100 Subject: [PATCH 4/6] Fix cast from function ptr to void ptr --- intercept/src/intercept.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/intercept/src/intercept.cpp b/intercept/src/intercept.cpp index 33522080..185d0533 100644 --- a/intercept/src/intercept.cpp +++ b/intercept/src/intercept.cpp @@ -14153,7 +14153,7 @@ bool CLIntercept::initDispatch( void ) #define INIT_NATIVE_LAYER_FUNC(funcname) \ { \ void** pfunc = (void**)( &layer_dispatch . funcname ); \ - *pfunc = CLIRN(funcname); \ + *pfunc = reinterpret_cast(CLIRN(funcname)); \ } void CLIntercept::initLayer(cl_icd_dispatch& layer_dispatch, const cl_icd_dispatch* target_dispatch) From 5e1b9aad6220101ccd8ca38ec3b1e7ffe949b697 Mon Sep 17 00:00:00 2001 From: mgiunton Date: Fri, 28 Aug 2026 13:33:32 +0100 Subject: [PATCH 5/6] Remove CLIRN to fix macOS build --- intercept/src/intercept.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/intercept/src/intercept.cpp b/intercept/src/intercept.cpp index 185d0533..705e1f47 100644 --- a/intercept/src/intercept.cpp +++ b/intercept/src/intercept.cpp @@ -14153,7 +14153,7 @@ bool CLIntercept::initDispatch( void ) #define INIT_NATIVE_LAYER_FUNC(funcname) \ { \ void** pfunc = (void**)( &layer_dispatch . funcname ); \ - *pfunc = reinterpret_cast(CLIRN(funcname)); \ + *pfunc = reinterpret_cast(funcname); \ } void CLIntercept::initLayer(cl_icd_dispatch& layer_dispatch, const cl_icd_dispatch* target_dispatch) From 1cd4ea6c34410478d7c5fbdee768165ddb991cc0 Mon Sep 17 00:00:00 2001 From: mgiunton Date: Mon, 31 Aug 2026 14:54:29 +0100 Subject: [PATCH 6/6] Reorganize initLayer extension entries --- intercept/src/intercept.cpp | 51 +++++++++++++++++++++++++++---------- 1 file changed, 37 insertions(+), 14 deletions(-) diff --git a/intercept/src/intercept.cpp b/intercept/src/intercept.cpp index 705e1f47..b1b46800 100644 --- a/intercept/src/intercept.cpp +++ b/intercept/src/intercept.cpp @@ -14226,24 +14226,49 @@ void CLIntercept::initLayer(cl_icd_dispatch& layer_dispatch, const cl_icd_dispat INIT_NATIVE_LAYER_FUNC( clEnqueueMarker ); INIT_NATIVE_LAYER_FUNC( clEnqueueWaitForEvents ); INIT_NATIVE_LAYER_FUNC( clEnqueueBarrier ); + INIT_NATIVE_LAYER_FUNC( clGetExtensionFunctionAddress ); + + // cl_khr_gl_sharing (optional) + // The entry points for this extension are exported from the ICD + // loader even though they are extension APIs. + INIT_NATIVE_LAYER_FUNC( clGetGLContextInfoKHR ); + +#if defined (_WIN32) || defined (__linux__) || defined (__FreeBSD__) INIT_NATIVE_LAYER_FUNC( clCreateFromGLBuffer ); INIT_NATIVE_LAYER_FUNC( clCreateFromGLTexture2D ); INIT_NATIVE_LAYER_FUNC( clCreateFromGLTexture3D ); INIT_NATIVE_LAYER_FUNC( clCreateFromGLRenderbuffer ); INIT_NATIVE_LAYER_FUNC( clGetGLObjectInfo ); - INIT_NATIVE_LAYER_FUNC( clGetGLTextureInfo ); + INIT_NATIVE_LAYER_FUNC( clGetGLTextureInfo ); // OpenCL 1.2 INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireGLObjects ); INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseGLObjects ); - INIT_NATIVE_LAYER_FUNC( clGetGLContextInfoKHR ); +#endif + #if defined(_WIN32) + // cl_khr_d3d10_sharing INIT_NATIVE_LAYER_FUNC( clGetDeviceIDsFromD3D10KHR ); INIT_NATIVE_LAYER_FUNC( clCreateFromD3D10BufferKHR ); INIT_NATIVE_LAYER_FUNC( clCreateFromD3D10Texture2DKHR ); INIT_NATIVE_LAYER_FUNC( clCreateFromD3D10Texture3DKHR ); INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireD3D10ObjectsKHR ); INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseD3D10ObjectsKHR ); + + // cl_khr_d3d11_sharing + INIT_NATIVE_LAYER_FUNC( clGetDeviceIDsFromD3D11KHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromD3D11BufferKHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromD3D11Texture2DKHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromD3D11Texture3DKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireD3D11ObjectsKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseD3D11ObjectsKHR ); + + // cl_khr_dx9_media_sharing + INIT_NATIVE_LAYER_FUNC( clGetDeviceIDsFromDX9MediaAdapterKHR ); + INIT_NATIVE_LAYER_FUNC( clCreateFromDX9MediaSurfaceKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireDX9MediaSurfacesKHR ); + INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseDX9MediaSurfacesKHR ); #endif + INIT_NATIVE_LAYER_FUNC( clSetEventCallback ); INIT_NATIVE_LAYER_FUNC( clCreateSubBuffer ); INIT_NATIVE_LAYER_FUNC( clSetMemObjectDestructorCallback ); @@ -14269,18 +14294,9 @@ void CLIntercept::initLayer(cl_icd_dispatch& layer_dispatch, const cl_icd_dispat INIT_NATIVE_LAYER_FUNC( clEnqueueBarrierWithWaitList ); INIT_NATIVE_LAYER_FUNC( clGetExtensionFunctionAddressForPlatform ); INIT_NATIVE_LAYER_FUNC( clCreateFromGLTexture ); -#if defined(_WIN32) - INIT_NATIVE_LAYER_FUNC( clGetDeviceIDsFromD3D11KHR ); - INIT_NATIVE_LAYER_FUNC( clCreateFromD3D11BufferKHR ); - INIT_NATIVE_LAYER_FUNC( clCreateFromD3D11Texture2DKHR ); - INIT_NATIVE_LAYER_FUNC( clCreateFromD3D11Texture3DKHR ); - INIT_NATIVE_LAYER_FUNC( clCreateFromDX9MediaSurfaceKHR ); - INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireD3D11ObjectsKHR ); - INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseD3D11ObjectsKHR ); - INIT_NATIVE_LAYER_FUNC( clGetDeviceIDsFromDX9MediaAdapterKHR ); - INIT_NATIVE_LAYER_FUNC( clEnqueueAcquireDX9MediaSurfacesKHR ); - INIT_NATIVE_LAYER_FUNC( clEnqueueReleaseDX9MediaSurfacesKHR ); -#endif + +#if !defined(__APPLE__) + // OpenCL 2.0 Entry Points (optional) INIT_NATIVE_LAYER_FUNC( clCreateCommandQueueWithProperties ); INIT_NATIVE_LAYER_FUNC( clCreatePipe ); INIT_NATIVE_LAYER_FUNC( clGetPipeInfo ); @@ -14295,6 +14311,8 @@ void CLIntercept::initLayer(cl_icd_dispatch& layer_dispatch, const cl_icd_dispat INIT_NATIVE_LAYER_FUNC( clSetKernelArgSVMPointer ); INIT_NATIVE_LAYER_FUNC( clSetKernelExecInfo ); INIT_NATIVE_LAYER_FUNC( clGetKernelSubGroupInfoKHR ); + + // OpenCL 2.1 Entry Points (optional) INIT_NATIVE_LAYER_FUNC( clCloneKernel ); INIT_NATIVE_LAYER_FUNC( clCreateProgramWithIL ); INIT_NATIVE_LAYER_FUNC( clEnqueueSVMMigrateMem ); @@ -14302,11 +14320,16 @@ void CLIntercept::initLayer(cl_icd_dispatch& layer_dispatch, const cl_icd_dispat INIT_NATIVE_LAYER_FUNC( clGetHostTimer ); INIT_NATIVE_LAYER_FUNC( clGetKernelSubGroupInfo ); INIT_NATIVE_LAYER_FUNC( clSetDefaultDeviceCommandQueue ); + + // OpenCL 2.2 Entry Points (optional) INIT_NATIVE_LAYER_FUNC( clSetProgramReleaseCallback ); INIT_NATIVE_LAYER_FUNC( clSetProgramSpecializationConstant ); + + // OpenCL 3.0 Entry Points (optional) INIT_NATIVE_LAYER_FUNC( clCreateBufferWithProperties ); INIT_NATIVE_LAYER_FUNC( clCreateImageWithProperties ); INIT_NATIVE_LAYER_FUNC( clSetContextDestructorCallback ); +#endif } ///////////////////////////////////////////////////////////////////////////////