43 #include "magick/studio.h"
44 #include "magick/artifact.h"
45 #include "magick/cache.h"
46 #include "magick/cache-private.h"
47 #include "magick/color.h"
48 #include "magick/compare.h"
49 #include "magick/constitute.h"
50 #include "magick/distort.h"
51 #include "magick/draw.h"
52 #include "magick/effect.h"
53 #include "magick/exception.h"
54 #include "magick/exception-private.h"
55 #include "magick/fx.h"
56 #include "magick/gem.h"
57 #include "magick/geometry.h"
58 #include "magick/image.h"
59 #include "magick/image-private.h"
60 #include "magick/layer.h"
61 #include "magick/locale_.h"
62 #include "magick/mime-private.h"
63 #include "magick/memory_.h"
64 #include "magick/memory-private.h"
65 #include "magick/monitor.h"
66 #include "magick/montage.h"
67 #include "magick/morphology.h"
68 #include "magick/nt-base.h"
69 #include "magick/nt-base-private.h"
70 #include "magick/opencl.h"
71 #include "magick/opencl-private.h"
72 #include "magick/option.h"
73 #include "magick/policy.h"
74 #include "magick/property.h"
75 #include "magick/quantize.h"
76 #include "magick/quantum.h"
77 #include "magick/random_.h"
78 #include "magick/random-private.h"
79 #include "magick/resample.h"
80 #include "magick/resource_.h"
81 #include "magick/splay-tree.h"
82 #include "magick/semaphore.h"
83 #include "magick/statistic.h"
84 #include "magick/string_.h"
85 #include "magick/token.h"
86 #include "magick/utility.h"
87 #include "magick/utility-private.h"
89 #ifdef MAGICKCORE_CLPERFMARKER
90 #include "CLPerfMarker.h"
94 #if defined(MAGICKCORE_OPENCL_SUPPORT)
96 #define NUM_CL_RAND_GENERATORS 1024
97 #define PROFILE_OCL_KERNELS 0
105 } KernelProfileRecord;
107 static const char *kernelNames[] = {
121 "LocalContrastBlurRow",
122 "LocalContrastBlurApplyColumn",
126 "RandomNumberGenerator",
129 "UnsharpMaskBlurColumn",
135 profileRecords[KERNEL_COUNT];
137 typedef struct _AccelerateTimer {
143 void startAccelerateTimer(AccelerateTimer* timer) {
145 QueryPerformanceCounter((LARGE_INTEGER*)&timer->_start);
151 timer->_start = (
long long)s.tv_sec * (
long long)1.0E3 + (
long long)s.tv_usec / (
long long)1.0E3;
155 void stopAccelerateTimer(AccelerateTimer* timer) {
158 QueryPerformanceCounter((LARGE_INTEGER*)&(n));
162 n = (
long long)s.tv_sec * (
long long)1.0E3+ (
long long)s.tv_usec / (
long long)1.0E3;
169 void resetAccelerateTimer(AccelerateTimer* timer) {
174 void initAccelerateTimer(AccelerateTimer* timer) {
176 QueryPerformanceFrequency((LARGE_INTEGER*)&timer->_freq);
178 timer->_freq = (
long long)1.0E3;
180 resetAccelerateTimer(timer);
183 double readAccelerateTimer(AccelerateTimer* timer) {
184 return (
double)timer->_clocks/(double)timer->_freq;
187 MagickPrivate MagickBooleanType RecordProfileData(
MagickCLEnv clEnv, ProfiledKernels kernel, cl_event event)
189 #if PROFILE_OCL_KERNELS
193 cl_ulong elapsed = 0;
194 clEnv->library->clWaitForEvents(1, &event);
195 status = clEnv->library->clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_START,
sizeof(cl_ulong), &start, NULL);
196 status &= clEnv->library->clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_END,
sizeof(cl_ulong), &end, NULL);
197 if (status == CL_SUCCESS) {
200 elapsed = end - start;
202 LockSemaphoreInfo(clEnv->commandQueuesLock);
203 if ((elapsed < profileRecords[kernel].min) || (profileRecords[kernel].count == 0))
204 profileRecords[kernel].min = elapsed;
205 if (elapsed > profileRecords[kernel].max)
206 profileRecords[kernel].max = elapsed;
207 profileRecords[kernel].total += elapsed;
208 profileRecords[kernel].count += 1;
209 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
213 magick_unreferenced(clEnv);
214 magick_unreferenced(kernel);
215 magick_unreferenced(event);
220 void DumpProfileData()
222 #if PROFILE_OCL_KERNELS
225 OpenCLLog(
"====================================================");
236 clEnv = GetDefaultOpenCLEnv();
238 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_VENDOR, 2048, buff, NULL);
241 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_NAME, 2048, buff, NULL);
244 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DRIVER_VERSION, 2048, buff, NULL);
248 OpenCLLog(
"====================================================");
249 OpenCLLog(
" ave\tcalls \tmin -> max");
250 OpenCLLog(
" ---\t----- \t----------");
251 for (i = 0; i < KERNEL_COUNT; ++i) {
254 (void) CopyMagickString(indent,
" ",
256 strncpy(indent, kernelNames[i], min(strlen(kernelNames[i]), strlen(indent) - 1));
257 (void) FormatLocaleString(buf,
sizeof(buf),
"%s%d\t(%d calls) \t%d -> %d",
258 indent, profileRecords[i].count > 0 ? (profileRecords[i].total /
259 profileRecords[i].count) : 0, profileRecords[i].count,
260 profileRecords[i].min, profileRecords[i].max);
266 OpenCLLog(
"====================================================");
275 #ifdef MAGICKCORE_WINDOWS_SUPPORT
281 void *OsLibraryLoad(
const char *libraryName)
283 #ifdef MAGICKCORE_WINDOWS_SUPPORT
284 return (
void *)LoadLibraryA(libraryName);
286 return (
void *)dlopen(libraryName, RTLD_NOW);
291 void *OsLibraryGetFunctionAddress(
void *library,
const char *functionName)
293 #ifdef MAGICKCORE_WINDOWS_SUPPORT
294 if (!library || !functionName)
298 return (
void *) GetProcAddress( (HMODULE)library, functionName);
300 if (!library || !functionName)
304 return (
void *)dlsym(library, functionName);
331 clEnv->commandQueuesPos=-1;
332 ActivateSemaphoreInfo(&clEnv->lock);
333 ActivateSemaphoreInfo(&clEnv->commandQueuesLock);
362 MagickPrivate MagickBooleanType RelinquishMagickOpenCLEnv(
MagickCLEnv clEnv)
366 while (clEnv->commandQueuesPos >= 0)
368 clEnv->library->clReleaseCommandQueue(
369 clEnv->commandQueues[clEnv->commandQueuesPos--]);
371 if (clEnv->programs[0] != (cl_program) NULL)
372 (void) clEnv->library->clReleaseProgram(clEnv->programs[0]);
373 if (clEnv->context != (cl_context) NULL)
374 clEnv->library->clReleaseContext(clEnv->context);
375 DestroySemaphoreInfo(&clEnv->lock);
376 DestroySemaphoreInfo(&clEnv->commandQueuesLock);
377 RelinquishMagickMemory(clEnv);
393 MagickLibrary * OpenCLLib;
397 static MagickBooleanType bindOpenCLFunctions(
void* library)
399 #ifdef MAGICKCORE_HAVE_OPENCL_CL_H
400 #define BIND(X) OpenCLLib->X= &X;
403 if ((OpenCLLib->X=(MAGICKpfn_##X)OsLibraryGetFunctionAddress(library,#X)) == NULL)\
407 BIND(clGetPlatformIDs);
408 BIND(clGetPlatformInfo);
410 BIND(clGetDeviceIDs);
411 BIND(clGetDeviceInfo);
413 BIND(clCreateContext);
414 BIND(clReleaseContext);
416 BIND(clCreateBuffer);
417 BIND(clRetainMemObject);
418 BIND(clReleaseMemObject);
420 BIND(clCreateProgramWithSource);
421 BIND(clCreateProgramWithBinary);
422 BIND(clBuildProgram);
423 BIND(clReleaseProgram);
424 BIND(clGetProgramInfo);
425 BIND(clGetProgramBuildInfo);
427 BIND(clCreateKernel);
428 BIND(clReleaseKernel);
429 BIND(clSetKernelArg);
434 BIND(clEnqueueNDRangeKernel);
435 BIND(clEnqueueReadBuffer);
436 BIND(clEnqueueMapBuffer);
437 BIND(clEnqueueUnmapMemObject);
439 BIND(clCreateCommandQueue);
440 BIND(clReleaseCommandQueue);
442 BIND(clGetEventProfilingInfo);
443 BIND(clGetEventInfo);
444 BIND(clWaitForEvents);
445 BIND(clReleaseEvent);
447 BIND(clSetEventCallback);
452 MagickLibrary * GetOpenCLLib()
454 if (OpenCLLib == NULL)
456 if (OpenCLLibLock == NULL)
458 ActivateSemaphoreInfo(&OpenCLLibLock);
461 LockSemaphoreInfo(OpenCLLibLock);
463 OpenCLLib = (MagickLibrary *) AcquireMagickMemory (
sizeof (MagickLibrary));
465 if (OpenCLLib != NULL)
467 MagickBooleanType status = MagickFalse;
468 void * library = NULL;
470 #ifdef MAGICKCORE_OPENCL_MACOSX
471 status = bindOpenCLFunctions(library);
474 memset(OpenCLLib, 0,
sizeof(MagickLibrary));
475 #ifdef MAGICKCORE_WINDOWS_SUPPORT
476 library = OsLibraryLoad(
"OpenCL.dll");
478 library = OsLibraryLoad(
"libOpenCL.so");
481 status = bindOpenCLFunctions(library);
483 if (status==MagickTrue)
484 OpenCLLib->base=library;
486 OpenCLLib=(MagickLibrary *)RelinquishMagickMemory(OpenCLLib);
490 UnlockSemaphoreInfo(OpenCLLibLock);
523 if (defaultCLEnv == NULL)
525 if (defaultCLEnvLock == NULL)
527 ActivateSemaphoreInfo(&defaultCLEnvLock);
529 LockSemaphoreInfo(defaultCLEnvLock);
530 if (defaultCLEnv == NULL)
531 defaultCLEnv = AcquireMagickOpenCLEnv();
532 UnlockSemaphoreInfo(defaultCLEnvLock);
537 static void LockDefaultOpenCLEnv() {
538 if (defaultCLEnvLock == NULL)
540 ActivateSemaphoreInfo(&defaultCLEnvLock);
542 LockSemaphoreInfo(defaultCLEnvLock);
545 static void UnlockDefaultOpenCLEnv() {
546 if (defaultCLEnvLock == NULL)
548 ActivateSemaphoreInfo(&defaultCLEnvLock);
551 UnlockSemaphoreInfo(defaultCLEnvLock);
581 LockDefaultOpenCLEnv();
582 oldEnv = defaultCLEnv;
583 defaultCLEnv = clEnv;
584 UnlockDefaultOpenCLEnv();
621 static MagickBooleanType SetMagickOpenCLEnvParamInternal(
MagickCLEnv clEnv, MagickOpenCLEnvParam param
624 MagickBooleanType status = MagickFalse;
632 case MAGICK_OPENCL_ENV_PARAM_DEVICE:
633 if (dataSize !=
sizeof(clEnv->device))
635 clEnv->device = *((cl_device_id*)data);
636 clEnv->OpenCLInitialized = MagickFalse;
640 case MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED:
641 if (dataSize !=
sizeof(clEnv->OpenCLDisabled))
643 clEnv->OpenCLDisabled = *((MagickBooleanType*)data);
644 clEnv->OpenCLInitialized = MagickFalse;
648 case MAGICK_OPENCL_ENV_PARAM_OPENCL_INITIALIZED:
649 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"SetMagickOpenCLEnvParm cannot modify the OpenCL initialization state.",
"'%s'",
".");
652 case MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED:
653 if (dataSize !=
sizeof(clEnv->disableProgramCache))
655 clEnv->disableProgramCache = *((MagickBooleanType*)data);
656 clEnv->OpenCLInitialized = MagickFalse;
660 case MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE:
661 if (dataSize !=
sizeof(clEnv->regenerateProfile))
663 clEnv->regenerateProfile = *((MagickBooleanType*)data);
664 clEnv->OpenCLInitialized = MagickFalse;
677 MagickBooleanType SetMagickOpenCLEnvParam(
MagickCLEnv clEnv, MagickOpenCLEnvParam param
679 MagickBooleanType status = MagickFalse;
681 LockSemaphoreInfo(clEnv->lock);
682 status = SetMagickOpenCLEnvParamInternal(clEnv,param,dataSize,data,exception);
683 UnlockSemaphoreInfo(clEnv->lock);
722 MagickBooleanType GetMagickOpenCLEnvParam(
MagickCLEnv clEnv, MagickOpenCLEnvParam param
731 magick_unreferenced(exception);
733 status = MagickFalse;
741 case MAGICK_OPENCL_ENV_PARAM_DEVICE:
742 if (dataSize !=
sizeof(cl_device_id))
744 *((cl_device_id*)data) = clEnv->device;
748 case MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED:
749 if (dataSize !=
sizeof(clEnv->OpenCLDisabled))
751 *((MagickBooleanType*)data) = clEnv->OpenCLDisabled;
755 case MAGICK_OPENCL_ENV_PARAM_OPENCL_INITIALIZED:
756 if (dataSize !=
sizeof(clEnv->OpenCLDisabled))
758 *((MagickBooleanType*)data) = clEnv->OpenCLInitialized;
762 case MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED:
763 if (dataSize !=
sizeof(clEnv->disableProgramCache))
765 *((MagickBooleanType*)data) = clEnv->disableProgramCache;
769 case MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE:
770 if (dataSize !=
sizeof(clEnv->regenerateProfile))
772 *((MagickBooleanType*)data) = clEnv->regenerateProfile;
776 case MAGICK_OPENCL_ENV_PARAM_PLATFORM_VENDOR:
777 if (dataSize !=
sizeof(
char *))
779 clEnv->library->clGetPlatformInfo(clEnv->platform,CL_PLATFORM_VENDOR,0,
781 *((
char **) data)=(
char *) AcquireQuantumMemory(length,
sizeof(
char));
782 clEnv->library->clGetPlatformInfo(clEnv->platform,CL_PLATFORM_VENDOR,
783 length,*((
char **) data),NULL);
787 case MAGICK_OPENCL_ENV_PARAM_DEVICE_NAME:
788 if (dataSize !=
sizeof(
char *))
790 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_NAME,0,NULL,
792 *((
char **) data)=(
char *) AcquireQuantumMemory(length,
sizeof(
char));
793 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_NAME,length,
794 *((
char **) data),NULL);
835 return clEnv->context;
838 static char* getBinaryCLProgramName(
MagickCLEnv clEnv, MagickOpenCLProgram prog,
unsigned int signature)
842 char path[MaxTextExtent];
843 char deviceName[MaxTextExtent];
844 const char* prefix =
"magick_opencl";
845 clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_NAME, MaxTextExtent, deviceName, NULL);
850 if ( *ptr ==
' ' || *ptr ==
'\\' || *ptr ==
'/' || *ptr ==
':' || *ptr ==
'*'
851 || *ptr ==
'?' || *ptr ==
'"' || *ptr ==
'<' || *ptr ==
'>' || *ptr ==
'|')
857 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s_%s_%02d_%08x_%.20g.bin",
858 GetOpenCLCachedFilesDirectory(),DirectorySeparator,prefix,deviceName,
859 (
unsigned int) prog,signature,(
double)
sizeof(
char*)*8);
860 name = (
char*)AcquireMagickMemory(strlen(path)+1);
861 CopyMagickString(name,path,strlen(path)+1);
865 static void saveBinaryCLProgram(
MagickCLEnv clEnv,MagickOpenCLProgram prog,
882 filename=getBinaryCLProgramName(clEnv,prog,signature);
883 status=clEnv->library->clGetProgramInfo(clEnv->programs[prog],
884 CL_PROGRAM_NUM_DEVICES,
sizeof(cl_uint),&num_devices,NULL);
885 if (status != CL_SUCCESS)
887 size=num_devices*
sizeof(*program_sizes);
888 program_sizes=(
size_t*) AcquireQuantumMemory(1,size);
889 if (program_sizes == (
size_t*) NULL)
891 status=clEnv->library->clGetProgramInfo(clEnv->programs[prog],
892 CL_PROGRAM_BINARY_SIZES,size,program_sizes,NULL);
893 if (status == CL_SUCCESS)
901 binary_program_size=num_devices*
sizeof(*binary_program);
902 binary_program=(
unsigned char **) AcquireQuantumMemory(1,
903 binary_program_size);
904 if (binary_program == (
unsigned char **) NULL)
906 program_sizes=(
size_t *) RelinquishMagickMemory(program_sizes);
909 for (i = 0; i < num_devices; i++)
911 binary_program[i]=AcquireQuantumMemory(MagickMax(*(program_sizes+i),1),
912 sizeof(**binary_program));
913 if (binary_program[i] == (
unsigned char *) NULL)
915 status=CL_OUT_OF_HOST_MEMORY;
919 if (status == CL_SUCCESS)
920 status=clEnv->library->clGetProgramInfo(clEnv->programs[prog],
921 CL_PROGRAM_BINARIES,binary_program_size,binary_program,NULL);
922 if (status == CL_SUCCESS)
924 for (i = 0; i < num_devices; i++)
932 program_size=*(program_sizes+i);
933 if (program_size < 1)
935 file=open_utf8(filename,O_WRONLY | O_CREAT | O_BINARY,S_MODE);
938 write(file,binary_program[i],program_size);
942 (
void) ThrowMagickException(exception,GetMagickModule(),
943 DelegateWarning,
"Saving kernel failed.",
"`%s'",filename);
947 for (i = 0; i < num_devices; i++)
948 binary_program[i]=(
unsigned char *) RelinquishMagickMemory(
950 binary_program=(
unsigned char **) RelinquishMagickMemory(binary_program);
952 program_sizes=(
size_t *) RelinquishMagickMemory(program_sizes);
955 static MagickBooleanType loadBinaryCLProgram(
MagickCLEnv clEnv, MagickOpenCLProgram prog,
unsigned int signature)
957 MagickBooleanType loadSuccessful;
958 unsigned char* binaryProgram;
959 char* binaryFileName;
962 #ifdef MAGICKCORE_CLPERFMARKER
963 clBeginPerfMarkerAMD(__FUNCTION__,
"");
966 binaryProgram = NULL;
967 binaryFileName = NULL;
969 loadSuccessful = MagickFalse;
971 binaryFileName = getBinaryCLProgramName(clEnv, prog, signature);
972 fileHandle = fopen(binaryFileName,
"rb");
973 if (fileHandle != NULL)
978 cl_int clBinaryStatus;
982 b_error |= fseek( fileHandle, 0, SEEK_END ) < 0;
983 b_error |= ( length = ftell( fileHandle ) ) <= 0;
984 b_error |= fseek( fileHandle, 0, SEEK_SET ) < 0;
988 binaryProgram = (
unsigned char*)AcquireMagickMemory(length);
989 if (binaryProgram == NULL)
992 memset(binaryProgram, 0, length);
993 b_error |= fread(binaryProgram, 1, length, fileHandle) != length;
995 clEnv->programs[prog] = clEnv->library->clCreateProgramWithBinary(clEnv->context, 1, &clEnv->device, &length, (
const unsigned char**)&binaryProgram, &clBinaryStatus, &clStatus);
996 if (clStatus != CL_SUCCESS
997 || clBinaryStatus != CL_SUCCESS)
1000 loadSuccessful = MagickTrue;
1004 if (fileHandle != NULL)
1006 if (binaryFileName != NULL)
1007 RelinquishMagickMemory(binaryFileName);
1008 if (binaryProgram != NULL)
1009 RelinquishMagickMemory(binaryProgram);
1011 #ifdef MAGICKCORE_CLPERFMARKER
1012 clEndPerfMarkerAMD();
1015 return loadSuccessful;
1018 static unsigned int stringSignature(
const char*
string)
1020 unsigned int stringLength;
1022 unsigned int signature;
1026 const unsigned int* u;
1029 #ifdef MAGICKCORE_CLPERFMARKER
1030 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1033 stringLength = (
unsigned int) strlen(
string);
1034 signature = stringLength;
1035 n = stringLength/
sizeof(
unsigned int);
1037 for (i = 0; i < n; i++)
1041 if (n *
sizeof(
unsigned int) != stringLength)
1044 j = n *
sizeof(
unsigned int);
1045 for (i = 0; i < 4; i++,j++)
1047 if (j < stringLength)
1056 #ifdef MAGICKCORE_CLPERFMARKER
1057 clEndPerfMarkerAMD();
1064 extern const char *accelerateKernels, *accelerateKernels2;
1068 MagickBooleanType status = MagickFalse;
1071 char* accelerateKernelsBuffer = NULL;
1074 const char* MagickOpenCLProgramStrings[MAGICK_OPENCL_NUM_PROGRAMS];
1076 char options[MaxTextExtent];
1077 unsigned int optionsSignature;
1079 #ifdef MAGICKCORE_CLPERFMARKER
1080 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1084 (void) FormatLocaleString(options, MaxTextExtent, CLOptions, (
float)QuantumRange,
1085 (float)QuantumScale, (
float)CLCharQuantumScale, (
float)MagickEpsilon, (
float)MagickPI, (
unsigned int)MaxMap, (
unsigned int)MAGICKCORE_QUANTUM_DEPTH);
1100 optionsSignature = stringSignature(options);
1103 accelerateKernelsBuffer = (
char*) AcquireQuantumMemory(1,strlen(accelerateKernels)+strlen(accelerateKernels2)+1);
1104 FormatLocaleString(accelerateKernelsBuffer,strlen(accelerateKernels)+
1105 strlen(accelerateKernels2)+1,"%s%s",accelerateKernels,accelerateKernels2);
1106 MagickOpenCLProgramStrings[MAGICK_OPENCL_ACCELERATE] = accelerateKernelsBuffer;
1108 for (i = 0; i < MAGICK_OPENCL_NUM_PROGRAMS; i++)
1110 MagickBooleanType loadSuccessful = MagickFalse;
1111 unsigned int programSignature = stringSignature(MagickOpenCLProgramStrings[i]) ^ optionsSignature;
1114 if (clEnv->disableProgramCache != MagickTrue
1115 && !getenv(
"MAGICK_OCL_REC"))
1116 loadSuccessful = loadBinaryCLProgram(clEnv, (MagickOpenCLProgram)i, programSignature);
1118 if (loadSuccessful == MagickFalse)
1121 size_t programLength = strlen(MagickOpenCLProgramStrings[i]);
1122 clEnv->programs[i] = clEnv->library->clCreateProgramWithSource(clEnv->context, 1, &(MagickOpenCLProgramStrings[i]), &programLength, &clStatus);
1123 if (clStatus!=CL_SUCCESS)
1125 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1126 "clCreateProgramWithSource failed.",
"(%d)", (int)clStatus);
1132 clStatus = clEnv->library->clBuildProgram(clEnv->programs[i], 1, &clEnv->device, options, NULL, NULL);
1133 if (clStatus!=CL_SUCCESS)
1135 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1136 "clBuildProgram failed.",
"(%d)", (int)clStatus);
1138 if (loadSuccessful == MagickFalse)
1140 char path[MaxTextExtent];
1144 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
1145 ,GetOpenCLCachedFilesDirectory()
1146 ,DirectorySeparator,
"magick_badcl.cl");
1147 fileHandle = fopen(path,
"wb");
1148 if (fileHandle != NULL)
1150 fwrite(MagickOpenCLProgramStrings[i],
sizeof(
char), strlen(MagickOpenCLProgramStrings[i]), fileHandle);
1158 clEnv->library->clGetProgramBuildInfo(clEnv->programs[i], clEnv->device, CL_PROGRAM_BUILD_LOG, 0, NULL, &logSize);
1159 log = (
char*)AcquireCriticalMemory(logSize);
1160 clEnv->library->clGetProgramBuildInfo(clEnv->programs[i], clEnv->device, CL_PROGRAM_BUILD_LOG, logSize, log, &logSize);
1162 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
1163 ,GetOpenCLCachedFilesDirectory()
1164 ,DirectorySeparator,
"magick_badcl_build.log");
1165 fileHandle = fopen(path,
"wb");
1166 if (fileHandle != NULL)
1168 const char* buildOptionsTitle =
"build options: ";
1169 fwrite(buildOptionsTitle,
sizeof(
char), strlen(buildOptionsTitle), fileHandle);
1170 fwrite(options,
sizeof(
char), strlen(options), fileHandle);
1171 fwrite(
"\n",
sizeof(
char), 1, fileHandle);
1172 fwrite(log,
sizeof(
char), logSize, fileHandle);
1175 RelinquishMagickMemory(log);
1181 if (loadSuccessful == MagickFalse)
1184 saveBinaryCLProgram(clEnv, (MagickOpenCLProgram)i, programSignature, exception);
1188 status = MagickTrue;
1192 if (accelerateKernelsBuffer!=NULL) RelinquishMagickMemory(accelerateKernelsBuffer);
1194 #ifdef MAGICKCORE_CLPERFMARKER
1195 clEndPerfMarkerAMD();
1204 cl_uint numPlatforms = 0;
1205 cl_platform_id *platforms = NULL;
1206 char* MAGICK_OCL_DEVICE = NULL;
1207 MagickBooleanType OpenCLAvailable = MagickFalse;
1209 #ifdef MAGICKCORE_CLPERFMARKER
1210 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1214 MAGICK_OCL_DEVICE = getenv(
"MAGICK_OCL_DEVICE");
1215 if (MAGICK_OCL_DEVICE == (
char *) NULL)
1216 return(MagickFalse);
1217 if (strcmp(MAGICK_OCL_DEVICE,
"CPU") == 0)
1218 clEnv->deviceType = CL_DEVICE_TYPE_CPU;
1219 else if (strcmp(MAGICK_OCL_DEVICE,
"GPU") == 0)
1220 clEnv->deviceType = CL_DEVICE_TYPE_GPU;
1221 else if (IsStringTrue(MAGICK_OCL_DEVICE) != MagickFalse)
1223 if (clEnv->deviceType == 0)
1224 clEnv->deviceType = CL_DEVICE_TYPE_ALL;
1227 return(MagickFalse);
1229 if (clEnv->device != NULL)
1231 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_PLATFORM,
sizeof(cl_platform_id), &clEnv->platform, NULL);
1232 if (status != CL_SUCCESS) {
1233 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1234 "Failed to get OpenCL platform from the selected device.",
"(%d)", status);
1238 else if (clEnv->platform != NULL)
1241 platforms = (cl_platform_id *) AcquireQuantumMemory(1,numPlatforms *
sizeof(cl_platform_id));
1242 if (platforms == (cl_platform_id *) NULL)
1244 (void) ThrowMagickException(exception, GetMagickModule(), ResourceLimitError,
1245 "AcquireMagickMemory failed.",
".");
1248 platforms[0] = clEnv->platform;
1252 clEnv->device = NULL;
1255 status = clEnv->library->clGetPlatformIDs(0, NULL, &numPlatforms);
1256 if (status != CL_SUCCESS)
1258 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1259 "clGetplatformIDs failed.",
"(%d)", status);
1264 if (numPlatforms == 0) {
1268 platforms = (cl_platform_id *) AcquireQuantumMemory(1,numPlatforms *
sizeof(cl_platform_id));
1269 if (platforms == (cl_platform_id *) NULL)
1271 (void) ThrowMagickException(exception, GetMagickModule(), ResourceLimitError,
1272 "AcquireMagickMemory failed.",
".");
1276 status = clEnv->library->clGetPlatformIDs(numPlatforms, platforms, NULL);
1277 if (status != CL_SUCCESS)
1279 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1280 "clGetPlatformIDs failed.",
"(%d)", status);
1286 clEnv->device = NULL;
1287 for (j = 0; j < 2; j++)
1290 cl_device_type deviceType;
1291 if (clEnv->deviceType == CL_DEVICE_TYPE_ALL)
1294 deviceType = CL_DEVICE_TYPE_GPU;
1296 deviceType = CL_DEVICE_TYPE_CPU;
1303 deviceType = clEnv->deviceType;
1305 for (i = 0; i < numPlatforms; i++)
1307 char version[MaxTextExtent];
1309 status = clEnv->library->clGetPlatformInfo(clEnv->platform, CL_PLATFORM_VERSION, MaxTextExtent, version, NULL);
1310 if (status != CL_SUCCESS)
1312 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1313 "clGetPlatformInfo failed.",
"(%d)", status);
1316 if (strncmp(version,
"OpenCL 1.0 ",11) == 0)
1318 status = clEnv->library->clGetDeviceIDs(platforms[i], deviceType, 1, &(clEnv->device), &numDevices);
1319 if (status != CL_SUCCESS)
1321 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1322 "clGetDeviceIDs failed.",
"(%d)", status);
1325 if (clEnv->device != NULL)
1327 clEnv->platform = platforms[i];
1334 if (platforms!=NULL)
1335 RelinquishMagickMemory(platforms);
1337 OpenCLAvailable = (clEnv->platform!=NULL
1338 && clEnv->device!=NULL)?MagickTrue:MagickFalse;
1340 #ifdef MAGICKCORE_CLPERFMARKER
1341 clEndPerfMarkerAMD();
1344 return OpenCLAvailable;
1347 static MagickBooleanType EnableOpenCLInternal(
MagickCLEnv clEnv) {
1348 if (clEnv->OpenCLInitialized != MagickFalse
1349 && clEnv->platform != NULL
1350 && clEnv->device != NULL) {
1351 clEnv->OpenCLDisabled = MagickFalse;
1354 clEnv->OpenCLDisabled = MagickTrue;
1385 static void RelinquishCommandQueues(
MagickCLEnv clEnv)
1390 LockSemaphoreInfo(clEnv->commandQueuesLock);
1391 while (clEnv->commandQueuesPos >= 0)
1392 clEnv->library->clReleaseCommandQueue(
1393 clEnv->commandQueues[clEnv->commandQueuesPos--]);
1394 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1399 MagickBooleanType status = MagickTrue;
1401 cl_context_properties cps[3];
1403 #ifdef MAGICKCORE_CLPERFMARKER
1405 int status = clInitializePerfMarkerAMD();
1406 if (status == AP_SUCCESS) {
1411 clEnv->OpenCLInitialized = MagickTrue;
1414 OpenCLLib=GetOpenCLLib();
1417 clEnv->library=OpenCLLib;
1422 MagickBooleanType flag;
1424 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
1425 ,
sizeof(MagickBooleanType), &flag, exception);
1428 if (clEnv->OpenCLDisabled != MagickFalse)
1431 clEnv->OpenCLDisabled = MagickTrue;
1433 status = InitOpenCLPlatformDevice(clEnv, exception);
1434 if (status == MagickFalse) {
1440 cps[0] = CL_CONTEXT_PLATFORM;
1441 cps[1] = (cl_context_properties)clEnv->platform;
1443 clEnv->context = clEnv->library->clCreateContext(cps, 1, &(clEnv->device), NULL, NULL, &clStatus);
1444 if (clStatus != CL_SUCCESS)
1446 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1447 "clCreateContext failed.",
"(%d)", clStatus);
1448 status = MagickFalse;
1452 RelinquishCommandQueues(clEnv);
1454 status = CompileOpenCLKernels(clEnv, exception);
1455 if (status == MagickFalse) {
1456 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1457 "clCreateCommandQueue failed.",
"(%d)", status);
1462 status = EnableOpenCLInternal(clEnv);
1471 MagickBooleanType status = MagickFalse;
1473 if ((clEnv == NULL) || (getenv(
"MAGICK_OCL_DEVICE") == (
const char *) NULL))
1476 #ifdef MAGICKCORE_CLPERFMARKER
1477 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1480 LockSemaphoreInfo(clEnv->lock);
1481 if (clEnv->OpenCLInitialized == MagickFalse) {
1482 if (clEnv->device==NULL && clEnv->OpenCLDisabled == MagickFalse)
1483 status = autoSelectDevice(clEnv, exception);
1485 status = InitOpenCLEnvInternal(clEnv, exception);
1487 UnlockSemaphoreInfo(clEnv->lock);
1489 #ifdef MAGICKCORE_CLPERFMARKER
1490 clEndPerfMarkerAMD();
1519 MagickPrivate cl_command_queue AcquireOpenCLCommandQueue(
MagickCLEnv clEnv)
1524 cl_command_queue_properties
1528 return (cl_command_queue) NULL;
1529 LockSemaphoreInfo(clEnv->commandQueuesLock);
1530 if (clEnv->commandQueuesPos >= 0) {
1531 queue=clEnv->commandQueues[clEnv->commandQueuesPos--];
1532 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1535 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1537 #if PROFILE_OCL_KERNELS
1538 properties=CL_QUEUE_PROFILING_ENABLE;
1540 queue=clEnv->library->clCreateCommandQueue(clEnv->context,clEnv->device,
1573 MagickPrivate MagickBooleanType RelinquishOpenCLCommandQueue(
MagickCLEnv clEnv,
1574 cl_command_queue queue)
1580 return(MagickFalse);
1582 LockSemaphoreInfo(clEnv->commandQueuesLock);
1584 if (clEnv->commandQueuesPos >= MAX_COMMAND_QUEUES-1)
1586 clEnv->library->clFinish(queue);
1587 status=(clEnv->library->clReleaseCommandQueue(queue) == CL_SUCCESS) ?
1588 MagickTrue : MagickFalse;
1592 clEnv->library->clFlush(queue);
1593 clEnv->commandQueues[++clEnv->commandQueuesPos]=queue;
1597 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1631 cl_kernel AcquireOpenCLKernel(
MagickCLEnv clEnv, MagickOpenCLProgram program,
const char* kernelName)
1634 cl_kernel kernel = NULL;
1635 if (clEnv != NULL && kernelName!=NULL)
1637 kernel = clEnv->library->clCreateKernel(clEnv->programs[program], kernelName, &clStatus);
1671 MagickBooleanType RelinquishOpenCLKernel(
MagickCLEnv clEnv, cl_kernel kernel)
1673 MagickBooleanType status = MagickFalse;
1674 if (clEnv != NULL && kernel != NULL)
1676 status = ((clEnv->library->clReleaseKernel(kernel) == CL_SUCCESS)?MagickTrue:MagickFalse);
1706 unsigned long GetOpenCLDeviceLocalMemorySize(
MagickCLEnv clEnv)
1708 cl_ulong localMemorySize;
1709 clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_LOCAL_MEM_SIZE,
sizeof(cl_ulong), &localMemorySize, NULL);
1710 return (
unsigned long)localMemorySize;
1714 unsigned long GetOpenCLDeviceMaxMemAllocSize(
MagickCLEnv clEnv)
1716 cl_ulong maxMemAllocSize;
1717 clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
sizeof(cl_ulong), &maxMemAllocSize, NULL);
1718 return (
unsigned long)maxMemAllocSize;
1729 ,DS_INVALID_PROFILE = 1000
1731 ,DS_INVALID_PERF_EVALUATOR_TYPE
1732 ,DS_INVALID_PERF_EVALUATOR
1733 ,DS_PERF_EVALUATOR_ERROR
1735 ,DS_UNKNOWN_DEVICE_TYPE
1736 ,DS_PROFILE_FILE_ERROR
1737 ,DS_SCORE_SERIALIZER_ERROR
1738 ,DS_SCORE_DESERIALIZER_ERROR
1743 DS_DEVICE_NATIVE_CPU = 0
1744 ,DS_DEVICE_OPENCL_DEVICE
1749 ds_device_type type;
1750 cl_device_type oclDeviceType;
1751 cl_device_id oclDeviceID;
1752 char* oclDeviceName;
1753 char* oclDriverVersion;
1754 cl_uint oclMaxClockFrequency;
1755 cl_uint oclMaxComputeUnits;
1760 unsigned int numDevices;
1762 const char* version;
1766 typedef ds_status (*ds_score_release)(
void* score);
1768 static ds_status releaseDeviceResource(ds_device* device, ds_score_release sr) {
1769 ds_status status = DS_SUCCESS;
1771 if (device->oclDeviceName) RelinquishMagickMemory(device->oclDeviceName);
1772 if (device->oclDriverVersion) RelinquishMagickMemory(device->oclDriverVersion);
1773 if (device->score) status = sr(device->score);
1778 static ds_status releaseDSProfile(ds_profile* profile, ds_score_release sr) {
1779 ds_status status = DS_SUCCESS;
1780 if (profile!=NULL) {
1781 if (profile->devices!=NULL && sr!=NULL) {
1783 for (i = 0; i < profile->numDevices; i++) {
1784 status = releaseDeviceResource(profile->devices+i,sr);
1785 if (status != DS_SUCCESS)
1788 RelinquishMagickMemory(profile->devices);
1790 RelinquishMagickMemory(profile);
1796 static ds_status initDSProfile(ds_profile** p,
const char* version) {
1798 cl_uint numPlatforms = 0;
1799 cl_platform_id* platforms = NULL;
1800 cl_device_id* devices = NULL;
1801 ds_status status = DS_SUCCESS;
1802 ds_profile* profile = NULL;
1803 unsigned int next = 0;
1807 return DS_INVALID_PROFILE;
1809 profile = (ds_profile*) AcquireMagickMemory(
sizeof(ds_profile));
1810 if (profile == NULL)
1811 return DS_MEMORY_ERROR;
1813 memset(profile, 0,
sizeof(ds_profile));
1815 OpenCLLib->clGetPlatformIDs(0, NULL, &numPlatforms);
1816 if (numPlatforms > 0) {
1817 platforms = (cl_platform_id*) AcquireQuantumMemory(numPlatforms,
sizeof(cl_platform_id));
1818 if (platforms == NULL) {
1819 status = DS_MEMORY_ERROR;
1822 OpenCLLib->clGetPlatformIDs(numPlatforms, platforms, NULL);
1823 for (i = 0; i < (
unsigned int)numPlatforms; i++) {
1825 if (OpenCLLib->clGetDeviceIDs(platforms[i], CL_DEVICE_TYPE_CPU | CL_DEVICE_TYPE_GPU, 0, NULL, &num) == CL_SUCCESS)
1830 profile->numDevices = numDevices+1;
1832 profile->devices = (ds_device*) AcquireQuantumMemory(profile->numDevices,
sizeof(ds_device));
1833 if (profile->devices == NULL) {
1834 profile->numDevices = 0;
1835 status = DS_MEMORY_ERROR;
1838 memset(profile->devices, 0, profile->numDevices*
sizeof(ds_device));
1840 if (numDevices > 0) {
1841 devices = (cl_device_id*) AcquireQuantumMemory(numDevices,
sizeof(cl_device_id));
1842 if (devices == NULL) {
1843 status = DS_MEMORY_ERROR;
1846 for (i = 0; i < (
unsigned int)numPlatforms; i++) {
1850 for (d = 0; d < 2; d++) {
1852 cl_device_type deviceType;
1855 deviceType = CL_DEVICE_TYPE_GPU;
1858 deviceType = CL_DEVICE_TYPE_CPU;
1864 if (OpenCLLib->clGetDeviceIDs(platforms[i], deviceType, numDevices, devices, &num) != CL_SUCCESS)
1866 for (j = 0; j < num; j++, next++) {
1869 profile->devices[next].type = DS_DEVICE_OPENCL_DEVICE;
1870 profile->devices[next].oclDeviceID = devices[j];
1872 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_NAME
1873 , 0, NULL, &length);
1874 profile->devices[next].oclDeviceName = (
char*) AcquireQuantumMemory(length,
sizeof(
char));
1875 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_NAME
1876 , length, profile->devices[next].oclDeviceName, NULL);
1878 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DRIVER_VERSION
1879 , 0, NULL, &length);
1880 profile->devices[next].oclDriverVersion = (
char*) AcquireQuantumMemory(length,
sizeof(
char));
1881 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DRIVER_VERSION
1882 , length, profile->devices[next].oclDriverVersion, NULL);
1884 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_MAX_CLOCK_FREQUENCY
1885 ,
sizeof(cl_uint), &profile->devices[next].oclMaxClockFrequency, NULL);
1887 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_MAX_COMPUTE_UNITS
1888 ,
sizeof(cl_uint), &profile->devices[next].oclMaxComputeUnits, NULL);
1890 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_TYPE
1891 ,
sizeof(cl_device_type), &profile->devices[next].oclDeviceType, NULL);
1897 profile->devices[next].type = DS_DEVICE_NATIVE_CPU;
1898 profile->version = version;
1901 if (platforms) RelinquishMagickMemory(platforms);
1902 if (devices) RelinquishMagickMemory(devices);
1903 if (status == DS_SUCCESS) {
1908 if (profile->devices)
1909 RelinquishMagickMemory(profile->devices);
1910 RelinquishMagickMemory(profile);
1920 typedef ds_status (*ds_perf_evaluator)(ds_device* device,
void* data);
1924 ,DS_EVALUATE_NEW_ONLY
1925 } ds_evaluation_type;
1927 static ds_status profileDevices(ds_profile* profile,
const ds_evaluation_type type
1928 ,ds_perf_evaluator evaluator,
void* evaluatorData,
unsigned int* numUpdates) {
1929 ds_status status = DS_SUCCESS;
1931 unsigned int updates = 0;
1933 if (profile == NULL) {
1934 return DS_INVALID_PROFILE;
1936 if (evaluator == NULL) {
1937 return DS_INVALID_PERF_EVALUATOR;
1940 for (i = 0; i < profile->numDevices; i++) {
1941 ds_status evaluatorStatus;
1944 case DS_EVALUATE_NEW_ONLY:
1945 if (profile->devices[i].score != NULL)
1948 case DS_EVALUATE_ALL:
1949 evaluatorStatus = evaluator(profile->devices+i,evaluatorData);
1950 if (evaluatorStatus != DS_SUCCESS) {
1951 status = evaluatorStatus;
1957 return DS_INVALID_PERF_EVALUATOR_TYPE;
1962 *numUpdates = updates;
1967 #define DS_TAG_VERSION "<version>"
1968 #define DS_TAG_VERSION_END "</version>"
1969 #define DS_TAG_DEVICE "<device>"
1970 #define DS_TAG_DEVICE_END "</device>"
1971 #define DS_TAG_SCORE "<score>"
1972 #define DS_TAG_SCORE_END "</score>"
1973 #define DS_TAG_DEVICE_TYPE "<type>"
1974 #define DS_TAG_DEVICE_TYPE_END "</type>"
1975 #define DS_TAG_DEVICE_NAME "<name>"
1976 #define DS_TAG_DEVICE_NAME_END "</name>"
1977 #define DS_TAG_DEVICE_DRIVER_VERSION "<driver>"
1978 #define DS_TAG_DEVICE_DRIVER_VERSION_END "</driver>"
1979 #define DS_TAG_DEVICE_MAX_COMPUTE_UNITS "<max cu>"
1980 #define DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END "</max cu>"
1981 #define DS_TAG_DEVICE_MAX_CLOCK_FREQ "<max clock>"
1982 #define DS_TAG_DEVICE_MAX_CLOCK_FREQ_END "</max clock>"
1984 #define DS_DEVICE_NATIVE_CPU_STRING "native_cpu"
1988 typedef ds_status (*ds_score_serializer)(ds_device* device,
void** serializedScore,
unsigned int* serializedScoreSize);
1989 static ds_status writeProfileToFile(ds_profile* profile, ds_score_serializer serializer,
const char* file) {
1990 ds_status status = DS_SUCCESS;
1991 FILE* profileFile = NULL;
1994 if (profile == NULL)
1995 return DS_INVALID_PROFILE;
1997 profileFile = fopen(file,
"wb");
1998 if (profileFile==NULL) {
1999 status = DS_FILE_ERROR;
2005 fwrite(DS_TAG_VERSION,
sizeof(
char), strlen(DS_TAG_VERSION), profileFile);
2006 fwrite(profile->version,
sizeof(
char), strlen(profile->version), profileFile);
2007 fwrite(DS_TAG_VERSION_END,
sizeof(
char), strlen(DS_TAG_VERSION_END), profileFile);
2008 fwrite(
"\n",
sizeof(
char), 1, profileFile);
2010 for (i = 0; i < profile->numDevices && status == DS_SUCCESS; i++) {
2011 void* serializedScore;
2012 unsigned int serializedScoreSize;
2014 fwrite(DS_TAG_DEVICE,
sizeof(
char), strlen(DS_TAG_DEVICE), profileFile);
2016 fwrite(DS_TAG_DEVICE_TYPE,
sizeof(
char), strlen(DS_TAG_DEVICE_TYPE), profileFile);
2017 fwrite(&profile->devices[i].type,
sizeof(ds_device_type),1, profileFile);
2018 fwrite(DS_TAG_DEVICE_TYPE_END,
sizeof(
char), strlen(DS_TAG_DEVICE_TYPE_END), profileFile);
2020 switch(profile->devices[i].type) {
2021 case DS_DEVICE_NATIVE_CPU:
2031 case DS_DEVICE_OPENCL_DEVICE:
2035 fwrite(DS_TAG_DEVICE_NAME,
sizeof(
char), strlen(DS_TAG_DEVICE_NAME), profileFile);
2036 fwrite(profile->devices[i].oclDeviceName,
sizeof(
char),strlen(profile->devices[i].oclDeviceName), profileFile);
2037 fwrite(DS_TAG_DEVICE_NAME_END,
sizeof(
char), strlen(DS_TAG_DEVICE_NAME_END), profileFile);
2039 fwrite(DS_TAG_DEVICE_DRIVER_VERSION,
sizeof(
char), strlen(DS_TAG_DEVICE_DRIVER_VERSION), profileFile);
2040 fwrite(profile->devices[i].oclDriverVersion,
sizeof(
char),strlen(profile->devices[i].oclDriverVersion), profileFile);
2041 fwrite(DS_TAG_DEVICE_DRIVER_VERSION_END,
sizeof(
char), strlen(DS_TAG_DEVICE_DRIVER_VERSION_END), profileFile);
2043 fwrite(DS_TAG_DEVICE_MAX_COMPUTE_UNITS,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_COMPUTE_UNITS), profileFile);
2044 (void) FormatLocaleString(tmp,
sizeof(tmp),
"%d",
2045 profile->devices[i].oclMaxComputeUnits);
2046 fwrite(tmp,
sizeof(
char),strlen(tmp), profileFile);
2047 fwrite(DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END), profileFile);
2049 fwrite(DS_TAG_DEVICE_MAX_CLOCK_FREQ,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_CLOCK_FREQ), profileFile);
2050 (void) FormatLocaleString(tmp,
sizeof(tmp),
"%d",
2051 profile->devices[i].oclMaxClockFrequency);
2052 fwrite(tmp,
sizeof(
char),strlen(tmp), profileFile);
2053 fwrite(DS_TAG_DEVICE_MAX_CLOCK_FREQ_END,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_CLOCK_FREQ_END), profileFile);
2057 status = DS_UNKNOWN_DEVICE_TYPE;
2061 fwrite(DS_TAG_SCORE,
sizeof(
char), strlen(DS_TAG_SCORE), profileFile);
2062 status = serializer(profile->devices+i, &serializedScore, &serializedScoreSize);
2063 if (status == DS_SUCCESS && serializedScore!=NULL && serializedScoreSize > 0) {
2064 fwrite(serializedScore,
sizeof(
char), serializedScoreSize, profileFile);
2065 RelinquishMagickMemory(serializedScore);
2067 fwrite(DS_TAG_SCORE_END,
sizeof(
char), strlen(DS_TAG_SCORE_END), profileFile);
2068 fwrite(DS_TAG_DEVICE_END,
sizeof(
char), strlen(DS_TAG_DEVICE_END), profileFile);
2069 fwrite(
"\n",
sizeof(
char),1,profileFile);
2071 fclose(profileFile);
2077 static ds_status readProFile(
const char* fileName,
char** content,
size_t* contentSize) {
2078 ds_status status = DS_SUCCESS;
2079 FILE * input = NULL;
2082 char* binary = NULL;
2087 input = fopen(fileName,
"rb");
2089 return DS_FILE_ERROR;
2092 fseek(input, 0L, SEEK_END);
2093 size = ftell(input);
2095 binary = (
char*) AcquireQuantumMemory(1,size);
2096 if(binary == NULL) {
2097 status = DS_FILE_ERROR;
2100 rsize = fread(binary,
sizeof(
char), size, input);
2103 status = DS_FILE_ERROR;
2106 *contentSize = size;
2110 if (input != NULL) fclose(input);
2111 if (status != DS_SUCCESS
2112 && binary != NULL) {
2113 RelinquishMagickMemory(binary);
2121 static const char* findString(
const char* contentStart,
const char* contentEnd,
const char*
string) {
2122 size_t stringLength;
2123 const char* currentPosition;
2126 stringLength = strlen(
string);
2127 currentPosition = contentStart;
2128 for(currentPosition = contentStart; currentPosition < contentEnd; currentPosition++) {
2129 if (*currentPosition ==
string[0]) {
2130 if (currentPosition+stringLength < contentEnd) {
2131 if (strncmp(currentPosition,
string, stringLength) == 0) {
2132 found = currentPosition;
2142 typedef ds_status (*ds_score_deserializer)(ds_device* device,
const unsigned char* serializedScore,
unsigned int serializedScoreSize);
2143 static ds_status readProfileFromFile(ds_profile* profile, ds_score_deserializer deserializer,
const char* file) {
2145 ds_status status = DS_SUCCESS;
2146 char* contentStart = NULL;
2147 const char* contentEnd = NULL;
2151 return DS_INVALID_PROFILE;
2153 status = readProFile(file, &contentStart, &contentSize);
2154 if (status == DS_SUCCESS) {
2155 const char* currentPosition;
2156 const char* dataStart;
2157 const char* dataEnd;
2158 size_t versionStringLength;
2160 contentEnd = contentStart + contentSize;
2161 currentPosition = contentStart;
2165 dataStart = findString(currentPosition, contentEnd, DS_TAG_VERSION);
2166 if (dataStart == NULL) {
2167 status = DS_PROFILE_FILE_ERROR;
2170 dataStart += strlen(DS_TAG_VERSION);
2172 dataEnd = findString(dataStart, contentEnd, DS_TAG_VERSION_END);
2173 if (dataEnd==NULL) {
2174 status = DS_PROFILE_FILE_ERROR;
2178 versionStringLength = strlen(profile->version);
2179 if (versionStringLength!=(
size_t)(dataEnd-dataStart)
2180 || strncmp(profile->version, dataStart, versionStringLength)!=(
int)0) {
2182 status = DS_PROFILE_FILE_ERROR;
2185 currentPosition = dataEnd+strlen(DS_TAG_VERSION_END);
2188 DisableMSCWarning(4127)
2193 const char* deviceTypeStart;
2194 const char* deviceTypeEnd;
2195 ds_device_type deviceType;
2197 const char* deviceNameStart;
2198 const char* deviceNameEnd;
2200 const char* deviceScoreStart;
2201 const char* deviceScoreEnd;
2203 const char* deviceDriverStart;
2204 const char* deviceDriverEnd;
2206 const char* tmpStart;
2210 cl_uint maxClockFrequency;
2211 cl_uint maxComputeUnits;
2213 dataStart = findString(currentPosition, contentEnd, DS_TAG_DEVICE);
2214 if (dataStart==NULL) {
2218 dataStart+=strlen(DS_TAG_DEVICE);
2219 dataEnd = findString(dataStart, contentEnd, DS_TAG_DEVICE_END);
2220 if (dataEnd==NULL) {
2221 status = DS_PROFILE_FILE_ERROR;
2226 deviceTypeStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_TYPE);
2227 if (deviceTypeStart==NULL) {
2228 status = DS_PROFILE_FILE_ERROR;
2231 deviceTypeStart+=strlen(DS_TAG_DEVICE_TYPE);
2232 deviceTypeEnd = findString(deviceTypeStart, contentEnd, DS_TAG_DEVICE_TYPE_END);
2233 if (deviceTypeEnd==NULL) {
2234 status = DS_PROFILE_FILE_ERROR;
2237 memcpy(&deviceType, deviceTypeStart,
sizeof(ds_device_type));
2241 if (deviceType == DS_DEVICE_OPENCL_DEVICE) {
2243 deviceNameStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_NAME);
2244 if (deviceNameStart==NULL) {
2245 status = DS_PROFILE_FILE_ERROR;
2248 deviceNameStart+=strlen(DS_TAG_DEVICE_NAME);
2249 deviceNameEnd = findString(deviceNameStart, contentEnd, DS_TAG_DEVICE_NAME_END);
2250 if (deviceNameEnd==NULL) {
2251 status = DS_PROFILE_FILE_ERROR;
2256 deviceDriverStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_DRIVER_VERSION);
2257 if (deviceDriverStart==NULL) {
2258 status = DS_PROFILE_FILE_ERROR;
2261 deviceDriverStart+=strlen(DS_TAG_DEVICE_DRIVER_VERSION);
2262 deviceDriverEnd = findString(deviceDriverStart, contentEnd, DS_TAG_DEVICE_DRIVER_VERSION_END);
2263 if (deviceDriverEnd ==NULL) {
2264 status = DS_PROFILE_FILE_ERROR;
2269 tmpStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_MAX_COMPUTE_UNITS);
2270 if (tmpStart==NULL) {
2271 status = DS_PROFILE_FILE_ERROR;
2274 tmpStart+=strlen(DS_TAG_DEVICE_MAX_COMPUTE_UNITS);
2275 tmpEnd = findString(tmpStart, contentEnd, DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END);
2276 if (tmpEnd ==NULL) {
2277 status = DS_PROFILE_FILE_ERROR;
2280 memcpy(tmp,tmpStart,tmpEnd-tmpStart);
2281 tmp[tmpEnd-tmpStart] =
'\0';
2282 maxComputeUnits = strtol(tmp,(
char **) NULL,10);
2285 tmpStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_MAX_CLOCK_FREQ);
2286 if (tmpStart==NULL) {
2287 status = DS_PROFILE_FILE_ERROR;
2290 tmpStart+=strlen(DS_TAG_DEVICE_MAX_CLOCK_FREQ);
2291 tmpEnd = findString(tmpStart, contentEnd, DS_TAG_DEVICE_MAX_CLOCK_FREQ_END);
2292 if (tmpEnd ==NULL) {
2293 status = DS_PROFILE_FILE_ERROR;
2296 memcpy(tmp,tmpStart,tmpEnd-tmpStart);
2297 tmp[tmpEnd-tmpStart] =
'\0';
2298 maxClockFrequency = strtol(tmp,(
char **) NULL,10);
2302 for (i = 0; i < profile->numDevices; i++) {
2303 if (profile->devices[i].type == DS_DEVICE_OPENCL_DEVICE) {
2304 size_t actualDeviceNameLength;
2305 size_t driverVersionLength;
2307 actualDeviceNameLength = strlen(profile->devices[i].oclDeviceName);
2308 driverVersionLength = strlen(profile->devices[i].oclDriverVersion);
2309 if (actualDeviceNameLength == (
size_t)(deviceNameEnd - deviceNameStart)
2310 && driverVersionLength == (
size_t)(deviceDriverEnd - deviceDriverStart)
2311 && maxComputeUnits == profile->devices[i].oclMaxComputeUnits
2312 && maxClockFrequency == profile->devices[i].oclMaxClockFrequency
2313 && strncmp(profile->devices[i].oclDeviceName, deviceNameStart, actualDeviceNameLength)==(
int)0
2314 && strncmp(profile->devices[i].oclDriverVersion, deviceDriverStart, driverVersionLength)==(
int)0) {
2316 deviceScoreStart = findString(dataStart, contentEnd, DS_TAG_SCORE);
2317 if (deviceNameStart==NULL) {
2318 status = DS_PROFILE_FILE_ERROR;
2321 deviceScoreStart+=strlen(DS_TAG_SCORE);
2322 deviceScoreEnd = findString(deviceScoreStart, contentEnd, DS_TAG_SCORE_END);
2323 status = deserializer(profile->devices+i, (
const unsigned char*)deviceScoreStart, deviceScoreEnd-deviceScoreStart);
2324 if (status != DS_SUCCESS) {
2332 else if (deviceType == DS_DEVICE_NATIVE_CPU) {
2333 for (i = 0; i < profile->numDevices; i++) {
2334 if (profile->devices[i].type == DS_DEVICE_NATIVE_CPU) {
2335 deviceScoreStart = findString(dataStart, contentEnd, DS_TAG_SCORE);
2336 if (deviceScoreStart==NULL) {
2337 status = DS_PROFILE_FILE_ERROR;
2340 deviceScoreStart+=strlen(DS_TAG_SCORE);
2341 deviceScoreEnd = findString(deviceScoreStart, contentEnd, DS_TAG_SCORE_END);
2342 status = deserializer(profile->devices+i, (
const unsigned char*)deviceScoreStart, deviceScoreEnd-deviceScoreStart);
2343 if (status != DS_SUCCESS) {
2351 currentPosition = dataEnd+strlen(DS_TAG_DEVICE_END);
2355 if (contentStart!=NULL) RelinquishMagickMemory(contentStart);
2361 static ds_status getNumDeviceWithEmptyScore(ds_profile* profile,
unsigned int* num) {
2363 if (profile == NULL || num==NULL)
2364 return DS_MEMORY_ERROR;
2366 for (i = 0; i < profile->numDevices; i++) {
2367 if (profile->devices[i].score == NULL) {
2380 typedef double AccelerateScoreType;
2382 static ds_status AcceleratePerfEvaluator(ds_device *device,
2383 void *magick_unused(data))
2385 #define ACCELERATE_PERF_DIMEN "2048x1536"
2387 #define ReturnStatus(status) \
2389 if (oldClEnv != (MagickCLEnv) NULL) \
2390 defaultCLEnv=oldClEnv; \
2391 if (clEnv != (MagickCLEnv) NULL) \
2392 (void) RelinquishMagickOpenCLEnv(clEnv); \
2409 magick_unreferenced(data);
2412 ReturnStatus(DS_PERF_EVALUATOR_ERROR);
2414 clEnv=AcquireMagickOpenCLEnv();
2415 exception=AcquireExceptionInfo();
2417 if (device->type == DS_DEVICE_NATIVE_CPU)
2420 MagickBooleanType flag=MagickTrue;
2421 SetMagickOpenCLEnvParamInternal(clEnv,
2422 MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED,
sizeof(MagickBooleanType),
2425 else if (device->type == DS_DEVICE_OPENCL_DEVICE)
2428 SetMagickOpenCLEnvParamInternal(clEnv,MAGICK_OPENCL_ENV_PARAM_DEVICE,
2429 sizeof(cl_device_id),&device->oclDeviceID,exception);
2432 ReturnStatus(DS_PERF_EVALUATOR_ERROR);
2435 clEnv->disableProgramCache = defaultCLEnv->disableProgramCache;
2437 status=InitOpenCLEnvInternal(clEnv,exception);
2438 oldClEnv=defaultCLEnv;
2442 if (status != MagickFalse)
2453 imageInfo=AcquireImageInfo();
2454 CloneString(&imageInfo->size,ACCELERATE_PERF_DIMEN);
2455 CopyMagickString(imageInfo->filename,
"xc:none",MaxTextExtent);
2456 inputImage=ReadImage(imageInfo,exception);
2458 initAccelerateTimer(&timer);
2460 for (i=0; i<=NUM_ITER; i++)
2474 startAccelerateTimer(&timer);
2476 #ifdef MAGICKCORE_CLPERFMARKER
2477 clBeginPerfMarkerAMD(
"PerfEvaluatorRegion",
"");
2480 bluredImage=BlurImage(inputImage,10.0f,3.5f,exception);
2481 unsharpedImage=UnsharpMaskImage(bluredImage,2.0f,2.0f,50.0f,10.0f,
2483 resizedImage=ResizeImage(unsharpedImage,640,480,LanczosFilter,1.0,
2490 if (device->type != DS_DEVICE_NATIVE_CPU)
2492 events=GetOpenCLEvents(resizedImage,&event_count);
2493 if (event_count > 0)
2494 clEnv->library->clWaitForEvents(event_count,events);
2495 events=(cl_event *) RelinquishMagickMemory(events);
2498 #ifdef MAGICKCORE_CLPERFMARKER
2499 clEndPerfMarkerAMD();
2503 stopAccelerateTimer(&timer);
2506 DestroyImage(bluredImage);
2508 DestroyImage(unsharpedImage);
2510 DestroyImage(resizedImage);
2512 DestroyImage(inputImage);
2516 if (device->score == NULL)
2517 device->score= AcquireMagickMemory(
sizeof(AccelerateScoreType));
2519 if (status != MagickFalse)
2520 *(AccelerateScoreType*) device->score=readAccelerateTimer(&timer);
2522 *(AccelerateScoreType*) device->score=42;
2524 ReturnStatus(DS_SUCCESS);
2527 ds_status AccelerateScoreSerializer(ds_device* device,
void** serializedScore,
unsigned int* serializedScoreSize) {
2531 char* s = (
char*) AcquireQuantumMemory(256,
sizeof(
char));
2532 (void) FormatLocaleString(s,256,
"%.4f",*((AccelerateScoreType*)
2534 *serializedScore = (
void*)s;
2535 *serializedScoreSize = (
unsigned int) strlen(s);
2539 return DS_SCORE_SERIALIZER_ERROR;
2543 ds_status AccelerateScoreDeserializer(ds_device* device,
const unsigned char* serializedScore,
unsigned int serializedScoreSize) {
2546 char* s = (
char*) AcquireQuantumMemory(1,serializedScoreSize+1);
2547 memcpy(s, serializedScore, serializedScoreSize);
2548 s[serializedScoreSize] = (char)
'\0';
2549 device->score = AcquireMagickMemory(
sizeof(AccelerateScoreType));
2550 *((AccelerateScoreType*)device->score) = (AccelerateScoreType)
2551 strtod(s, (
char **) NULL);
2552 RelinquishMagickMemory(s);
2556 return DS_SCORE_DESERIALIZER_ERROR;
2560 ds_status AccelerateScoreRelease(
void* score) {
2562 RelinquishMagickMemory(score);
2567 ds_status canWriteProfileToFile(
const char *path)
2569 FILE* profileFile = fopen(path,
"ab");
2571 if (profileFile==NULL)
2572 return DS_FILE_ERROR;
2574 fclose(profileFile);
2579 #define IMAGEMAGICK_PROFILE_VERSION "ImageMagick Device Selection v0.9"
2580 #define IMAGEMAGICK_PROFILE_FILE "ImagemagickOpenCLDeviceProfile"
2583 MagickBooleanType mStatus = MagickFalse;
2585 ds_profile* profile;
2586 unsigned int numDeviceProfiled = 0;
2588 unsigned int bestDeviceIndex;
2589 AccelerateScoreType bestScore;
2590 char path[MaxTextExtent];
2591 MagickBooleanType flag;
2592 ds_evaluation_type profileType;
2594 LockDefaultOpenCLEnv();
2598 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2599 ,
sizeof(MagickBooleanType), &flag, exception);
2602 OpenCLLib=GetOpenCLLib();
2603 if (OpenCLLib==NULL)
2605 mStatus=InitOpenCLEnvInternal(clEnv, exception);
2609 clEnv->library=OpenCLLib;
2611 status = initDSProfile(&profile, IMAGEMAGICK_PROFILE_VERSION);
2612 if (status!=DS_SUCCESS) {
2613 (void) ThrowMagickException(exception, GetMagickModule(), ModuleFatalError,
"Error when initializing the profile",
"'%s'",
".");
2617 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
2618 ,GetOpenCLCachedFilesDirectory()
2619 ,DirectorySeparator,IMAGEMAGICK_PROFILE_FILE);
2621 if (canWriteProfileToFile(path) != DS_SUCCESS) {
2625 bestDeviceIndex = 0;
2626 for (i = 1; i < profile->numDevices; i++) {
2627 if ((profile->devices[i].type == DS_DEVICE_OPENCL_DEVICE) && (profile->devices[i].oclDeviceType == CL_DEVICE_TYPE_GPU)) {
2628 bestDeviceIndex = i;
2634 if (clEnv->regenerateProfile != MagickFalse) {
2635 profileType = DS_EVALUATE_ALL;
2638 readProfileFromFile(profile, AccelerateScoreDeserializer, path);
2639 profileType = DS_EVALUATE_NEW_ONLY;
2641 status = profileDevices(profile, profileType, AcceleratePerfEvaluator, NULL, &numDeviceProfiled);
2643 if (status!=DS_SUCCESS) {
2644 (void) ThrowMagickException(exception, GetMagickModule(), ModuleFatalError,
"Error when initializing the profile",
"'%s'",
".");
2647 if (numDeviceProfiled > 0) {
2648 status = writeProfileToFile(profile, AccelerateScoreSerializer, path);
2649 if (status!=DS_SUCCESS) {
2650 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"Error when saving the profile into a file",
"'%s'",
".");
2655 bestDeviceIndex = 0;
2656 bestScore = *(AccelerateScoreType*)profile->devices[bestDeviceIndex].score;
2657 for (i = 1; i < profile->numDevices; i++) {
2658 AccelerateScoreType score = *(AccelerateScoreType*)profile->devices[i].score;
2659 if (score < bestScore) {
2660 bestDeviceIndex = i;
2667 if (profile->devices[bestDeviceIndex].type == DS_DEVICE_NATIVE_CPU) {
2670 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2671 ,
sizeof(MagickBooleanType), &flag, exception);
2673 else if (profile->devices[bestDeviceIndex].type == DS_DEVICE_OPENCL_DEVICE) {
2676 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2677 ,
sizeof(MagickBooleanType), &flag, exception);
2678 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2679 ,
sizeof(cl_device_id), &profile->devices[bestDeviceIndex].oclDeviceID,exception);
2682 status = DS_PERF_EVALUATOR_ERROR;
2685 mStatus=InitOpenCLEnvInternal(clEnv, exception);
2687 status = releaseDSProfile(profile, AccelerateScoreRelease);
2688 if (status!=DS_SUCCESS) {
2689 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"Error when releasing the profile",
"'%s'",
".");
2694 UnlockDefaultOpenCLEnv();
2732 MagickExport MagickBooleanType InitImageMagickOpenCL(
2733 ImageMagickOpenCLMode mode,
void *userSelectedDevice,
void *selectedDevice,
2736 MagickBooleanType status = MagickFalse;
2738 MagickBooleanType flag;
2740 clEnv = GetDefaultOpenCLEnv();
2744 case MAGICK_OPENCL_OFF:
2746 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2747 ,
sizeof(MagickBooleanType), &flag, exception);
2748 status = InitOpenCLEnv(clEnv, exception);
2751 *(cl_device_id*)selectedDevice = NULL;
2754 case MAGICK_OPENCL_DEVICE_SELECT_USER:
2756 if (userSelectedDevice == NULL)
2760 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2761 ,
sizeof(MagickBooleanType), &flag, exception);
2763 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2764 ,
sizeof(cl_device_id), userSelectedDevice,exception);
2766 status = InitOpenCLEnv(clEnv, exception);
2767 if (selectedDevice) {
2768 GetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2769 ,
sizeof(cl_device_id), selectedDevice, exception);
2773 case MAGICK_OPENCL_DEVICE_SELECT_AUTO_CLEAR_CACHE:
2775 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED
2776 ,
sizeof(MagickBooleanType), &flag, exception);
2778 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE
2779 ,
sizeof(MagickBooleanType), &flag, exception);
2782 case MAGICK_OPENCL_DEVICE_SELECT_AUTO:
2785 cl_device_id d = NULL;
2787 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2788 ,
sizeof(MagickBooleanType), &flag, exception);
2789 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2790 ,
sizeof(cl_device_id), &d,exception);
2791 status = InitOpenCLEnv(clEnv, exception);
2792 if (selectedDevice) {
2793 GetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2794 ,
sizeof(cl_device_id), selectedDevice, exception);
2805 MagickBooleanType OpenCLThrowMagickException(
ExceptionInfo *exception,
2806 const char *module,
const char *
function,
const size_t line,
2807 const ExceptionType severity,
const char *tag,
const char *format,...) {
2813 status = MagickTrue;
2815 clEnv = GetDefaultOpenCLEnv();
2818 assert(exception->signature == MagickCoreSignature);
2821 cl_device_type dType;
2822 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_TYPE ,
sizeof(cl_device_type),&dType,NULL);
2823 if (dType == CL_DEVICE_TYPE_CPU) {
2824 char buffer[MaxTextExtent];
2825 clEnv->library->clGetPlatformInfo(clEnv->platform, CL_PLATFORM_NAME, MaxTextExtent, buffer, NULL);
2829 if (strncmp(buffer,
"Intel",5) == 0) {
2831 InitImageMagickOpenCL(MAGICK_OPENCL_OFF, NULL, NULL, exception);
2836 #ifdef OPENCLLOG_ENABLED
2840 va_start(operands,format);
2841 status=ThrowMagickExceptionList(exception,module,
function,line,severity,tag, format,operands);
2845 magick_unreferenced(module);
2846 magick_unreferenced(
function);
2847 magick_unreferenced(line);
2848 magick_unreferenced(tag);
2849 magick_unreferenced(format);
2855 char* openclCachedFilesDirectory;
2859 const char* GetOpenCLCachedFilesDirectory() {
2860 if (openclCachedFilesDirectory == NULL) {
2861 if (openclCachedFilesDirectoryLock == NULL)
2863 ActivateSemaphoreInfo(&openclCachedFilesDirectoryLock);
2865 LockSemaphoreInfo(openclCachedFilesDirectoryLock);
2866 if (openclCachedFilesDirectory == NULL) {
2867 char path[MaxTextExtent];
2870 struct stat attributes;
2871 MagickBooleanType status;
2872 int mkdirStatus = 0;
2876 home=GetEnvironmentValue(
"MAGICK_OPENCL_CACHE_DIR");
2877 if (home == (
char *) NULL)
2879 home=GetEnvironmentValue(
"XDG_CACHE_HOME");
2880 #if defined(MAGICKCORE_WINDOWS_SUPPORT) || defined(__MINGW32__)
2881 if (home == (
char *) NULL)
2882 home=GetEnvironmentValue(
"LOCALAPPDATA");
2883 if (home == (
char *) NULL)
2884 home=GetEnvironmentValue(
"APPDATA");
2885 if (home == (
char *) NULL)
2886 home=GetEnvironmentValue(
"USERPROFILE");
2890 if (home != (
char *) NULL)
2893 (void) FormatLocaleString(path,MaxTextExtent,
"%s",home);
2894 status=GetPathAttributes(path,&attributes);
2895 if (status == MagickFalse)
2898 #ifdef MAGICKCORE_WINDOWS_SUPPORT
2899 mkdirStatus = mkdir(path);
2901 mkdirStatus = mkdir(path, 0777);
2908 (void) FormatLocaleString(path,MaxTextExtent,
2909 "%s%sImageMagick",home,DirectorySeparator);
2911 status=GetPathAttributes(path,&attributes);
2912 if (status == MagickFalse)
2914 #ifdef MAGICKCORE_WINDOWS_SUPPORT
2915 mkdirStatus = mkdir(path);
2917 mkdirStatus = mkdir(path, 0777);
2924 temp = (
char*)AcquireCriticalMemory(strlen(path)+1);
2925 CopyMagickString(temp,path,strlen(path)+1);
2927 home=DestroyString(home);
2929 home=GetEnvironmentValue(
"HOME");
2930 if (home != (
char *) NULL)
2936 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s.cache",
2937 home,DirectorySeparator);
2938 status=GetPathAttributes(path,&attributes);
2939 if (status == MagickFalse)
2942 #ifdef MAGICKCORE_WINDOWS_SUPPORT
2943 mkdirStatus = mkdir(path);
2945 mkdirStatus = mkdir(path, 0777);
2952 (void) FormatLocaleString(path,MaxTextExtent,
2953 "%s%s.cache%sImageMagick",home,DirectorySeparator,
2954 DirectorySeparator);
2956 status=GetPathAttributes(path,&attributes);
2957 if (status == MagickFalse)
2959 #ifdef MAGICKCORE_WINDOWS_SUPPORT
2960 mkdirStatus = mkdir(path);
2962 mkdirStatus = mkdir(path, 0777);
2969 temp = (
char*)AcquireCriticalMemory(strlen(path)+1);
2970 CopyMagickString(temp,path,strlen(path)+1);
2972 home=DestroyString(home);
2975 openclCachedFilesDirectory = temp;
2977 UnlockSemaphoreInfo(openclCachedFilesDirectoryLock);
2979 return openclCachedFilesDirectory;
2984 void OpenCLLog(
const char* message) {
2986 #ifdef OPENCLLOG_ENABLED
2987 #define OPENCL_LOG_FILE "ImageMagickOpenCL.log"
2990 if (getenv(
"MAGICK_OCL_LOG"))
2993 char path[MaxTextExtent];
2994 unsigned long allocSize;
2998 clEnv = GetDefaultOpenCLEnv();
3001 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
3002 ,GetOpenCLCachedFilesDirectory()
3003 ,DirectorySeparator,OPENCL_LOG_FILE);
3006 log = fopen(path,
"ab");
3007 if (log == (FILE *) NULL)
3009 fwrite(message,
sizeof(
char), strlen(message), log);
3010 fwrite(
"\n",
sizeof(
char), 1, log);
3012 if (clEnv->OpenCLInitialized && !clEnv->OpenCLDisabled)
3014 allocSize = GetOpenCLDeviceMaxMemAllocSize(clEnv);
3015 fprintf(log,
"Devic Max Memory Alloc Size: %lu\n", allocSize);
3022 magick_unreferenced(message);
3026 MagickPrivate
void OpenCLTerminus()
3029 if (openclCachedFilesDirectory != (
char *) NULL)
3030 openclCachedFilesDirectory=DestroyString(openclCachedFilesDirectory);
3032 DestroySemaphoreInfo(&openclCachedFilesDirectoryLock);
3035 (void) RelinquishMagickOpenCLEnv(defaultCLEnv);
3039 DestroySemaphoreInfo(&defaultCLEnvLock);
3040 if (OpenCLLib != (MagickLibrary *)NULL)
3042 if (OpenCLLib->base != (
void *) NULL)
3043 (
void) lt_dlclose(OpenCLLib->base);
3044 OpenCLLib=(MagickLibrary *)RelinquishMagickMemory(OpenCLLib);
3047 DestroySemaphoreInfo(&OpenCLLibLock);
3053 MagickBooleanType OpenCLInitialized;
3064 MagickExport MagickBooleanType SetMagickOpenCLEnvParam(
3065 MagickCLEnv magick_unused(clEnv),MagickOpenCLEnvParam magick_unused(param),
3066 size_t magick_unused(dataSize),
void *magick_unused(data),
3069 magick_unreferenced(clEnv);
3070 magick_unreferenced(param);
3071 magick_unreferenced(dataSize);
3072 magick_unreferenced(data);
3073 magick_unreferenced(exception);
3074 return(MagickFalse);
3077 MagickExport MagickBooleanType GetMagickOpenCLEnvParam(
3078 MagickCLEnv magick_unused(clEnv),MagickOpenCLEnvParam magick_unused(param),
3079 size_t magick_unused(dataSize),
void *magick_unused(data),
3082 magick_unreferenced(clEnv);
3083 magick_unreferenced(param);
3084 magick_unreferenced(dataSize);
3085 magick_unreferenced(data);
3086 magick_unreferenced(exception);
3087 return(MagickFalse);
3090 MagickExport MagickBooleanType InitOpenCLEnv(
MagickCLEnv magick_unused(clEnv),
3093 magick_unreferenced(clEnv);
3094 magick_unreferenced(exception);
3095 return(MagickFalse);
3098 MagickExport MagickBooleanType InitImageMagickOpenCL(
3099 ImageMagickOpenCLMode magick_unused(mode),
3100 void *magick_unused(userSelectedDevice),
void *magick_unused(selectedDevice),
3103 magick_unreferenced(mode);
3104 magick_unreferenced(userSelectedDevice);
3105 magick_unreferenced(selectedDevice);
3106 magick_unreferenced(exception);
3107 return(MagickFalse);