1 /* 2 Copyright 1999-2016 ImageMagick Studio LLC, a non-profit organization 3 dedicated to making software imaging solutions freely available. 4 5 You may not use this file except in compliance with the License. 6 obtain a copy of the License at 7 8 http://www.imagemagick.org/script/license.php 9 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. 15 16 MagickCore OpenCL private methods. 17 */ 18 #ifndef MAGICKCORE_OPENCL_PRIVATE_H 19 #define MAGICKCORE_OPENCL_PRIVATE_H 20 21 /* 22 Include declarations. 23 */ 24 #include "MagickCore/studio.h" 25 #include "MagickCore/opencl.h" 26 #include "MagickCore/thread_.h" 27 28 #if defined(__cplusplus) || defined(c_plusplus) 29 extern "C" { 30 #endif 31 32 #if !defined(MAGICKCORE_OPENCL_SUPPORT) 33 typedef void* MagickCLCacheInfo; 34 #else 35 typedef struct _MagickCLCacheInfo 36 { 37 cl_event 38 *events; 39 40 cl_mem 41 buffer; 42 43 cl_uint 44 event_count; 45 46 MagickCLDevice 47 device; 48 49 MagickSizeType 50 length; 51 52 Quantum 53 *pixels; 54 }* MagickCLCacheInfo; 55 56 /* 57 Define declarations. 58 */ 59 #define MAGICKCORE_OPENCL_UNDEFINED_SCORE -1.0 60 #define MAGICKCORE_OPENCL_COMMAND_QUEUES 16 61 62 /* Platform APIs */ 63 typedef CL_API_ENTRY cl_int 64 (CL_API_CALL *MAGICKpfn_clGetPlatformIDs)(cl_uint num_entries, 65 cl_platform_id *platforms,cl_uint *num_platforms) CL_API_SUFFIX__VERSION_1_0; 66 67 typedef CL_API_ENTRY cl_int 68 (CL_API_CALL *MAGICKpfn_clGetPlatformInfo)(cl_platform_id platform, 69 cl_platform_info param_name,size_t param_value_size,void *param_value, 70 size_t *param_value_size_ret) CL_API_SUFFIX__VERSION_1_0; 71 72 73 /* Device APIs */ 74 typedef CL_API_ENTRY cl_int 75 (CL_API_CALL *MAGICKpfn_clGetDeviceIDs)(cl_platform_id platform, 76 cl_device_type device_type,cl_uint num_entries,cl_device_id *devices, 77 cl_uint *num_devices) CL_API_SUFFIX__VERSION_1_0; 78 79 typedef CL_API_ENTRY cl_int 80 (CL_API_CALL *MAGICKpfn_clGetDeviceInfo)(cl_device_id device, 81 cl_device_info param_name,size_t param_value_size,void *param_value, 82 size_t *param_value_size_ret) CL_API_SUFFIX__VERSION_1_0; 83 84 85 /* Context APIs */ 86 typedef CL_API_ENTRY cl_context 87 (CL_API_CALL *MAGICKpfn_clCreateContext)( 88 const cl_context_properties *properties,cl_uint num_devices, 89 const cl_device_id *devices,void (CL_CALLBACK *pfn_notify)(const char *, 90 const void *,size_t,void *),void *user_data,cl_int *errcode_ret) 91 CL_API_SUFFIX__VERSION_1_0; 92 93 typedef CL_API_ENTRY cl_int 94 (CL_API_CALL *MAGICKpfn_clReleaseContext)(cl_context context) 95 CL_API_SUFFIX__VERSION_1_0; 96 97 98 /* Command Queue APIs */ 99 typedef CL_API_ENTRY cl_command_queue 100 (CL_API_CALL *MAGICKpfn_clCreateCommandQueue)(cl_context context, 101 cl_device_id device,cl_command_queue_properties properties, 102 cl_int *errcode_ret) CL_API_SUFFIX__VERSION_1_0; 103 104 typedef CL_API_ENTRY cl_int 105 (CL_API_CALL *MAGICKpfn_clReleaseCommandQueue)( 106 cl_command_queue command_queue) CL_API_SUFFIX__VERSION_1_0; 107 108 typedef CL_API_ENTRY cl_int 109 (CL_API_CALL *MAGICKpfn_clFlush)(cl_command_queue command_queue) 110 CL_API_SUFFIX__VERSION_1_0; 111 112 typedef CL_API_ENTRY cl_int 113 (CL_API_CALL *MAGICKpfn_clFinish)(cl_command_queue command_queue) 114 CL_API_SUFFIX__VERSION_1_0; 115 116 117 /* Memory Object APIs */ 118 typedef CL_API_ENTRY cl_mem 119 (CL_API_CALL *MAGICKpfn_clCreateBuffer)(cl_context context, 120 cl_mem_flags flags,size_t size,void *host_ptr,cl_int *errcode_ret) 121 CL_API_SUFFIX__VERSION_1_0; 122 123 typedef CL_API_ENTRY cl_int 124 (CL_API_CALL *MAGICKpfn_clReleaseMemObject)(cl_mem memobj) 125 CL_API_SUFFIX__VERSION_1_0; 126 127 128 /* Program Object APIs */ 129 typedef CL_API_ENTRY cl_program 130 (CL_API_CALL *MAGICKpfn_clCreateProgramWithSource)(cl_context context, 131 cl_uint count,const char **strings,const size_t *lengths, 132 cl_int *errcode_ret) CL_API_SUFFIX__VERSION_1_0; 133 134 typedef CL_API_ENTRY cl_program 135 (CL_API_CALL *MAGICKpfn_clCreateProgramWithBinary)(cl_context context, 136 cl_uint num_devices,const cl_device_id *device_list,const size_t *lengths, 137 const unsigned char **binaries,cl_int *binary_status,cl_int *errcode_ret) 138 CL_API_SUFFIX__VERSION_1_0; 139 140 typedef CL_API_ENTRY cl_int 141 (CL_API_CALL *MAGICKpfn_clReleaseProgram)(cl_program program) 142 CL_API_SUFFIX__VERSION_1_0; 143 144 typedef CL_API_ENTRY cl_int 145 (CL_API_CALL *MAGICKpfn_clBuildProgram)(cl_program program, 146 cl_uint num_devices,const cl_device_id *device_list,const char *options, 147 void (CL_CALLBACK *pfn_notify)(cl_program program,void * user_data), 148 void *user_data) CL_API_SUFFIX__VERSION_1_0; 149 150 typedef CL_API_ENTRY cl_int 151 (CL_API_CALL *MAGICKpfn_clGetProgramBuildInfo)(cl_program program, 152 cl_device_id device,cl_program_build_info param_name,size_t param_value_size, 153 void *param_value,size_t *param_value_size_ret) CL_API_SUFFIX__VERSION_1_0; 154 155 typedef CL_API_ENTRY cl_int 156 (CL_API_CALL *MAGICKpfn_clGetProgramInfo)(cl_program program, 157 cl_program_info param_name,size_t param_value_size,void *param_value, 158 size_t *param_value_size_ret) CL_API_SUFFIX__VERSION_1_0; 159 160 161 /* Kernel Object APIs */ 162 typedef CL_API_ENTRY cl_kernel 163 (CL_API_CALL *MAGICKpfn_clCreateKernel)(cl_program program, 164 const char *kernel_name,cl_int *errcode_ret) CL_API_SUFFIX__VERSION_1_0; 165 166 typedef CL_API_ENTRY cl_int 167 (CL_API_CALL *MAGICKpfn_clReleaseKernel)(cl_kernel kernel) 168 CL_API_SUFFIX__VERSION_1_0; 169 170 typedef CL_API_ENTRY cl_int 171 (CL_API_CALL *MAGICKpfn_clSetKernelArg)(cl_kernel kernel,cl_uint arg_index, 172 size_t arg_size,const void * arg_value) CL_API_SUFFIX__VERSION_1_0; 173 174 typedef CL_API_ENTRY cl_int 175 (CL_API_CALL *MAGICKpfn_clGetKernelInfo)(cl_kernel kernel, 176 cl_kernel_info param_name,size_t param_value_size,void *param_value, 177 size_t *param_value_size_ret) CL_API_SUFFIX__VERSION_1_0; 178 179 180 /* Enqueued Commands APIs */ 181 typedef CL_API_ENTRY cl_int 182 (CL_API_CALL *MAGICKpfn_clEnqueueReadBuffer)(cl_command_queue command_queue, 183 cl_mem buffer,cl_bool blocking_read,size_t offset,size_t cb,void *ptr, 184 cl_uint num_events_in_wait_list,const cl_event *event_wait_list, 185 cl_event *event) CL_API_SUFFIX__VERSION_1_0; 186 187 typedef CL_API_ENTRY void 188 *(CL_API_CALL *MAGICKpfn_clEnqueueMapBuffer)(cl_command_queue command_queue, 189 cl_mem buffer,cl_bool blocking_map,cl_map_flags map_flags,size_t offset, 190 size_t cb,cl_uint num_events_in_wait_list,const cl_event *event_wait_list, 191 cl_event *event,cl_int *errcode_ret) CL_API_SUFFIX__VERSION_1_0; 192 193 typedef CL_API_ENTRY cl_int 194 (CL_API_CALL *MAGICKpfn_clEnqueueUnmapMemObject)( 195 cl_command_queue command_queue,cl_mem memobj,void *mapped_ptr, 196 cl_uint num_events_in_wait_list,const cl_event *event_wait_list, 197 cl_event *event) CL_API_SUFFIX__VERSION_1_0; 198 199 typedef CL_API_ENTRY cl_int 200 (CL_API_CALL *MAGICKpfn_clEnqueueNDRangeKernel)( 201 cl_command_queue command_queue,cl_kernel kernel,cl_uint work_dim, 202 const size_t *global_work_offset,const size_t *global_work_size, 203 const size_t *local_work_size,cl_uint num_events_in_wait_list, 204 const cl_event * event_wait_list,cl_event *event) 205 CL_API_SUFFIX__VERSION_1_0; 206 207 208 /* Events APIs */ 209 typedef CL_API_ENTRY cl_int 210 (CL_API_CALL *MAGICKpfn_clWaitForEvents)(cl_uint num_events, 211 const cl_event *event_list) CL_API_SUFFIX__VERSION_1_0; 212 213 typedef CL_API_ENTRY cl_int 214 (CL_API_CALL *MAGICKpfn_clReleaseEvent)(cl_event event) 215 CL_API_SUFFIX__VERSION_1_0; 216 217 typedef CL_API_ENTRY cl_int 218 (CL_API_CALL *MAGICKpfn_clRetainEvent)(cl_event event) 219 CL_API_SUFFIX__VERSION_1_0; 220 221 typedef CL_API_ENTRY cl_int 222 (CL_API_CALL *MAGICKpfn_clSetEventCallback)(cl_event event, 223 cl_int command_exec_callback_type,void (CL_CALLBACK *MAGICKpfn_notify)( 224 cl_event,cl_int,void *),void *user_data) CL_API_SUFFIX__VERSION_1_1; 225 226 227 /* Profiling APIs */ 228 typedef CL_API_ENTRY cl_int 229 (CL_API_CALL *MAGICKpfn_clGetEventProfilingInfo)(cl_event event, 230 cl_profiling_info param_name,size_t param_value_size,void *param_value, 231 size_t *param_value_size_ret) CL_API_SUFFIX__VERSION_1_0; 232 233 typedef struct MagickLibraryRec MagickLibrary; 234 235 struct MagickLibraryRec 236 { 237 MAGICKpfn_clGetPlatformIDs clGetPlatformIDs; 238 MAGICKpfn_clGetPlatformInfo clGetPlatformInfo; 239 240 MAGICKpfn_clGetDeviceIDs clGetDeviceIDs; 241 MAGICKpfn_clGetDeviceInfo clGetDeviceInfo; 242 243 MAGICKpfn_clCreateContext clCreateContext; 244 MAGICKpfn_clReleaseContext clReleaseContext; 245 246 MAGICKpfn_clCreateCommandQueue clCreateCommandQueue; 247 MAGICKpfn_clReleaseCommandQueue clReleaseCommandQueue; 248 MAGICKpfn_clFlush clFlush; 249 MAGICKpfn_clFinish clFinish; 250 251 MAGICKpfn_clCreateBuffer clCreateBuffer; 252 MAGICKpfn_clReleaseMemObject clReleaseMemObject; 253 254 MAGICKpfn_clCreateProgramWithSource clCreateProgramWithSource; 255 MAGICKpfn_clCreateProgramWithBinary clCreateProgramWithBinary; 256 MAGICKpfn_clReleaseProgram clReleaseProgram; 257 MAGICKpfn_clBuildProgram clBuildProgram; 258 MAGICKpfn_clGetProgramBuildInfo clGetProgramBuildInfo; 259 MAGICKpfn_clGetProgramInfo clGetProgramInfo; 260 261 MAGICKpfn_clCreateKernel clCreateKernel; 262 MAGICKpfn_clReleaseKernel clReleaseKernel; 263 MAGICKpfn_clSetKernelArg clSetKernelArg; 264 MAGICKpfn_clGetKernelInfo clGetKernelInfo; 265 266 MAGICKpfn_clEnqueueReadBuffer clEnqueueReadBuffer; 267 MAGICKpfn_clEnqueueMapBuffer clEnqueueMapBuffer; 268 MAGICKpfn_clEnqueueUnmapMemObject clEnqueueUnmapMemObject; 269 MAGICKpfn_clEnqueueNDRangeKernel clEnqueueNDRangeKernel; 270 271 MAGICKpfn_clWaitForEvents clWaitForEvents; 272 MAGICKpfn_clReleaseEvent clReleaseEvent; 273 MAGICKpfn_clRetainEvent clRetainEvent; 274 MAGICKpfn_clSetEventCallback clSetEventCallback; 275 276 MAGICKpfn_clGetEventProfilingInfo clGetEventProfilingInfo; 277 }; 278 279 struct _MagickCLDevice 280 { 281 char 282 *name, 283 *platform_name, 284 *version; 285 286 cl_command_queue 287 command_queues[MAGICKCORE_OPENCL_COMMAND_QUEUES]; 288 289 cl_context 290 context; 291 292 cl_device_id 293 deviceID; 294 295 cl_device_type 296 type; 297 298 cl_program 299 program; 300 301 cl_uint 302 max_clock_frequency, 303 max_compute_units; 304 305 cl_ulong 306 local_memory_size; 307 308 double 309 score; 310 311 KernelProfileRecord 312 *profile_records; 313 314 MagickBooleanType 315 enabled, 316 profile_kernels; 317 318 SemaphoreInfo 319 *lock; 320 321 size_t 322 requested; 323 324 ssize_t 325 command_queues_index; 326 }; 327 328 typedef struct _MagickCLEnv 329 { 330 cl_context 331 *contexts; 332 333 double 334 cpu_score; 335 336 MagickBooleanType 337 enabled, 338 initialized; 339 340 MagickCLDevice 341 *devices; 342 343 MagickLibrary 344 *library; 345 346 MagickThreadType 347 benchmark_thread_id; 348 349 SemaphoreInfo 350 *lock; 351 352 size_t 353 number_contexts, 354 number_devices; 355 } *MagickCLEnv; 356 357 #if defined(MAGICKCORE_HDRI_SUPPORT) 358 #define CLOptions "-cl-single-precision-constant -cl-mad-enable -DMAGICKCORE_HDRI_SUPPORT=1 "\ 359 "-DCLQuantum=float -DCLSignedQuantum=float -DCLPixelType=float4 -DQuantumRange=%ff " \ 360 "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\ 361 "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u" 362 #define CLQuantum cl_float 363 #define CLPixelPacket cl_float4 364 #define CLCharQuantumScale 1.0f 365 #elif (MAGICKCORE_QUANTUM_DEPTH == 8) 366 #define CLOptions "-cl-single-precision-constant -cl-mad-enable " \ 367 "-DCLQuantum=uchar -DCLSignedQuantum=char -DCLPixelType=uchar4 -DQuantumRange=%ff " \ 368 "-DQuantumScale=%ff -DCharQuantumScale=%ff -DMagickEpsilon=%ff -DMagickPI=%ff "\ 369 "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u" 370 #define CLQuantum cl_uchar 371 #define CLPixelPacket cl_uchar4 372 #define CLCharQuantumScale 1.0f 373 #elif (MAGICKCORE_QUANTUM_DEPTH == 16) 374 #define CLOptions "-cl-single-precision-constant -cl-mad-enable " \ 375 "-DCLQuantum=ushort -DCLSignedQuantum=short -DCLPixelType=ushort4 -DQuantumRange=%ff "\ 376 "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\ 377 "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u" 378 #define CLQuantum cl_ushort 379 #define CLPixelPacket cl_ushort4 380 #define CLCharQuantumScale 257.0f 381 #elif (MAGICKCORE_QUANTUM_DEPTH == 32) 382 #define CLOptions "-cl-single-precision-constant -cl-mad-enable " \ 383 "-DCLQuantum=uint -DCLSignedQuantum=int -DCLPixelType=uint4 -DQuantumRange=%ff "\ 384 "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\ 385 "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u" 386 #define CLQuantum cl_uint 387 #define CLPixelPacket cl_uint4 388 #define CLCharQuantumScale 16843009.0f 389 #elif (MAGICKCORE_QUANTUM_DEPTH == 64) 390 #define CLOptions "-cl-single-precision-constant -cl-mad-enable " \ 391 "-DCLQuantum=ulong -DCLSignedQuantum=long -DCLPixelType=ulong4 -DQuantumRange=%ff "\ 392 "-DQuantumScale=%f -DCharQuantumScale=%f -DMagickEpsilon=%f -DMagickPI=%f "\ 393 "-DMaxMap=%u -DMAGICKCORE_QUANTUM_DEPTH=%u" 394 #define CLQuantum cl_ulong 395 #define CLPixelPacket cl_ulong4 396 #define CLCharQuantumScale 72340172838076673.0f 397 #endif 398 399 extern MagickPrivate cl_command_queue 400 AcquireOpenCLCommandQueue(MagickCLDevice); 401 402 extern MagickPrivate cl_int 403 SetOpenCLKernelArg(cl_kernel,cl_uint,size_t,const void *); 404 405 extern MagickPrivate cl_kernel 406 AcquireOpenCLKernel(MagickCLDevice,const char *); 407 408 extern MagickPrivate cl_mem 409 CreateOpenCLBuffer(MagickCLDevice,cl_mem_flags,size_t,void *); 410 411 extern MagickPrivate MagickBooleanType 412 EnqueueOpenCLKernel(cl_command_queue,cl_kernel,cl_uint,const size_t *, 413 const size_t *,const size_t *,const Image *,const Image *,ExceptionInfo *), 414 InitializeOpenCL(MagickCLEnv,ExceptionInfo *), 415 OpenCLThrowMagickException(MagickCLDevice,ExceptionInfo *, 416 const char *,const char *,const size_t,const ExceptionType,const char *, 417 const char *,...), 418 RecordProfileData(MagickCLDevice,cl_kernel,cl_event); 419 420 extern MagickPrivate MagickCLCacheInfo 421 AcquireMagickCLCacheInfo(MagickCLDevice,Quantum *,const MagickSizeType), 422 CopyMagickCLCacheInfo(MagickCLCacheInfo), 423 RelinquishMagickCLCacheInfo(MagickCLCacheInfo,const MagickBooleanType); 424 425 extern MagickPrivate MagickCLDevice 426 RequestOpenCLDevice(MagickCLEnv); 427 428 extern MagickPrivate MagickCLEnv 429 GetCurrentOpenCLEnv(void); 430 431 extern MagickPrivate unsigned long 432 GetOpenCLDeviceLocalMemorySize(const MagickCLDevice); 433 434 extern MagickPrivate void 435 DumpOpenCLProfileData(), 436 OpenCLTerminus(), 437 ReleaseOpenCLCommandQueue(MagickCLDevice,cl_command_queue), 438 ReleaseOpenCLDevice(MagickCLDevice), 439 ReleaseOpenCLKernel(cl_kernel), 440 ReleaseOpenCLMemObject(cl_mem), 441 RetainOpenCLEvent(cl_event); 442 443 #endif 444 445 #if defined(__cplusplus) || defined(c_plusplus) 446 } 447 #endif 448 449 #endif 450
