opencl-private.h revision c062b6cf5680afdf8024bad74e563e15d99f3fac 
1/*
2Copyright 1999-2016 ImageMagick Studio LLC, a non-profit organization
3dedicated to making software imaging solutions freely available.
4
5You may not use this file except in compliance with the License.
6obtain a copy of the License at
7
8http://www.imagemagick.org/script/license.php
9
10Unless required by applicable law or agreed to in writing, software
11distributed under the License is distributed on an "AS IS" BASIS,
12WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13See the License for the specific language governing permissions and
14limitations under the License.
15
16MagickCore OpenCL private methods.
17*/
18#ifndef _MAGICKCORE_OPENCL_PRIVATE_H
19#define _MAGICKCORE_OPENCL_PRIVATE_H
20
21/*
22Include declarations.
23*/
24#include "MagickCore/studio.h"
25#include "MagickCore/opencl.h"
26
27#if defined(__cplusplus) || defined(c_plusplus)
28extern "C" {
29#endif
30
31#if !defined(MAGICKCORE_OPENCL_SUPPORT)
32  typedef void* cl_context;
33  typedef void* cl_command_queue;
34  typedef void* cl_device_id;
35  typedef void* cl_event;
36  typedef void* cl_kernel;
37  typedef void* cl_mem;
38  typedef void* cl_platform_id;
39  typedef struct { unsigned char t[8]; } cl_device_type; /* 64-bit */
40#else
41
42#define MAX_COMMAND_QUEUES 16
43
44/*
45 *
46 * function pointer typedefs
47 *
48 */
49
50/* Platform APIs */
51typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetPlatformIDs)(
52                 cl_uint          num_entries,
53                 cl_platform_id * platforms,
54                 cl_uint *        num_platforms) CL_API_SUFFIX__VERSION_1_0;
55
56typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetPlatformInfo)(
57    cl_platform_id   platform,
58    cl_platform_info param_name,
59    size_t           param_value_size,
60    void *           param_value,
61    size_t *         param_value_size_ret) CL_API_SUFFIX__VERSION_1_0;
62
63/* Device APIs */
64typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetDeviceIDs)(
65    cl_platform_id   platform,
66    cl_device_type   device_type,
67    cl_uint          num_entries,
68    cl_device_id *   devices,
69    cl_uint *        num_devices) CL_API_SUFFIX__VERSION_1_0;
70
71typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetDeviceInfo)(
72    cl_device_id    device,
73    cl_device_info  param_name,
74    size_t          param_value_size,
75    void *          param_value,
76    size_t *        param_value_size_ret) CL_API_SUFFIX__VERSION_1_0;
77
78/* Context APIs */
79typedef CL_API_ENTRY cl_context (CL_API_CALL *MAGICKpfn_clCreateContext)(
80    const cl_context_properties * properties,
81    cl_uint                 num_devices,
82    const cl_device_id *    devices,
83    void (CL_CALLBACK *pfn_notify)(const char *, const void *, size_t, void *),
84    void *                  user_data,
85    cl_int *                errcode_ret) CL_API_SUFFIX__VERSION_1_0;
86
87typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseContext)(
88    cl_context context) CL_API_SUFFIX__VERSION_1_0;
89
90/* Command Queue APIs */
91typedef CL_API_ENTRY cl_command_queue (CL_API_CALL *MAGICKpfn_clCreateCommandQueue)(
92    cl_context                     context,
93    cl_device_id                   device,
94    cl_command_queue_properties    properties,
95    cl_int *                       errcode_ret) CL_API_SUFFIX__VERSION_1_0;
96
97typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseCommandQueue)(
98    cl_command_queue command_queue) CL_API_SUFFIX__VERSION_1_0;
99
100/* Memory Object APIs */
101typedef CL_API_ENTRY cl_mem (CL_API_CALL *MAGICKpfn_clCreateBuffer)(
102    cl_context   context,
103    cl_mem_flags flags,
104    size_t       size,
105    void *       host_ptr,
106    cl_int *     errcode_ret) CL_API_SUFFIX__VERSION_1_0;
107
108typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseMemObject)(cl_mem memobj) CL_API_SUFFIX__VERSION_1_0;
109
110/* Program Object APIs */
111typedef CL_API_ENTRY cl_program (CL_API_CALL *MAGICKpfn_clCreateProgramWithSource)(
112    cl_context        context,
113    cl_uint           count,
114    const char **     strings,
115    const size_t *    lengths,
116    cl_int *          errcode_ret) CL_API_SUFFIX__VERSION_1_0;
117
118typedef CL_API_ENTRY cl_program (CL_API_CALL *MAGICKpfn_clCreateProgramWithBinary)(
119    cl_context                     context,
120    cl_uint                        num_devices,
121    const cl_device_id *           device_list,
122    const size_t *                 lengths,
123    const unsigned char **         binaries,
124    cl_int *                       binary_status,
125    cl_int *                       errcode_ret) CL_API_SUFFIX__VERSION_1_0;
126
127typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseProgram)(cl_program program) CL_API_SUFFIX__VERSION_1_0;
128
129typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clBuildProgram)(
130    cl_program           program,
131    cl_uint              num_devices,
132    const cl_device_id * device_list,
133    const char *         options,
134    void (CL_CALLBACK *pfn_notify)(cl_program program, void * user_data),
135    void *               user_data) CL_API_SUFFIX__VERSION_1_0;
136
137typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetProgramInfo)(
138    cl_program         program,
139    cl_program_info    param_name,
140    size_t             param_value_size,
141    void *             param_value,
142    size_t *           param_value_size_ret) CL_API_SUFFIX__VERSION_1_0;
143
144typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clGetProgramBuildInfo)(
145    cl_program            program,
146    cl_device_id          device,
147    cl_program_build_info param_name,
148    size_t                param_value_size,
149    void *                param_value,
150    size_t *              param_value_size_ret) CL_API_SUFFIX__VERSION_1_0;
151
152/* Kernel Object APIs */
153typedef CL_API_ENTRY cl_kernel (CL_API_CALL *MAGICKpfn_clCreateKernel)(
154    cl_program      program,
155    const char *    kernel_name,
156    cl_int *        errcode_ret) CL_API_SUFFIX__VERSION_1_0;
157
158typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clReleaseKernel)(cl_kernel   kernel) CL_API_SUFFIX__VERSION_1_0;
159
160typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clSetKernelArg)(
161    cl_kernel    kernel,
162    cl_uint      arg_index,
163    size_t       arg_size,
164    const void * arg_value) CL_API_SUFFIX__VERSION_1_0;
165
166/* Flush and Finish APIs */
167typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clFlush)(cl_command_queue command_queue) CL_API_SUFFIX__VERSION_1_0;
168
169typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clFinish)(cl_command_queue command_queue) CL_API_SUFFIX__VERSION_1_0;
170
171/* Enqueued Commands APIs */
172typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clEnqueueReadBuffer)(
173    cl_command_queue    command_queue,
174    cl_mem              buffer,
175    cl_bool             blocking_read,
176    size_t              offset,
177    size_t              cb,
178    void *              ptr,
179    cl_uint             num_events_in_wait_list,
180    const cl_event *    event_wait_list,
181    cl_event *          event) CL_API_SUFFIX__VERSION_1_0;
182
183typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clEnqueueWriteBuffer)(
184    cl_command_queue   command_queue,
185    cl_mem             buffer,
186    cl_bool            blocking_write,
187    size_t             offset,
188    size_t             cb,
189    const void *       ptr,
190    cl_uint            num_events_in_wait_list,
191    const cl_event *   event_wait_list,
192    cl_event *         event) CL_API_SUFFIX__VERSION_1_0;
193
194typedef CL_API_ENTRY void * (CL_API_CALL *MAGICKpfn_clEnqueueMapBuffer)(
195    cl_command_queue command_queue,
196    cl_mem           buffer,
197    cl_bool          blocking_map,
198    cl_map_flags     map_flags,
199    size_t           offset,
200    size_t           cb,
201    cl_uint          num_events_in_wait_list,
202    const cl_event * event_wait_list,
203    cl_event *       event,
204    cl_int *         errcode_ret) CL_API_SUFFIX__VERSION_1_0;
205
206typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clEnqueueUnmapMemObject)(
207    cl_command_queue command_queue,
208    cl_mem           memobj,
209    void *           mapped_ptr,
210    cl_uint          num_events_in_wait_list,
211    const cl_event *  event_wait_list,
212    cl_event *        event) CL_API_SUFFIX__VERSION_1_0;
213
214typedef CL_API_ENTRY cl_int (CL_API_CALL *MAGICKpfn_clEnqueueNDRangeKernel)(
215    cl_command_queue command_queue,
216    cl_kernel        kernel,
217    cl_uint          work_dim,
218    const size_t *   global_work_offset,
219    const size_t *   global_work_size,
220    const size_t *   local_work_size,
221    cl_uint          num_events_in_wait_list,
222    const cl_event * event_wait_list,
223    cl_event *       event) CL_API_SUFFIX__VERSION_1_0;
224
225typedef CL_API_ENTRY cl_int(CL_API_CALL *MAGICKpfn_clGetEventProfilingInfo)(
226    cl_event event,
227    cl_profiling_info param_name,
228    size_t param_value_size,
229    void *param_value,
230    size_t *param_value_size_ret) CL_API_SUFFIX__VERSION_1_0;
231
232typedef CL_API_ENTRY cl_int(CL_API_CALL *MAGICKpfn_clWaitForEvents)(
233    cl_uint num_events,
234    const cl_event *event_list) CL_API_SUFFIX__VERSION_1_0;
235
236typedef CL_API_ENTRY cl_int(CL_API_CALL *MAGICKpfn_clReleaseEvent)(
237    cl_event event) CL_API_SUFFIX__VERSION_1_0;
238
239/*
240 *
241 * vendor dispatch table structure
242 *
243 * note that the types in the structure KHRicdVendorDispatch mirror the function
244 * names listed in the string table khrIcdVendorDispatchFunctionNames
245 *
246 */
247
248typedef struct MagickLibraryRec MagickLibrary;
249
250struct MagickLibraryRec
251{
252  void * base;
253
254  MAGICKpfn_clGetPlatformIDs                         clGetPlatformIDs;
255  MAGICKpfn_clGetPlatformInfo                        clGetPlatformInfo;
256  MAGICKpfn_clGetDeviceIDs                           clGetDeviceIDs;
257  MAGICKpfn_clGetDeviceInfo                          clGetDeviceInfo;
258  MAGICKpfn_clCreateContext                          clCreateContext;
259  MAGICKpfn_clCreateCommandQueue                     clCreateCommandQueue;
260  MAGICKpfn_clReleaseCommandQueue                    clReleaseCommandQueue;
261  MAGICKpfn_clCreateBuffer                           clCreateBuffer;
262  MAGICKpfn_clReleaseMemObject                       clReleaseMemObject;
263  MAGICKpfn_clCreateProgramWithSource                clCreateProgramWithSource;
264  MAGICKpfn_clCreateProgramWithBinary                clCreateProgramWithBinary;
265  MAGICKpfn_clReleaseProgram                         clReleaseProgram;
266  MAGICKpfn_clBuildProgram                           clBuildProgram;
267  MAGICKpfn_clGetProgramInfo                         clGetProgramInfo;
268  MAGICKpfn_clGetProgramBuildInfo                    clGetProgramBuildInfo;
269  MAGICKpfn_clCreateKernel                           clCreateKernel;
270  MAGICKpfn_clReleaseKernel                          clReleaseKernel;
271  MAGICKpfn_clSetKernelArg                           clSetKernelArg;
272  MAGICKpfn_clFlush                                  clFlush;
273  MAGICKpfn_clFinish                                 clFinish;
274  MAGICKpfn_clEnqueueReadBuffer                      clEnqueueReadBuffer;
275  MAGICKpfn_clEnqueueWriteBuffer                     clEnqueueWriteBuffer;
276  MAGICKpfn_clEnqueueMapBuffer                       clEnqueueMapBuffer;
277  MAGICKpfn_clEnqueueUnmapMemObject                  clEnqueueUnmapMemObject;
278  MAGICKpfn_clEnqueueNDRangeKernel                   clEnqueueNDRangeKernel;
279  MAGICKpfn_clGetEventProfilingInfo                  clGetEventProfilingInfo;
280  MAGICKpfn_clWaitForEvents                          clWaitForEvents;
281  MAGICKpfn_clReleaseEvent                           clReleaseEvent;
282};
283
284struct _MagickCLEnv {
285  MagickBooleanType OpenCLInitialized;  /* whether OpenCL environment is initialized. */
286  MagickBooleanType OpenCLDisabled;	/* whether if OpenCL has been explicitely disabled. */
287
288  MagickLibrary * library;
289
290  /*OpenCL objects */
291  cl_platform_id platform;
292  cl_device_type deviceType;
293  cl_device_id device;
294  cl_context context;
295
296  MagickBooleanType disableProgramCache; /* disable the OpenCL program cache */
297  cl_program programs[MAGICK_OPENCL_NUM_PROGRAMS]; /* one program object maps one kernel source file */
298
299  MagickBooleanType regenerateProfile;   /* re-run the microbenchmark in auto device selection mode */
300
301  /* Random number generator seeds */
302  unsigned int numGenerators;
303  float randNormalize;
304  cl_mem seeds;
305  SemaphoreInfo* seedsLock;
306
307  SemaphoreInfo* lock;
308
309  cl_command_queue commandQueues[MAX_COMMAND_QUEUES];
310  ssize_t commandQueuesPos;
311  SemaphoreInfo* commandQueuesLock;
312};
313
314#endif
315
316#if defined(MAGICKCORE_HDRI_SUPPORT)
317#define CLOptions "-cl-single-precision-constant -cl-mad-enable -DMAGICKCORE_HDRI_SUPPORT=1 "\
318  "-DCLQuantum=float -DCLSignedQuantum=float -DCLPixelType=float4 -DQuantumRange=%f " \
319  "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\
320  " -DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
321#define CLPixelPacket  cl_float4
322#define CLCharQuantumScale 1.0f
323#elif (MAGICKCORE_QUANTUM_DEPTH == 8)
324#define CLOptions "-cl-single-precision-constant -cl-mad-enable " \
325  "-DCLQuantum=uchar -DCLSignedQuantum=char -DCLPixelType=uchar4 -DQuantumRange=%ff " \
326  "-DQuantumScale=%ff -DCharQuantumScale=%ff -DMagickEpsilon=%ff -DMagickPI=%ff "\
327  "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
328#define CLPixelPacket  cl_uchar4
329#define CLCharQuantumScale 1.0f
330#elif (MAGICKCORE_QUANTUM_DEPTH == 16)
331#define CLOptions "-cl-single-precision-constant -cl-mad-enable " \
332  "-DCLQuantum=ushort -DCLSignedQuantum=short -DCLPixelType=ushort4 -DQuantumRange=%f "\
333  "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\
334  "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
335#define CLPixelPacket  cl_ushort4
336#define CLCharQuantumScale 257.0f
337#elif (MAGICKCORE_QUANTUM_DEPTH == 32)
338#define CLOptions "-cl-single-precision-constant -cl-mad-enable " \
339  "-DCLQuantum=uint -DCLSignedQuantum=int -DCLPixelType=uint4 -DQuantumRange=%f "\
340  "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\
341  "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
342#define CLPixelPacket  cl_uint4
343#define CLCharQuantumScale 16843009.0f
344#elif (MAGICKCORE_QUANTUM_DEPTH == 64)
345#define CLOptions "-cl-single-precision-constant -cl-mad-enable " \
346  "-DCLQuantum=ulong -DCLSignedQuantum=long -DCLPixelType=ulong4 -DQuantumRange=%f "\
347  "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\
348  "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u"
349#define CLPixelPacket  cl_ulong4
350#define CLCharQuantumScale 72340172838076673.0f
351#endif
352
353typedef enum {
354  AddNoiseKernel,
355  BlurRowKernel,
356  BlurColumnKernel,
357  CompositeKernel,
358  ComputeFunctionKernel,
359  ContrastKernel,
360  ContrastStretchKernel,
361  ConvolveKernel,
362  EqualizeKernel,
363  GrayScaleKernel,
364  HistogramKernel,
365  HullPass1Kernel,
366  HullPass2Kernel,
367  LocalContrastBlurRowKernel,
368  LocalContrastBlurApplyColumnKernel,
369  ModulateKernel,
370  MotionBlurKernel,
371  RandomNumberGeneratorKernel,
372  ResizeHorizontalKernel,
373  ResizeVerticalKernel,
374  RotationalBlurKernel,
375  UnsharpMaskBlurColumnKernel,
376  UnsharpMaskKernel,
377  WaveletDenoiseKernel,
378  KERNEL_COUNT
379} ProfiledKernels;
380
381extern MagickPrivate cl_context
382  GetOpenCLContext(MagickCLEnv);
383
384extern MagickPrivate cl_kernel
385  AcquireOpenCLKernel(MagickCLEnv, MagickOpenCLProgram, const char*);
386
387extern MagickPrivate cl_command_queue
388  AcquireOpenCLCommandQueue(MagickCLEnv);
389
390extern MagickPrivate MagickBooleanType
391  OpenCLThrowMagickException(ExceptionInfo *,
392    const char *,const char *,const size_t,
393    const ExceptionType,const char *,const char *,...),
394  RelinquishOpenCLCommandQueue(MagickCLEnv, cl_command_queue),
395  RelinquishOpenCLKernel(MagickCLEnv, cl_kernel);
396
397extern MagickPrivate unsigned long
398  GetOpenCLDeviceLocalMemorySize(MagickCLEnv),
399  GetOpenCLDeviceMaxMemAllocSize(MagickCLEnv);
400
401extern MagickPrivate const char*
402  GetOpenCLCachedFilesDirectory();
403
404extern MagickPrivate void
405  OpenCLLog(const char*),
406  UnlockRandSeedBuffer(MagickCLEnv);
407
408extern MagickPrivate cl_mem
409  GetAndLockRandSeedBuffer(MagickCLEnv);
410
411extern MagickPrivate unsigned int
412  GetNumRandGenerators(MagickCLEnv);
413
414extern MagickPrivate float
415  GetRandNormalize(MagickCLEnv);
416
417extern MagickPrivate void
418  OpenCLTerminus(),
419  RecordProfileData(MagickCLEnv,ProfiledKernels,cl_event);
420
421/* #define OPENCLLOG_ENABLED 1 */
422static inline void OpenCLLogException(const char* function,
423                        const unsigned int line,
424                        ExceptionInfo* exception) {
425#ifdef OPENCLLOG_ENABLED
426  if (exception->severity!=0) {
427    char message[MagickPathExtent];
428    /*  dump the source into a file */
429    (void) FormatLocaleString(message,MagickPathExtent,"%s:%d Exception(%d):%s "
430        ,function,line,exception->severity,exception->reason);
431    OpenCLLog(message);
432  }
433#else
434  magick_unreferenced(function);
435  magick_unreferenced(line);
436  magick_unreferenced(exception);
437#endif
438}
439
440
441#if defined(__cplusplus) || defined(c_plusplus)
442}
443#endif
444
445#endif
446