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