2 Copyright 1999-2014 ImageMagick Studio LLC, a non-profit organization
3 dedicated to making software imaging solutions freely available.
5 You may not use this file except in compliance with the License.
6 obtain a copy of the License at
8 http://www.imagemagick.org/script/license.php
10 Unless required by applicable law or agreed to in writing, software
11 distributed under the License is distributed on an "AS IS" BASIS,
12 WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13 See the License for the specific language governing permissions and
14 limitations under the License.
16 MagickCore OpenCL private methods.
18 #ifndef _MAGICKCORE_OPENCL_PRIVATE_H
19 #define _MAGICKCORE_OPENCL_PRIVATE_H
24 #include "MagickCore/studio.h"
25 #include "MagickCore/opencl.h"
27 #if defined(__cplusplus) || defined(c_plusplus)
31 #if !defined(MAGICKCORE_OPENCL_SUPPORT)
32 typedef void* cl_platform_id;
33 typedef void* cl_device_id;
34 typedef void* cl_context;
35 typedef void* cl_command_queue;
36 typedef void* cl_kernel;
38 typedef struct { unsigned char t[8]; } cl_device_type; /* 64-bit */
42 * function pointer typedefs
47 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetPlatformIDs)(
49 cl_platform_id * platforms,
50 cl_uint * num_platforms) CL_API_SUFFIX__VERSION_1_0;
52 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetPlatformInfo)(
53 cl_platform_id platform,
54 cl_platform_info param_name,
55 size_t param_value_size,
57 size_t * param_value_size_ret) CL_API_SUFFIX__VERSION_1_0;
60 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetDeviceIDs)(
61 cl_platform_id platform,
62 cl_device_type device_type,
64 cl_device_id * devices,
65 cl_uint * num_devices) CL_API_SUFFIX__VERSION_1_0;
67 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetDeviceInfo)(
69 cl_device_info param_name,
70 size_t param_value_size,
72 size_t * param_value_size_ret) CL_API_SUFFIX__VERSION_1_0;
75 typedef CL_API_ENTRY cl_context (CL_API_CALL *MAGICKpfn_clCreateContext)(
76 const cl_context_properties * properties,
78 const cl_device_id * devices,
79 void (CL_CALLBACK *pfn_notify)(const char *, const void *, size_t, void *),
81 cl_int * errcode_ret) CL_API_SUFFIX__VERSION_1_0;
83 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseContext)(
84 cl_context context) CL_API_SUFFIX__VERSION_1_0;
86 /* Command Queue APIs */
87 typedef CL_API_ENTRY cl_command_queue (CL_API_CALL *MAGICKpfn_clCreateCommandQueue)(
90 cl_command_queue_properties properties,
91 cl_int * errcode_ret) CL_API_SUFFIX__VERSION_1_0;
93 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseCommandQueue)(
94 cl_command_queue command_queue) CL_API_SUFFIX__VERSION_1_0;
96 /* Memory Object APIs */
97 typedef CL_API_ENTRY cl_mem (CL_API_CALL *MAGICKpfn_clCreateBuffer)(
102 cl_int * errcode_ret) CL_API_SUFFIX__VERSION_1_0;
104 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseMemObject)(cl_mem memobj) CL_API_SUFFIX__VERSION_1_0;
106 /* Program Object APIs */
107 typedef CL_API_ENTRY cl_program (CL_API_CALL *MAGICKpfn_clCreateProgramWithSource)(
110 const char ** strings,
111 const size_t * lengths,
112 cl_int * errcode_ret) CL_API_SUFFIX__VERSION_1_0;
114 typedef CL_API_ENTRY cl_program (CL_API_CALL *MAGICKpfn_clCreateProgramWithBinary)(
117 const cl_device_id * device_list,
118 const size_t * lengths,
119 const unsigned char ** binaries,
120 cl_int * binary_status,
121 cl_int * errcode_ret) CL_API_SUFFIX__VERSION_1_0;
123 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseProgram)(cl_program program) CL_API_SUFFIX__VERSION_1_0;
125 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clBuildProgram)(
128 const cl_device_id * device_list,
129 const char * options,
130 void (CL_CALLBACK *pfn_notify)(cl_program program, void * user_data),
131 void * user_data) CL_API_SUFFIX__VERSION_1_0;
133 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetProgramInfo)(
135 cl_program_info param_name,
136 size_t param_value_size,
138 size_t * param_value_size_ret) CL_API_SUFFIX__VERSION_1_0;
140 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetProgramBuildInfo)(
143 cl_program_build_info param_name,
144 size_t param_value_size,
146 size_t * param_value_size_ret) CL_API_SUFFIX__VERSION_1_0;
148 /* Kernel Object APIs */
149 typedef CL_API_ENTRY cl_kernel (CL_API_CALL *MAGICKpfn_clCreateKernel)(
151 const char * kernel_name,
152 cl_int * errcode_ret) CL_API_SUFFIX__VERSION_1_0;
154 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseKernel)(cl_kernel kernel) CL_API_SUFFIX__VERSION_1_0;
156 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clSetKernelArg)(
160 const void * arg_value) CL_API_SUFFIX__VERSION_1_0;
162 /* Flush and Finish APIs */
163 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clFlush)(cl_command_queue command_queue) CL_API_SUFFIX__VERSION_1_0;
165 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clFinish)(cl_command_queue command_queue) CL_API_SUFFIX__VERSION_1_0;
167 /* Enqueued Commands APIs */
168 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clEnqueueReadBuffer)(
169 cl_command_queue command_queue,
171 cl_bool blocking_read,
175 cl_uint num_events_in_wait_list,
176 const cl_event * event_wait_list,
177 cl_event * event) CL_API_SUFFIX__VERSION_1_0;
179 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clEnqueueWriteBuffer)(
180 cl_command_queue command_queue,
182 cl_bool blocking_write,
186 cl_uint num_events_in_wait_list,
187 const cl_event * event_wait_list,
188 cl_event * event) CL_API_SUFFIX__VERSION_1_0;
190 typedef CL_API_ENTRY void * (CL_API_CALL *MAGICKpfn_clEnqueueMapBuffer)(
191 cl_command_queue command_queue,
193 cl_bool blocking_map,
194 cl_map_flags map_flags,
197 cl_uint num_events_in_wait_list,
198 const cl_event * event_wait_list,
200 cl_int * errcode_ret) CL_API_SUFFIX__VERSION_1_0;
202 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clEnqueueUnmapMemObject)(
203 cl_command_queue command_queue,
206 cl_uint num_events_in_wait_list,
207 const cl_event * event_wait_list,
208 cl_event * event) CL_API_SUFFIX__VERSION_1_0;
210 typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clEnqueueNDRangeKernel)(
211 cl_command_queue command_queue,
214 const size_t * global_work_offset,
215 const size_t * global_work_size,
216 const size_t * local_work_size,
217 cl_uint num_events_in_wait_list,
218 const cl_event * event_wait_list,
219 cl_event * event) CL_API_SUFFIX__VERSION_1_0;
223 * vendor dispatch table structure
225 * note that the types in the structure KHRicdVendorDispatch mirror the function
226 * names listed in the string table khrIcdVendorDispatchFunctionNames
230 typedef struct MagickLibraryRec MagickLibrary;
232 struct MagickLibraryRec
236 MAGICKpfn_clGetPlatformIDs clGetPlatformIDs;
237 MAGICKpfn_clGetPlatformInfo clGetPlatformInfo;
238 MAGICKpfn_clGetDeviceIDs clGetDeviceIDs;
239 MAGICKpfn_clGetDeviceInfo clGetDeviceInfo;
240 MAGICKpfn_clCreateContext clCreateContext;
241 MAGICKpfn_clCreateCommandQueue clCreateCommandQueue;
242 MAGICKpfn_clReleaseCommandQueue clReleaseCommandQueue;
243 MAGICKpfn_clCreateBuffer clCreateBuffer;
244 MAGICKpfn_clReleaseMemObject clReleaseMemObject;
245 MAGICKpfn_clCreateProgramWithSource clCreateProgramWithSource;
246 MAGICKpfn_clCreateProgramWithBinary clCreateProgramWithBinary;
247 MAGICKpfn_clReleaseProgram clReleaseProgram;
248 MAGICKpfn_clBuildProgram clBuildProgram;
249 MAGICKpfn_clGetProgramInfo clGetProgramInfo;
250 MAGICKpfn_clGetProgramBuildInfo clGetProgramBuildInfo;
251 MAGICKpfn_clCreateKernel clCreateKernel;
252 MAGICKpfn_clReleaseKernel clReleaseKernel;
253 MAGICKpfn_clSetKernelArg clSetKernelArg;
254 MAGICKpfn_clFlush clFlush;
255 MAGICKpfn_clFinish clFinish;
256 MAGICKpfn_clEnqueueReadBuffer clEnqueueReadBuffer;
257 MAGICKpfn_clEnqueueWriteBuffer clEnqueueWriteBuffer;
258 MAGICKpfn_clEnqueueMapBuffer clEnqueueMapBuffer;
259 MAGICKpfn_clEnqueueUnmapMemObject clEnqueueUnmapMemObject;
260 MAGICKpfn_clEnqueueNDRangeKernel clEnqueueNDRangeKernel;
263 struct _MagickCLEnv {
264 MagickBooleanType OpenCLInitialized; /* whether OpenCL environment is initialized. */
265 MagickBooleanType OpenCLDisabled; /* whether if OpenCL has been explicitely disabled. */
267 MagickLibrary * library;
270 cl_platform_id platform;
271 cl_device_type deviceType;
275 MagickBooleanType disableProgramCache; /* disable the OpenCL program cache */
276 cl_program programs[MAGICK_OPENCL_NUM_PROGRAMS]; /* one program object maps one kernel source file */
278 MagickBooleanType regenerateProfile; /* re-run the microbenchmark in auto device selection mode */
280 /* Random number generator seeds */
281 unsigned int numGenerators;
284 SemaphoreInfo* seedsLock;
291 #if defined(MAGICKCORE_HDRI_SUPPORT)
292 #define CLOptions "-cl-single-precision-constant -cl-mad-enable -DMAGICKCORE_HDRI_SUPPORT=1 "\
293 "-DCLQuantum=float -DCLSignedQuantum=float -DCLPixelType=float4 -DQuantumRange=%f " \
294 "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\
295 " -DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
296 #define CLPixelPacket cl_float4
297 #define CLCharQuantumScale 1.0f
298 #elif (MAGICKCORE_QUANTUM_DEPTH == 8)
299 #define CLOptions "-cl-single-precision-constant -cl-mad-enable " \
300 "-DCLQuantum=uchar -DCLSignedQuantum=char -DCLPixelType=uchar4 -DQuantumRange=%ff " \
301 "-DQuantumScale=%ff -DCharQuantumScale=%ff -DMagickEpsilon=%ff -DMagickPI=%ff "\
302 "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
303 #define CLPixelPacket cl_uchar4
304 #define CLCharQuantumScale 1.0f
305 #elif (MAGICKCORE_QUANTUM_DEPTH == 16)
306 #define CLOptions "-cl-single-precision-constant -cl-mad-enable " \
307 "-DCLQuantum=ushort -DCLSignedQuantum=short -DCLPixelType=ushort4 -DQuantumRange=%f "\
308 "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\
309 "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
310 #define CLPixelPacket cl_ushort4
311 #define CLCharQuantumScale 257.0f
312 #elif (MAGICKCORE_QUANTUM_DEPTH == 32)
313 #define CLOptions "-cl-single-precision-constant -cl-mad-enable " \
314 "-DCLQuantum=uint -DCLSignedQuantum=int -DCLPixelType=uint4 -DQuantumRange=%f "\
315 "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\
316 "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
317 #define CLPixelPacket cl_uint4
318 #define CLCharQuantumScale 16843009.0f
319 #elif (MAGICKCORE_QUANTUM_DEPTH == 64)
320 #define CLOptions "-cl-single-precision-constant -cl-mad-enable " \
321 "-DCLQuantum=ulong -DCLSignedQuantum=long -DCLPixelType=ulong4 -DQuantumRange=%f "\
322 "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\
323 "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
324 #define CLPixelPacket cl_ulong4
325 #define CLCharQuantumScale 72340172838076673.0f
328 extern MagickPrivate cl_context
329 GetOpenCLContext(MagickCLEnv);
331 extern MagickPrivate cl_kernel
332 AcquireOpenCLKernel(MagickCLEnv, MagickOpenCLProgram, const char*);
334 extern MagickPrivate cl_command_queue
335 AcquireOpenCLCommandQueue(MagickCLEnv);
337 extern MagickPrivate MagickBooleanType
338 OpenCLThrowMagickException(ExceptionInfo *,
339 const char *,const char *,const size_t,
340 const ExceptionType,const char *,const char *,...),
341 RelinquishOpenCLCommandQueue(MagickCLEnv, cl_command_queue),
342 RelinquishOpenCLKernel(MagickCLEnv, cl_kernel);
344 extern MagickPrivate unsigned long
345 GetOpenCLDeviceLocalMemorySize(MagickCLEnv),
346 GetOpenCLDeviceMaxMemAllocSize(MagickCLEnv);
348 extern MagickPrivate const char*
349 GetOpenCLCachedFilesDirectory();
351 extern MagickPrivate void
352 UnlockRandSeedBuffer(MagickCLEnv),
353 OpenCLLog(const char*);
355 extern MagickPrivate cl_mem
356 GetAndLockRandSeedBuffer(MagickCLEnv);
358 extern MagickPrivate unsigned int
359 GetNumRandGenerators(MagickCLEnv);
361 extern MagickPrivate float
362 GetRandNormalize(MagickCLEnv clEnv);
364 typedef struct _AccelerateTimer {
371 void startAccelerateTimer(AccelerateTimer* timer);
372 void stopAccelerateTimer(AccelerateTimer* timer);
373 void resetAccelerateTimer(AccelerateTimer* timer);
374 void initAccelerateTimer(AccelerateTimer* timer);
375 double readAccelerateTimer(AccelerateTimer* timer);
377 /* #define OPENCLLOG_ENABLED 1 */
378 static inline void OpenCLLogException(const char* function,
379 const unsigned int line,
380 ExceptionInfo* exception) {
381 #ifdef OPENCLLOG_ENABLED
382 if (exception->severity!=0) {
383 char message[MaxTextExtent];
384 /* dump the source into a file */
385 (void) FormatLocaleString(message,MaxTextExtent,"%s:%d Exception(%d):%s "
386 ,function,line,exception->severity,exception->reason);
390 magick_unreferenced(function);
391 magick_unreferenced(line);
392 magick_unreferenced(exception);
397 #if defined(__cplusplus) || defined(c_plusplus)