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/mime-private.h"
62 #include "magick/memory_.h"
63 #include "magick/memory-private.h"
64 #include "magick/monitor.h"
65 #include "magick/montage.h"
66 #include "magick/morphology.h"
67 #include "magick/nt-base.h"
68 #include "magick/nt-base-private.h"
69 #include "magick/opencl.h"
70 #include "magick/opencl-private.h"
71 #include "magick/option.h"
72 #include "magick/policy.h"
73 #include "magick/property.h"
74 #include "magick/quantize.h"
75 #include "magick/quantum.h"
76 #include "magick/random_.h"
77 #include "magick/random-private.h"
78 #include "magick/resample.h"
79 #include "magick/resource_.h"
80 #include "magick/splay-tree.h"
81 #include "magick/semaphore.h"
82 #include "magick/statistic.h"
83 #include "magick/string_.h"
84 #include "magick/token.h"
85 #include "magick/utility.h"
86 #include "magick/utility-private.h"
88 #ifdef MAGICKCORE_CLPERFMARKER
89 #include "CLPerfMarker.h"
93 #if defined(MAGICKCORE_OPENCL_SUPPORT)
95 #define NUM_CL_RAND_GENERATORS 1024
96 #define PROFILE_OCL_KERNELS 0
104 } KernelProfileRecord;
106 static const char *kernelNames[] = {
120 "LocalContrastBlurRow",
121 "LocalContrastBlurApplyColumn",
125 "RandomNumberGenerator",
128 "UnsharpMaskBlurColumn",
134 profileRecords[KERNEL_COUNT];
136 typedef struct _AccelerateTimer {
142 void startAccelerateTimer(AccelerateTimer* timer) {
144 QueryPerformanceCounter((LARGE_INTEGER*)&timer->_start);
150 timer->_start = (
long long)s.tv_sec * (
long long)1.0E3 + (
long long)s.tv_usec / (
long long)1.0E3;
154 void stopAccelerateTimer(AccelerateTimer* timer) {
157 QueryPerformanceCounter((LARGE_INTEGER*)&(n));
161 n = (
long long)s.tv_sec * (
long long)1.0E3+ (
long long)s.tv_usec / (
long long)1.0E3;
168 void resetAccelerateTimer(AccelerateTimer* timer) {
173 void initAccelerateTimer(AccelerateTimer* timer) {
175 QueryPerformanceFrequency((LARGE_INTEGER*)&timer->_freq);
177 timer->_freq = (
long long)1.0E3;
179 resetAccelerateTimer(timer);
182 double readAccelerateTimer(AccelerateTimer* timer) {
183 return (
double)timer->_clocks/(double)timer->_freq;
186 MagickPrivate MagickBooleanType RecordProfileData(
MagickCLEnv clEnv, ProfiledKernels kernel, cl_event event)
188 #if PROFILE_OCL_KERNELS
192 cl_ulong elapsed = 0;
193 clEnv->library->clWaitForEvents(1, &event);
194 status = clEnv->library->clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_START,
sizeof(cl_ulong), &start, NULL);
195 status &= clEnv->library->clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_END,
sizeof(cl_ulong), &end, NULL);
196 if (status == CL_SUCCESS) {
199 elapsed = end - start;
201 LockSemaphoreInfo(clEnv->commandQueuesLock);
202 if ((elapsed < profileRecords[kernel].min) || (profileRecords[kernel].count == 0))
203 profileRecords[kernel].min = elapsed;
204 if (elapsed > profileRecords[kernel].max)
205 profileRecords[kernel].max = elapsed;
206 profileRecords[kernel].total += elapsed;
207 profileRecords[kernel].count += 1;
208 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
212 magick_unreferenced(clEnv);
213 magick_unreferenced(kernel);
214 magick_unreferenced(event);
219 void DumpProfileData()
221 #if PROFILE_OCL_KERNELS
224 OpenCLLog(
"====================================================");
235 clEnv = GetDefaultOpenCLEnv();
237 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_VENDOR, 2048, buff, NULL);
240 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_NAME, 2048, buff, NULL);
243 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DRIVER_VERSION, 2048, buff, NULL);
247 OpenCLLog(
"====================================================");
248 OpenCLLog(
" ave\tcalls \tmin -> max");
249 OpenCLLog(
" ---\t----- \t----------");
250 for (i = 0; i < KERNEL_COUNT; ++i) {
254 strncpy(indent, kernelNames[i], min(strlen(kernelNames[i]), strlen(indent) - 1));
255 sprintf(buf,
"%s%d\t(%d calls) \t%d -> %d", indent, profileRecords[i].count > 0 ? (profileRecords[i].total / profileRecords[i].count) : 0, profileRecords[i].count, profileRecords[i].min, profileRecords[i].max);
261 OpenCLLog(
"====================================================");
270 #ifdef MAGICKCORE_WINDOWS_SUPPORT
276 void *OsLibraryLoad(
const char *libraryName)
278 #ifdef MAGICKCORE_WINDOWS_SUPPORT
279 return (
void *)LoadLibraryA(libraryName);
281 return (
void *)dlopen(libraryName, RTLD_NOW);
286 void *OsLibraryGetFunctionAddress(
void *library,
const char *functionName)
288 #ifdef MAGICKCORE_WINDOWS_SUPPORT
289 if (!library || !functionName)
293 return (
void *) GetProcAddress( (HMODULE)library, functionName);
295 if (!library || !functionName)
299 return (
void *)dlsym(library, functionName);
326 clEnv->commandQueuesPos=-1;
327 ActivateSemaphoreInfo(&clEnv->lock);
328 ActivateSemaphoreInfo(&clEnv->commandQueuesLock);
357 MagickPrivate MagickBooleanType RelinquishMagickOpenCLEnv(
MagickCLEnv clEnv)
361 while (clEnv->commandQueuesPos >= 0)
363 clEnv->library->clReleaseCommandQueue(
364 clEnv->commandQueues[clEnv->commandQueuesPos--]);
366 if (clEnv->programs[0] != (cl_program) NULL)
367 (void) clEnv->library->clReleaseProgram(clEnv->programs[0]);
368 if (clEnv->context != (cl_context) NULL)
369 clEnv->library->clReleaseContext(clEnv->context);
370 DestroySemaphoreInfo(&clEnv->lock);
371 DestroySemaphoreInfo(&clEnv->commandQueuesLock);
372 RelinquishMagickMemory(clEnv);
388 MagickLibrary * OpenCLLib;
392 static MagickBooleanType bindOpenCLFunctions(
void* library)
394 #ifdef MAGICKCORE_HAVE_OPENCL_CL_H
395 #define BIND(X) OpenCLLib->X= &X;
398 if ((OpenCLLib->X=(MAGICKpfn_##X)OsLibraryGetFunctionAddress(library,#X)) == NULL)\
402 BIND(clGetPlatformIDs);
403 BIND(clGetPlatformInfo);
405 BIND(clGetDeviceIDs);
406 BIND(clGetDeviceInfo);
408 BIND(clCreateContext);
409 BIND(clReleaseContext);
411 BIND(clCreateBuffer);
412 BIND(clRetainMemObject);
413 BIND(clReleaseMemObject);
415 BIND(clCreateProgramWithSource);
416 BIND(clCreateProgramWithBinary);
417 BIND(clBuildProgram);
418 BIND(clReleaseProgram);
419 BIND(clGetProgramInfo);
420 BIND(clGetProgramBuildInfo);
422 BIND(clCreateKernel);
423 BIND(clReleaseKernel);
424 BIND(clSetKernelArg);
429 BIND(clEnqueueNDRangeKernel);
430 BIND(clEnqueueReadBuffer);
431 BIND(clEnqueueMapBuffer);
432 BIND(clEnqueueUnmapMemObject);
434 BIND(clCreateCommandQueue);
435 BIND(clReleaseCommandQueue);
437 BIND(clGetEventProfilingInfo);
438 BIND(clGetEventInfo);
439 BIND(clWaitForEvents);
440 BIND(clReleaseEvent);
442 BIND(clSetEventCallback);
447 MagickLibrary * GetOpenCLLib()
449 if (OpenCLLib == NULL)
451 if (OpenCLLibLock == NULL)
453 ActivateSemaphoreInfo(&OpenCLLibLock);
456 LockSemaphoreInfo(OpenCLLibLock);
458 OpenCLLib = (MagickLibrary *) AcquireMagickMemory (
sizeof (MagickLibrary));
460 if (OpenCLLib != NULL)
462 MagickBooleanType status = MagickFalse;
463 void * library = NULL;
465 #ifdef MAGICKCORE_OPENCL_MACOSX
466 status = bindOpenCLFunctions(library);
469 memset(OpenCLLib, 0,
sizeof(MagickLibrary));
470 #ifdef MAGICKCORE_WINDOWS_SUPPORT
471 library = OsLibraryLoad(
"OpenCL.dll");
473 library = OsLibraryLoad(
"libOpenCL.so");
476 status = bindOpenCLFunctions(library);
478 if (status==MagickTrue)
479 OpenCLLib->base=library;
481 OpenCLLib=(MagickLibrary *)RelinquishMagickMemory(OpenCLLib);
485 UnlockSemaphoreInfo(OpenCLLibLock);
518 if (defaultCLEnv == NULL)
520 if (defaultCLEnvLock == NULL)
522 ActivateSemaphoreInfo(&defaultCLEnvLock);
524 LockSemaphoreInfo(defaultCLEnvLock);
525 if (defaultCLEnv == NULL)
526 defaultCLEnv = AcquireMagickOpenCLEnv();
527 UnlockSemaphoreInfo(defaultCLEnvLock);
532 static void LockDefaultOpenCLEnv() {
533 if (defaultCLEnvLock == NULL)
535 ActivateSemaphoreInfo(&defaultCLEnvLock);
537 LockSemaphoreInfo(defaultCLEnvLock);
540 static void UnlockDefaultOpenCLEnv() {
541 if (defaultCLEnvLock == NULL)
543 ActivateSemaphoreInfo(&defaultCLEnvLock);
546 UnlockSemaphoreInfo(defaultCLEnvLock);
576 LockDefaultOpenCLEnv();
577 oldEnv = defaultCLEnv;
578 defaultCLEnv = clEnv;
579 UnlockDefaultOpenCLEnv();
616 static MagickBooleanType SetMagickOpenCLEnvParamInternal(
MagickCLEnv clEnv, MagickOpenCLEnvParam param
619 MagickBooleanType status = MagickFalse;
627 case MAGICK_OPENCL_ENV_PARAM_DEVICE:
628 if (dataSize !=
sizeof(clEnv->device))
630 clEnv->device = *((cl_device_id*)data);
631 clEnv->OpenCLInitialized = MagickFalse;
635 case MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED:
636 if (dataSize !=
sizeof(clEnv->OpenCLDisabled))
638 clEnv->OpenCLDisabled = *((MagickBooleanType*)data);
639 clEnv->OpenCLInitialized = MagickFalse;
643 case MAGICK_OPENCL_ENV_PARAM_OPENCL_INITIALIZED:
644 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"SetMagickOpenCLEnvParm cannot modify the OpenCL initialization state.",
"'%s'",
".");
647 case MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED:
648 if (dataSize !=
sizeof(clEnv->disableProgramCache))
650 clEnv->disableProgramCache = *((MagickBooleanType*)data);
651 clEnv->OpenCLInitialized = MagickFalse;
655 case MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE:
656 if (dataSize !=
sizeof(clEnv->regenerateProfile))
658 clEnv->regenerateProfile = *((MagickBooleanType*)data);
659 clEnv->OpenCLInitialized = MagickFalse;
672 MagickBooleanType SetMagickOpenCLEnvParam(
MagickCLEnv clEnv, MagickOpenCLEnvParam param
674 MagickBooleanType status = MagickFalse;
676 LockSemaphoreInfo(clEnv->lock);
677 status = SetMagickOpenCLEnvParamInternal(clEnv,param,dataSize,data,exception);
678 UnlockSemaphoreInfo(clEnv->lock);
717 MagickBooleanType GetMagickOpenCLEnvParam(
MagickCLEnv clEnv, MagickOpenCLEnvParam param
726 magick_unreferenced(exception);
728 status = MagickFalse;
736 case MAGICK_OPENCL_ENV_PARAM_DEVICE:
737 if (dataSize !=
sizeof(cl_device_id))
739 *((cl_device_id*)data) = clEnv->device;
743 case MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED:
744 if (dataSize !=
sizeof(clEnv->OpenCLDisabled))
746 *((MagickBooleanType*)data) = clEnv->OpenCLDisabled;
750 case MAGICK_OPENCL_ENV_PARAM_OPENCL_INITIALIZED:
751 if (dataSize !=
sizeof(clEnv->OpenCLDisabled))
753 *((MagickBooleanType*)data) = clEnv->OpenCLInitialized;
757 case MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED:
758 if (dataSize !=
sizeof(clEnv->disableProgramCache))
760 *((MagickBooleanType*)data) = clEnv->disableProgramCache;
764 case MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE:
765 if (dataSize !=
sizeof(clEnv->regenerateProfile))
767 *((MagickBooleanType*)data) = clEnv->regenerateProfile;
771 case MAGICK_OPENCL_ENV_PARAM_PLATFORM_VENDOR:
772 if (dataSize !=
sizeof(
char *))
774 clEnv->library->clGetPlatformInfo(clEnv->platform,CL_PLATFORM_VENDOR,0,
776 *((
char **) data)=(
char *) AcquireQuantumMemory(length,
sizeof(
char));
777 clEnv->library->clGetPlatformInfo(clEnv->platform,CL_PLATFORM_VENDOR,
778 length,*((
char **) data),NULL);
782 case MAGICK_OPENCL_ENV_PARAM_DEVICE_NAME:
783 if (dataSize !=
sizeof(
char *))
785 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_NAME,0,NULL,
787 *((
char **) data)=(
char *) AcquireQuantumMemory(length,
sizeof(
char));
788 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_NAME,length,
789 *((
char **) data),NULL);
830 return clEnv->context;
833 static char* getBinaryCLProgramName(
MagickCLEnv clEnv, MagickOpenCLProgram prog,
unsigned int signature)
837 char path[MaxTextExtent];
838 char deviceName[MaxTextExtent];
839 const char* prefix =
"magick_opencl";
840 clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_NAME, MaxTextExtent, deviceName, NULL);
845 if ( *ptr ==
' ' || *ptr ==
'\\' || *ptr ==
'/' || *ptr ==
':' || *ptr ==
'*'
846 || *ptr ==
'?' || *ptr ==
'"' || *ptr ==
'<' || *ptr ==
'>' || *ptr ==
'|')
852 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s_%s_%02d_%08x_%.20g.bin",
853 GetOpenCLCachedFilesDirectory(),DirectorySeparator,prefix,deviceName,
854 (
unsigned int) prog,signature,(
double)
sizeof(
char*)*8);
855 name = (
char*)AcquireMagickMemory(strlen(path)+1);
856 CopyMagickString(name,path,strlen(path)+1);
860 static void saveBinaryCLProgram(
MagickCLEnv clEnv,MagickOpenCLProgram prog,
877 filename=getBinaryCLProgramName(clEnv,prog,signature);
878 status=clEnv->library->clGetProgramInfo(clEnv->programs[prog],
879 CL_PROGRAM_NUM_DEVICES,
sizeof(cl_uint),&num_devices,NULL);
880 if (status != CL_SUCCESS)
882 size=num_devices*
sizeof(*program_sizes);
883 program_sizes=(
size_t*) AcquireQuantumMemory(1,size);
884 if (program_sizes == (
size_t*) NULL)
886 status=clEnv->library->clGetProgramInfo(clEnv->programs[prog],
887 CL_PROGRAM_BINARY_SIZES,size,program_sizes,NULL);
888 if (status == CL_SUCCESS)
896 binary_program_size=num_devices*
sizeof(*binary_program);
897 binary_program=(
unsigned char **) AcquireQuantumMemory(1,
898 binary_program_size);
899 if (binary_program == (
unsigned char **) NULL)
901 program_sizes=(
size_t *) RelinquishMagickMemory(program_sizes);
904 for (i = 0; i < num_devices; i++)
906 binary_program[i]=AcquireQuantumMemory(MagickMax(*(program_sizes+i),1),
907 sizeof(**binary_program));
908 if (binary_program[i] == (
unsigned char *) NULL)
910 status=CL_OUT_OF_HOST_MEMORY;
914 if (status == CL_SUCCESS)
915 status=clEnv->library->clGetProgramInfo(clEnv->programs[prog],
916 CL_PROGRAM_BINARIES,binary_program_size,binary_program,NULL);
917 if (status == CL_SUCCESS)
919 for (i = 0; i < num_devices; i++)
927 program_size=*(program_sizes+i);
928 if (program_size < 1)
930 file=open_utf8(filename,O_WRONLY | O_CREAT | O_BINARY,S_MODE);
933 write(file,binary_program[i],program_size);
937 (
void) ThrowMagickException(exception,GetMagickModule(),
938 DelegateWarning,
"Saving kernel failed.",
"`%s'",filename);
942 for (i = 0; i < num_devices; i++)
943 binary_program[i]=(
unsigned char *) RelinquishMagickMemory(
945 binary_program=(
unsigned char **) RelinquishMagickMemory(binary_program);
947 program_sizes=(
size_t *) RelinquishMagickMemory(program_sizes);
950 static MagickBooleanType loadBinaryCLProgram(
MagickCLEnv clEnv, MagickOpenCLProgram prog,
unsigned int signature)
952 MagickBooleanType loadSuccessful;
953 unsigned char* binaryProgram;
954 char* binaryFileName;
957 #ifdef MAGICKCORE_CLPERFMARKER
958 clBeginPerfMarkerAMD(__FUNCTION__,
"");
961 binaryProgram = NULL;
962 binaryFileName = NULL;
964 loadSuccessful = MagickFalse;
966 binaryFileName = getBinaryCLProgramName(clEnv, prog, signature);
967 fileHandle = fopen(binaryFileName,
"rb");
968 if (fileHandle != NULL)
973 cl_int clBinaryStatus;
977 b_error |= fseek( fileHandle, 0, SEEK_END ) < 0;
978 b_error |= ( length = ftell( fileHandle ) ) <= 0;
979 b_error |= fseek( fileHandle, 0, SEEK_SET ) < 0;
983 binaryProgram = (
unsigned char*)AcquireMagickMemory(length);
984 if (binaryProgram == NULL)
987 memset(binaryProgram, 0, length);
988 b_error |= fread(binaryProgram, 1, length, fileHandle) != length;
990 clEnv->programs[prog] = clEnv->library->clCreateProgramWithBinary(clEnv->context, 1, &clEnv->device, &length, (
const unsigned char**)&binaryProgram, &clBinaryStatus, &clStatus);
991 if (clStatus != CL_SUCCESS
992 || clBinaryStatus != CL_SUCCESS)
995 loadSuccessful = MagickTrue;
999 if (fileHandle != NULL)
1001 if (binaryFileName != NULL)
1002 RelinquishMagickMemory(binaryFileName);
1003 if (binaryProgram != NULL)
1004 RelinquishMagickMemory(binaryProgram);
1006 #ifdef MAGICKCORE_CLPERFMARKER
1007 clEndPerfMarkerAMD();
1010 return loadSuccessful;
1013 static unsigned int stringSignature(
const char*
string)
1015 unsigned int stringLength;
1017 unsigned int signature;
1021 const unsigned int* u;
1024 #ifdef MAGICKCORE_CLPERFMARKER
1025 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1028 stringLength = (
unsigned int) strlen(
string);
1029 signature = stringLength;
1030 n = stringLength/
sizeof(
unsigned int);
1032 for (i = 0; i < n; i++)
1036 if (n *
sizeof(
unsigned int) != stringLength)
1039 j = n *
sizeof(
unsigned int);
1040 for (i = 0; i < 4; i++,j++)
1042 if (j < stringLength)
1051 #ifdef MAGICKCORE_CLPERFMARKER
1052 clEndPerfMarkerAMD();
1059 extern const char *accelerateKernels, *accelerateKernels2;
1063 MagickBooleanType status = MagickFalse;
1066 char* accelerateKernelsBuffer = NULL;
1069 const char* MagickOpenCLProgramStrings[MAGICK_OPENCL_NUM_PROGRAMS];
1071 char options[MaxTextExtent];
1072 unsigned int optionsSignature;
1074 #ifdef MAGICKCORE_CLPERFMARKER
1075 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1079 (void) FormatLocaleString(options, MaxTextExtent, CLOptions, (
float)QuantumRange,
1080 (float)QuantumScale, (
float)CLCharQuantumScale, (
float)MagickEpsilon, (
float)MagickPI, (
unsigned int)MaxMap, (
unsigned int)MAGICKCORE_QUANTUM_DEPTH);
1095 optionsSignature = stringSignature(options);
1098 accelerateKernelsBuffer = (
char*) AcquireQuantumMemory(1,strlen(accelerateKernels)+strlen(accelerateKernels2)+1);
1099 sprintf(accelerateKernelsBuffer,"%s%s",accelerateKernels,accelerateKernels2);
1100 MagickOpenCLProgramStrings[MAGICK_OPENCL_ACCELERATE] = accelerateKernelsBuffer;
1102 for (i = 0; i < MAGICK_OPENCL_NUM_PROGRAMS; i++)
1104 MagickBooleanType loadSuccessful = MagickFalse;
1105 unsigned int programSignature = stringSignature(MagickOpenCLProgramStrings[i]) ^ optionsSignature;
1108 if (clEnv->disableProgramCache != MagickTrue
1109 && !getenv(
"MAGICK_OCL_REC"))
1110 loadSuccessful = loadBinaryCLProgram(clEnv, (MagickOpenCLProgram)i, programSignature);
1112 if (loadSuccessful == MagickFalse)
1115 size_t programLength = strlen(MagickOpenCLProgramStrings[i]);
1116 clEnv->programs[i] = clEnv->library->clCreateProgramWithSource(clEnv->context, 1, &(MagickOpenCLProgramStrings[i]), &programLength, &clStatus);
1117 if (clStatus!=CL_SUCCESS)
1119 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1120 "clCreateProgramWithSource failed.",
"(%d)", (int)clStatus);
1126 clStatus = clEnv->library->clBuildProgram(clEnv->programs[i], 1, &clEnv->device, options, NULL, NULL);
1127 if (clStatus!=CL_SUCCESS)
1129 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1130 "clBuildProgram failed.",
"(%d)", (int)clStatus);
1132 if (loadSuccessful == MagickFalse)
1134 char path[MaxTextExtent];
1138 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
1139 ,GetOpenCLCachedFilesDirectory()
1140 ,DirectorySeparator,
"magick_badcl.cl");
1141 fileHandle = fopen(path,
"wb");
1142 if (fileHandle != NULL)
1144 fwrite(MagickOpenCLProgramStrings[i],
sizeof(
char), strlen(MagickOpenCLProgramStrings[i]), fileHandle);
1152 clEnv->library->clGetProgramBuildInfo(clEnv->programs[i], clEnv->device, CL_PROGRAM_BUILD_LOG, 0, NULL, &logSize);
1153 log = (
char*)AcquireCriticalMemory(logSize);
1154 clEnv->library->clGetProgramBuildInfo(clEnv->programs[i], clEnv->device, CL_PROGRAM_BUILD_LOG, logSize, log, &logSize);
1156 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
1157 ,GetOpenCLCachedFilesDirectory()
1158 ,DirectorySeparator,
"magick_badcl_build.log");
1159 fileHandle = fopen(path,
"wb");
1160 if (fileHandle != NULL)
1162 const char* buildOptionsTitle =
"build options: ";
1163 fwrite(buildOptionsTitle,
sizeof(
char), strlen(buildOptionsTitle), fileHandle);
1164 fwrite(options,
sizeof(
char), strlen(options), fileHandle);
1165 fwrite(
"\n",
sizeof(
char), 1, fileHandle);
1166 fwrite(log,
sizeof(
char), logSize, fileHandle);
1169 RelinquishMagickMemory(log);
1175 if (loadSuccessful == MagickFalse)
1178 saveBinaryCLProgram(clEnv, (MagickOpenCLProgram)i, programSignature, exception);
1182 status = MagickTrue;
1186 if (accelerateKernelsBuffer!=NULL) RelinquishMagickMemory(accelerateKernelsBuffer);
1188 #ifdef MAGICKCORE_CLPERFMARKER
1189 clEndPerfMarkerAMD();
1198 cl_uint numPlatforms = 0;
1199 cl_platform_id *platforms = NULL;
1200 char* MAGICK_OCL_DEVICE = NULL;
1201 MagickBooleanType OpenCLAvailable = MagickFalse;
1203 #ifdef MAGICKCORE_CLPERFMARKER
1204 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1208 MAGICK_OCL_DEVICE = getenv(
"MAGICK_OCL_DEVICE");
1209 if (MAGICK_OCL_DEVICE == (
char *) NULL)
1210 return(MagickFalse);
1211 if (strcmp(MAGICK_OCL_DEVICE,
"CPU") == 0)
1212 clEnv->deviceType = CL_DEVICE_TYPE_CPU;
1213 else if (strcmp(MAGICK_OCL_DEVICE,
"GPU") == 0)
1214 clEnv->deviceType = CL_DEVICE_TYPE_GPU;
1215 else if (IsStringTrue(MAGICK_OCL_DEVICE) != MagickFalse)
1217 if (clEnv->deviceType == 0)
1218 clEnv->deviceType = CL_DEVICE_TYPE_ALL;
1221 return(MagickFalse);
1223 if (clEnv->device != NULL)
1225 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_PLATFORM,
sizeof(cl_platform_id), &clEnv->platform, NULL);
1226 if (status != CL_SUCCESS) {
1227 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1228 "Failed to get OpenCL platform from the selected device.",
"(%d)", status);
1232 else if (clEnv->platform != NULL)
1235 platforms = (cl_platform_id *) AcquireQuantumMemory(1,numPlatforms *
sizeof(cl_platform_id));
1236 if (platforms == (cl_platform_id *) NULL)
1238 (void) ThrowMagickException(exception, GetMagickModule(), ResourceLimitError,
1239 "AcquireMagickMemory failed.",
".");
1242 platforms[0] = clEnv->platform;
1246 clEnv->device = NULL;
1249 status = clEnv->library->clGetPlatformIDs(0, NULL, &numPlatforms);
1250 if (status != CL_SUCCESS)
1252 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1253 "clGetplatformIDs failed.",
"(%d)", status);
1258 if (numPlatforms == 0) {
1262 platforms = (cl_platform_id *) AcquireQuantumMemory(1,numPlatforms *
sizeof(cl_platform_id));
1263 if (platforms == (cl_platform_id *) NULL)
1265 (void) ThrowMagickException(exception, GetMagickModule(), ResourceLimitError,
1266 "AcquireMagickMemory failed.",
".");
1270 status = clEnv->library->clGetPlatformIDs(numPlatforms, platforms, NULL);
1271 if (status != CL_SUCCESS)
1273 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1274 "clGetPlatformIDs failed.",
"(%d)", status);
1280 clEnv->device = NULL;
1281 for (j = 0; j < 2; j++)
1284 cl_device_type deviceType;
1285 if (clEnv->deviceType == CL_DEVICE_TYPE_ALL)
1288 deviceType = CL_DEVICE_TYPE_GPU;
1290 deviceType = CL_DEVICE_TYPE_CPU;
1297 deviceType = clEnv->deviceType;
1299 for (i = 0; i < numPlatforms; i++)
1301 char version[MaxTextExtent];
1303 status = clEnv->library->clGetPlatformInfo(clEnv->platform, CL_PLATFORM_VERSION, MaxTextExtent, version, NULL);
1304 if (status != CL_SUCCESS)
1306 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1307 "clGetPlatformInfo failed.",
"(%d)", status);
1310 if (strncmp(version,
"OpenCL 1.0 ",11) == 0)
1312 status = clEnv->library->clGetDeviceIDs(platforms[i], deviceType, 1, &(clEnv->device), &numDevices);
1313 if (status != CL_SUCCESS)
1315 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1316 "clGetDeviceIDs failed.",
"(%d)", status);
1319 if (clEnv->device != NULL)
1321 clEnv->platform = platforms[i];
1328 if (platforms!=NULL)
1329 RelinquishMagickMemory(platforms);
1331 OpenCLAvailable = (clEnv->platform!=NULL
1332 && clEnv->device!=NULL)?MagickTrue:MagickFalse;
1334 #ifdef MAGICKCORE_CLPERFMARKER
1335 clEndPerfMarkerAMD();
1338 return OpenCLAvailable;
1341 static MagickBooleanType EnableOpenCLInternal(
MagickCLEnv clEnv) {
1342 if (clEnv->OpenCLInitialized != MagickFalse
1343 && clEnv->platform != NULL
1344 && clEnv->device != NULL) {
1345 clEnv->OpenCLDisabled = MagickFalse;
1348 clEnv->OpenCLDisabled = MagickTrue;
1379 static void RelinquishCommandQueues(
MagickCLEnv clEnv)
1384 LockSemaphoreInfo(clEnv->commandQueuesLock);
1385 while (clEnv->commandQueuesPos >= 0)
1386 clEnv->library->clReleaseCommandQueue(
1387 clEnv->commandQueues[clEnv->commandQueuesPos--]);
1388 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1393 MagickBooleanType status = MagickTrue;
1395 cl_context_properties cps[3];
1397 #ifdef MAGICKCORE_CLPERFMARKER
1399 int status = clInitializePerfMarkerAMD();
1400 if (status == AP_SUCCESS) {
1405 clEnv->OpenCLInitialized = MagickTrue;
1408 OpenCLLib=GetOpenCLLib();
1411 clEnv->library=OpenCLLib;
1416 MagickBooleanType flag;
1418 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
1419 ,
sizeof(MagickBooleanType), &flag, exception);
1422 if (clEnv->OpenCLDisabled != MagickFalse)
1425 clEnv->OpenCLDisabled = MagickTrue;
1427 status = InitOpenCLPlatformDevice(clEnv, exception);
1428 if (status == MagickFalse) {
1434 cps[0] = CL_CONTEXT_PLATFORM;
1435 cps[1] = (cl_context_properties)clEnv->platform;
1437 clEnv->context = clEnv->library->clCreateContext(cps, 1, &(clEnv->device), NULL, NULL, &clStatus);
1438 if (clStatus != CL_SUCCESS)
1440 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1441 "clCreateContext failed.",
"(%d)", clStatus);
1442 status = MagickFalse;
1446 RelinquishCommandQueues(clEnv);
1448 status = CompileOpenCLKernels(clEnv, exception);
1449 if (status == MagickFalse) {
1450 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1451 "clCreateCommandQueue failed.",
"(%d)", status);
1456 status = EnableOpenCLInternal(clEnv);
1465 MagickBooleanType status = MagickFalse;
1467 if ((clEnv == NULL) || (getenv(
"MAGICK_OCL_DEVICE") == (
const char *) NULL))
1470 #ifdef MAGICKCORE_CLPERFMARKER
1471 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1474 LockSemaphoreInfo(clEnv->lock);
1475 if (clEnv->OpenCLInitialized == MagickFalse) {
1476 if (clEnv->device==NULL && clEnv->OpenCLDisabled == MagickFalse)
1477 status = autoSelectDevice(clEnv, exception);
1479 status = InitOpenCLEnvInternal(clEnv, exception);
1481 UnlockSemaphoreInfo(clEnv->lock);
1483 #ifdef MAGICKCORE_CLPERFMARKER
1484 clEndPerfMarkerAMD();
1513 MagickPrivate cl_command_queue AcquireOpenCLCommandQueue(
MagickCLEnv clEnv)
1518 cl_command_queue_properties
1522 return (cl_command_queue) NULL;
1523 LockSemaphoreInfo(clEnv->commandQueuesLock);
1524 if (clEnv->commandQueuesPos >= 0) {
1525 queue=clEnv->commandQueues[clEnv->commandQueuesPos--];
1526 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1529 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1531 #if PROFILE_OCL_KERNELS
1532 properties=CL_QUEUE_PROFILING_ENABLE;
1534 queue=clEnv->library->clCreateCommandQueue(clEnv->context,clEnv->device,
1567 MagickPrivate MagickBooleanType RelinquishOpenCLCommandQueue(
MagickCLEnv clEnv,
1568 cl_command_queue queue)
1574 return(MagickFalse);
1576 LockSemaphoreInfo(clEnv->commandQueuesLock);
1578 if (clEnv->commandQueuesPos >= MAX_COMMAND_QUEUES-1)
1580 clEnv->library->clFinish(queue);
1581 status=(clEnv->library->clReleaseCommandQueue(queue) == CL_SUCCESS) ?
1582 MagickTrue : MagickFalse;
1586 clEnv->library->clFlush(queue);
1587 clEnv->commandQueues[++clEnv->commandQueuesPos]=queue;
1591 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1625 cl_kernel AcquireOpenCLKernel(
MagickCLEnv clEnv, MagickOpenCLProgram program,
const char* kernelName)
1628 cl_kernel kernel = NULL;
1629 if (clEnv != NULL && kernelName!=NULL)
1631 kernel = clEnv->library->clCreateKernel(clEnv->programs[program], kernelName, &clStatus);
1665 MagickBooleanType RelinquishOpenCLKernel(
MagickCLEnv clEnv, cl_kernel kernel)
1667 MagickBooleanType status = MagickFalse;
1668 if (clEnv != NULL && kernel != NULL)
1670 status = ((clEnv->library->clReleaseKernel(kernel) == CL_SUCCESS)?MagickTrue:MagickFalse);
1700 unsigned long GetOpenCLDeviceLocalMemorySize(
MagickCLEnv clEnv)
1702 cl_ulong localMemorySize;
1703 clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_LOCAL_MEM_SIZE,
sizeof(cl_ulong), &localMemorySize, NULL);
1704 return (
unsigned long)localMemorySize;
1708 unsigned long GetOpenCLDeviceMaxMemAllocSize(
MagickCLEnv clEnv)
1710 cl_ulong maxMemAllocSize;
1711 clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
sizeof(cl_ulong), &maxMemAllocSize, NULL);
1712 return (
unsigned long)maxMemAllocSize;
1723 ,DS_INVALID_PROFILE = 1000
1725 ,DS_INVALID_PERF_EVALUATOR_TYPE
1726 ,DS_INVALID_PERF_EVALUATOR
1727 ,DS_PERF_EVALUATOR_ERROR
1729 ,DS_UNKNOWN_DEVICE_TYPE
1730 ,DS_PROFILE_FILE_ERROR
1731 ,DS_SCORE_SERIALIZER_ERROR
1732 ,DS_SCORE_DESERIALIZER_ERROR
1737 DS_DEVICE_NATIVE_CPU = 0
1738 ,DS_DEVICE_OPENCL_DEVICE
1743 ds_device_type type;
1744 cl_device_type oclDeviceType;
1745 cl_device_id oclDeviceID;
1746 char* oclDeviceName;
1747 char* oclDriverVersion;
1748 cl_uint oclMaxClockFrequency;
1749 cl_uint oclMaxComputeUnits;
1754 unsigned int numDevices;
1756 const char* version;
1760 typedef ds_status (*ds_score_release)(
void* score);
1762 static ds_status releaseDeviceResource(ds_device* device, ds_score_release sr) {
1763 ds_status status = DS_SUCCESS;
1765 if (device->oclDeviceName) RelinquishMagickMemory(device->oclDeviceName);
1766 if (device->oclDriverVersion) RelinquishMagickMemory(device->oclDriverVersion);
1767 if (device->score) status = sr(device->score);
1772 static ds_status releaseDSProfile(ds_profile* profile, ds_score_release sr) {
1773 ds_status status = DS_SUCCESS;
1774 if (profile!=NULL) {
1775 if (profile->devices!=NULL && sr!=NULL) {
1777 for (i = 0; i < profile->numDevices; i++) {
1778 status = releaseDeviceResource(profile->devices+i,sr);
1779 if (status != DS_SUCCESS)
1782 RelinquishMagickMemory(profile->devices);
1784 RelinquishMagickMemory(profile);
1790 static ds_status initDSProfile(ds_profile** p,
const char* version) {
1792 cl_uint numPlatforms = 0;
1793 cl_platform_id* platforms = NULL;
1794 cl_device_id* devices = NULL;
1795 ds_status status = DS_SUCCESS;
1796 ds_profile* profile = NULL;
1797 unsigned int next = 0;
1801 return DS_INVALID_PROFILE;
1803 profile = (ds_profile*) AcquireMagickMemory(
sizeof(ds_profile));
1804 if (profile == NULL)
1805 return DS_MEMORY_ERROR;
1807 memset(profile, 0,
sizeof(ds_profile));
1809 OpenCLLib->clGetPlatformIDs(0, NULL, &numPlatforms);
1810 if (numPlatforms > 0) {
1811 platforms = (cl_platform_id*) AcquireQuantumMemory(numPlatforms,
sizeof(cl_platform_id));
1812 if (platforms == NULL) {
1813 status = DS_MEMORY_ERROR;
1816 OpenCLLib->clGetPlatformIDs(numPlatforms, platforms, NULL);
1817 for (i = 0; i < (
unsigned int)numPlatforms; i++) {
1819 if (OpenCLLib->clGetDeviceIDs(platforms[i], CL_DEVICE_TYPE_CPU | CL_DEVICE_TYPE_GPU, 0, NULL, &num) == CL_SUCCESS)
1824 profile->numDevices = numDevices+1;
1826 profile->devices = (ds_device*) AcquireQuantumMemory(profile->numDevices,
sizeof(ds_device));
1827 if (profile->devices == NULL) {
1828 profile->numDevices = 0;
1829 status = DS_MEMORY_ERROR;
1832 memset(profile->devices, 0, profile->numDevices*
sizeof(ds_device));
1834 if (numDevices > 0) {
1835 devices = (cl_device_id*) AcquireQuantumMemory(numDevices,
sizeof(cl_device_id));
1836 if (devices == NULL) {
1837 status = DS_MEMORY_ERROR;
1840 for (i = 0; i < (
unsigned int)numPlatforms; i++) {
1844 for (d = 0; d < 2; d++) {
1846 cl_device_type deviceType;
1849 deviceType = CL_DEVICE_TYPE_GPU;
1852 deviceType = CL_DEVICE_TYPE_CPU;
1858 if (OpenCLLib->clGetDeviceIDs(platforms[i], deviceType, numDevices, devices, &num) != CL_SUCCESS)
1860 for (j = 0; j < num; j++, next++) {
1863 profile->devices[next].type = DS_DEVICE_OPENCL_DEVICE;
1864 profile->devices[next].oclDeviceID = devices[j];
1866 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_NAME
1867 , 0, NULL, &length);
1868 profile->devices[next].oclDeviceName = (
char*) AcquireQuantumMemory(length,
sizeof(
char));
1869 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_NAME
1870 , length, profile->devices[next].oclDeviceName, NULL);
1872 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DRIVER_VERSION
1873 , 0, NULL, &length);
1874 profile->devices[next].oclDriverVersion = (
char*) AcquireQuantumMemory(length,
sizeof(
char));
1875 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DRIVER_VERSION
1876 , length, profile->devices[next].oclDriverVersion, NULL);
1878 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_MAX_CLOCK_FREQUENCY
1879 ,
sizeof(cl_uint), &profile->devices[next].oclMaxClockFrequency, NULL);
1881 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_MAX_COMPUTE_UNITS
1882 ,
sizeof(cl_uint), &profile->devices[next].oclMaxComputeUnits, NULL);
1884 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_TYPE
1885 ,
sizeof(cl_device_type), &profile->devices[next].oclDeviceType, NULL);
1891 profile->devices[next].type = DS_DEVICE_NATIVE_CPU;
1892 profile->version = version;
1895 if (platforms) RelinquishMagickMemory(platforms);
1896 if (devices) RelinquishMagickMemory(devices);
1897 if (status == DS_SUCCESS) {
1902 if (profile->devices)
1903 RelinquishMagickMemory(profile->devices);
1904 RelinquishMagickMemory(profile);
1914 typedef ds_status (*ds_perf_evaluator)(ds_device* device,
void* data);
1918 ,DS_EVALUATE_NEW_ONLY
1919 } ds_evaluation_type;
1921 static ds_status profileDevices(ds_profile* profile,
const ds_evaluation_type type
1922 ,ds_perf_evaluator evaluator,
void* evaluatorData,
unsigned int* numUpdates) {
1923 ds_status status = DS_SUCCESS;
1925 unsigned int updates = 0;
1927 if (profile == NULL) {
1928 return DS_INVALID_PROFILE;
1930 if (evaluator == NULL) {
1931 return DS_INVALID_PERF_EVALUATOR;
1934 for (i = 0; i < profile->numDevices; i++) {
1935 ds_status evaluatorStatus;
1938 case DS_EVALUATE_NEW_ONLY:
1939 if (profile->devices[i].score != NULL)
1942 case DS_EVALUATE_ALL:
1943 evaluatorStatus = evaluator(profile->devices+i,evaluatorData);
1944 if (evaluatorStatus != DS_SUCCESS) {
1945 status = evaluatorStatus;
1951 return DS_INVALID_PERF_EVALUATOR_TYPE;
1956 *numUpdates = updates;
1961 #define DS_TAG_VERSION "<version>"
1962 #define DS_TAG_VERSION_END "</version>"
1963 #define DS_TAG_DEVICE "<device>"
1964 #define DS_TAG_DEVICE_END "</device>"
1965 #define DS_TAG_SCORE "<score>"
1966 #define DS_TAG_SCORE_END "</score>"
1967 #define DS_TAG_DEVICE_TYPE "<type>"
1968 #define DS_TAG_DEVICE_TYPE_END "</type>"
1969 #define DS_TAG_DEVICE_NAME "<name>"
1970 #define DS_TAG_DEVICE_NAME_END "</name>"
1971 #define DS_TAG_DEVICE_DRIVER_VERSION "<driver>"
1972 #define DS_TAG_DEVICE_DRIVER_VERSION_END "</driver>"
1973 #define DS_TAG_DEVICE_MAX_COMPUTE_UNITS "<max cu>"
1974 #define DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END "</max cu>"
1975 #define DS_TAG_DEVICE_MAX_CLOCK_FREQ "<max clock>"
1976 #define DS_TAG_DEVICE_MAX_CLOCK_FREQ_END "</max clock>"
1978 #define DS_DEVICE_NATIVE_CPU_STRING "native_cpu"
1982 typedef ds_status (*ds_score_serializer)(ds_device* device,
void** serializedScore,
unsigned int* serializedScoreSize);
1983 static ds_status writeProfileToFile(ds_profile* profile, ds_score_serializer serializer,
const char* file) {
1984 ds_status status = DS_SUCCESS;
1985 FILE* profileFile = NULL;
1988 if (profile == NULL)
1989 return DS_INVALID_PROFILE;
1991 profileFile = fopen(file,
"wb");
1992 if (profileFile==NULL) {
1993 status = DS_FILE_ERROR;
1999 fwrite(DS_TAG_VERSION,
sizeof(
char), strlen(DS_TAG_VERSION), profileFile);
2000 fwrite(profile->version,
sizeof(
char), strlen(profile->version), profileFile);
2001 fwrite(DS_TAG_VERSION_END,
sizeof(
char), strlen(DS_TAG_VERSION_END), profileFile);
2002 fwrite(
"\n",
sizeof(
char), 1, profileFile);
2004 for (i = 0; i < profile->numDevices && status == DS_SUCCESS; i++) {
2005 void* serializedScore;
2006 unsigned int serializedScoreSize;
2008 fwrite(DS_TAG_DEVICE,
sizeof(
char), strlen(DS_TAG_DEVICE), profileFile);
2010 fwrite(DS_TAG_DEVICE_TYPE,
sizeof(
char), strlen(DS_TAG_DEVICE_TYPE), profileFile);
2011 fwrite(&profile->devices[i].type,
sizeof(ds_device_type),1, profileFile);
2012 fwrite(DS_TAG_DEVICE_TYPE_END,
sizeof(
char), strlen(DS_TAG_DEVICE_TYPE_END), profileFile);
2014 switch(profile->devices[i].type) {
2015 case DS_DEVICE_NATIVE_CPU:
2025 case DS_DEVICE_OPENCL_DEVICE:
2029 fwrite(DS_TAG_DEVICE_NAME,
sizeof(
char), strlen(DS_TAG_DEVICE_NAME), profileFile);
2030 fwrite(profile->devices[i].oclDeviceName,
sizeof(
char),strlen(profile->devices[i].oclDeviceName), profileFile);
2031 fwrite(DS_TAG_DEVICE_NAME_END,
sizeof(
char), strlen(DS_TAG_DEVICE_NAME_END), profileFile);
2033 fwrite(DS_TAG_DEVICE_DRIVER_VERSION,
sizeof(
char), strlen(DS_TAG_DEVICE_DRIVER_VERSION), profileFile);
2034 fwrite(profile->devices[i].oclDriverVersion,
sizeof(
char),strlen(profile->devices[i].oclDriverVersion), profileFile);
2035 fwrite(DS_TAG_DEVICE_DRIVER_VERSION_END,
sizeof(
char), strlen(DS_TAG_DEVICE_DRIVER_VERSION_END), profileFile);
2037 fwrite(DS_TAG_DEVICE_MAX_COMPUTE_UNITS,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_COMPUTE_UNITS), profileFile);
2038 sprintf(tmp,
"%d",profile->devices[i].oclMaxComputeUnits);
2039 fwrite(tmp,
sizeof(
char),strlen(tmp), profileFile);
2040 fwrite(DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END), profileFile);
2042 fwrite(DS_TAG_DEVICE_MAX_CLOCK_FREQ,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_CLOCK_FREQ), profileFile);
2043 sprintf(tmp,
"%d",profile->devices[i].oclMaxClockFrequency);
2044 fwrite(tmp,
sizeof(
char),strlen(tmp), profileFile);
2045 fwrite(DS_TAG_DEVICE_MAX_CLOCK_FREQ_END,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_CLOCK_FREQ_END), profileFile);
2049 status = DS_UNKNOWN_DEVICE_TYPE;
2053 fwrite(DS_TAG_SCORE,
sizeof(
char), strlen(DS_TAG_SCORE), profileFile);
2054 status = serializer(profile->devices+i, &serializedScore, &serializedScoreSize);
2055 if (status == DS_SUCCESS && serializedScore!=NULL && serializedScoreSize > 0) {
2056 fwrite(serializedScore,
sizeof(
char), serializedScoreSize, profileFile);
2057 RelinquishMagickMemory(serializedScore);
2059 fwrite(DS_TAG_SCORE_END,
sizeof(
char), strlen(DS_TAG_SCORE_END), profileFile);
2060 fwrite(DS_TAG_DEVICE_END,
sizeof(
char), strlen(DS_TAG_DEVICE_END), profileFile);
2061 fwrite(
"\n",
sizeof(
char),1,profileFile);
2063 fclose(profileFile);
2069 static ds_status readProFile(
const char* fileName,
char** content,
size_t* contentSize) {
2070 ds_status status = DS_SUCCESS;
2071 FILE * input = NULL;
2074 char* binary = NULL;
2079 input = fopen(fileName,
"rb");
2081 return DS_FILE_ERROR;
2084 fseek(input, 0L, SEEK_END);
2085 size = ftell(input);
2087 binary = (
char*) AcquireQuantumMemory(1,size);
2088 if(binary == NULL) {
2089 status = DS_FILE_ERROR;
2092 rsize = fread(binary,
sizeof(
char), size, input);
2095 status = DS_FILE_ERROR;
2098 *contentSize = size;
2102 if (input != NULL) fclose(input);
2103 if (status != DS_SUCCESS
2104 && binary != NULL) {
2105 RelinquishMagickMemory(binary);
2113 static const char* findString(
const char* contentStart,
const char* contentEnd,
const char*
string) {
2114 size_t stringLength;
2115 const char* currentPosition;
2118 stringLength = strlen(
string);
2119 currentPosition = contentStart;
2120 for(currentPosition = contentStart; currentPosition < contentEnd; currentPosition++) {
2121 if (*currentPosition ==
string[0]) {
2122 if (currentPosition+stringLength < contentEnd) {
2123 if (strncmp(currentPosition,
string, stringLength) == 0) {
2124 found = currentPosition;
2134 typedef ds_status (*ds_score_deserializer)(ds_device* device,
const unsigned char* serializedScore,
unsigned int serializedScoreSize);
2135 static ds_status readProfileFromFile(ds_profile* profile, ds_score_deserializer deserializer,
const char* file) {
2137 ds_status status = DS_SUCCESS;
2138 char* contentStart = NULL;
2139 const char* contentEnd = NULL;
2143 return DS_INVALID_PROFILE;
2145 status = readProFile(file, &contentStart, &contentSize);
2146 if (status == DS_SUCCESS) {
2147 const char* currentPosition;
2148 const char* dataStart;
2149 const char* dataEnd;
2150 size_t versionStringLength;
2152 contentEnd = contentStart + contentSize;
2153 currentPosition = contentStart;
2157 dataStart = findString(currentPosition, contentEnd, DS_TAG_VERSION);
2158 if (dataStart == NULL) {
2159 status = DS_PROFILE_FILE_ERROR;
2162 dataStart += strlen(DS_TAG_VERSION);
2164 dataEnd = findString(dataStart, contentEnd, DS_TAG_VERSION_END);
2165 if (dataEnd==NULL) {
2166 status = DS_PROFILE_FILE_ERROR;
2170 versionStringLength = strlen(profile->version);
2171 if (versionStringLength!=(
size_t)(dataEnd-dataStart)
2172 || strncmp(profile->version, dataStart, versionStringLength)!=(
int)0) {
2174 status = DS_PROFILE_FILE_ERROR;
2177 currentPosition = dataEnd+strlen(DS_TAG_VERSION_END);
2180 DisableMSCWarning(4127)
2185 const char* deviceTypeStart;
2186 const char* deviceTypeEnd;
2187 ds_device_type deviceType;
2189 const char* deviceNameStart;
2190 const char* deviceNameEnd;
2192 const char* deviceScoreStart;
2193 const char* deviceScoreEnd;
2195 const char* deviceDriverStart;
2196 const char* deviceDriverEnd;
2198 const char* tmpStart;
2202 cl_uint maxClockFrequency;
2203 cl_uint maxComputeUnits;
2205 dataStart = findString(currentPosition, contentEnd, DS_TAG_DEVICE);
2206 if (dataStart==NULL) {
2210 dataStart+=strlen(DS_TAG_DEVICE);
2211 dataEnd = findString(dataStart, contentEnd, DS_TAG_DEVICE_END);
2212 if (dataEnd==NULL) {
2213 status = DS_PROFILE_FILE_ERROR;
2218 deviceTypeStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_TYPE);
2219 if (deviceTypeStart==NULL) {
2220 status = DS_PROFILE_FILE_ERROR;
2223 deviceTypeStart+=strlen(DS_TAG_DEVICE_TYPE);
2224 deviceTypeEnd = findString(deviceTypeStart, contentEnd, DS_TAG_DEVICE_TYPE_END);
2225 if (deviceTypeEnd==NULL) {
2226 status = DS_PROFILE_FILE_ERROR;
2229 memcpy(&deviceType, deviceTypeStart,
sizeof(ds_device_type));
2233 if (deviceType == DS_DEVICE_OPENCL_DEVICE) {
2235 deviceNameStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_NAME);
2236 if (deviceNameStart==NULL) {
2237 status = DS_PROFILE_FILE_ERROR;
2240 deviceNameStart+=strlen(DS_TAG_DEVICE_NAME);
2241 deviceNameEnd = findString(deviceNameStart, contentEnd, DS_TAG_DEVICE_NAME_END);
2242 if (deviceNameEnd==NULL) {
2243 status = DS_PROFILE_FILE_ERROR;
2248 deviceDriverStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_DRIVER_VERSION);
2249 if (deviceDriverStart==NULL) {
2250 status = DS_PROFILE_FILE_ERROR;
2253 deviceDriverStart+=strlen(DS_TAG_DEVICE_DRIVER_VERSION);
2254 deviceDriverEnd = findString(deviceDriverStart, contentEnd, DS_TAG_DEVICE_DRIVER_VERSION_END);
2255 if (deviceDriverEnd ==NULL) {
2256 status = DS_PROFILE_FILE_ERROR;
2261 tmpStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_MAX_COMPUTE_UNITS);
2262 if (tmpStart==NULL) {
2263 status = DS_PROFILE_FILE_ERROR;
2266 tmpStart+=strlen(DS_TAG_DEVICE_MAX_COMPUTE_UNITS);
2267 tmpEnd = findString(tmpStart, contentEnd, DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END);
2268 if (tmpEnd ==NULL) {
2269 status = DS_PROFILE_FILE_ERROR;
2272 memcpy(tmp,tmpStart,tmpEnd-tmpStart);
2273 tmp[tmpEnd-tmpStart] =
'\0';
2274 maxComputeUnits = strtol(tmp,(
char **) NULL,10);
2277 tmpStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_MAX_CLOCK_FREQ);
2278 if (tmpStart==NULL) {
2279 status = DS_PROFILE_FILE_ERROR;
2282 tmpStart+=strlen(DS_TAG_DEVICE_MAX_CLOCK_FREQ);
2283 tmpEnd = findString(tmpStart, contentEnd, DS_TAG_DEVICE_MAX_CLOCK_FREQ_END);
2284 if (tmpEnd ==NULL) {
2285 status = DS_PROFILE_FILE_ERROR;
2288 memcpy(tmp,tmpStart,tmpEnd-tmpStart);
2289 tmp[tmpEnd-tmpStart] =
'\0';
2290 maxClockFrequency = strtol(tmp,(
char **) NULL,10);
2294 for (i = 0; i < profile->numDevices; i++) {
2295 if (profile->devices[i].type == DS_DEVICE_OPENCL_DEVICE) {
2296 size_t actualDeviceNameLength;
2297 size_t driverVersionLength;
2299 actualDeviceNameLength = strlen(profile->devices[i].oclDeviceName);
2300 driverVersionLength = strlen(profile->devices[i].oclDriverVersion);
2301 if (actualDeviceNameLength == (
size_t)(deviceNameEnd - deviceNameStart)
2302 && driverVersionLength == (
size_t)(deviceDriverEnd - deviceDriverStart)
2303 && maxComputeUnits == profile->devices[i].oclMaxComputeUnits
2304 && maxClockFrequency == profile->devices[i].oclMaxClockFrequency
2305 && strncmp(profile->devices[i].oclDeviceName, deviceNameStart, actualDeviceNameLength)==(
int)0
2306 && strncmp(profile->devices[i].oclDriverVersion, deviceDriverStart, driverVersionLength)==(
int)0) {
2308 deviceScoreStart = findString(dataStart, contentEnd, DS_TAG_SCORE);
2309 if (deviceNameStart==NULL) {
2310 status = DS_PROFILE_FILE_ERROR;
2313 deviceScoreStart+=strlen(DS_TAG_SCORE);
2314 deviceScoreEnd = findString(deviceScoreStart, contentEnd, DS_TAG_SCORE_END);
2315 status = deserializer(profile->devices+i, (
const unsigned char*)deviceScoreStart, deviceScoreEnd-deviceScoreStart);
2316 if (status != DS_SUCCESS) {
2324 else if (deviceType == DS_DEVICE_NATIVE_CPU) {
2325 for (i = 0; i < profile->numDevices; i++) {
2326 if (profile->devices[i].type == DS_DEVICE_NATIVE_CPU) {
2327 deviceScoreStart = findString(dataStart, contentEnd, DS_TAG_SCORE);
2328 if (deviceScoreStart==NULL) {
2329 status = DS_PROFILE_FILE_ERROR;
2332 deviceScoreStart+=strlen(DS_TAG_SCORE);
2333 deviceScoreEnd = findString(deviceScoreStart, contentEnd, DS_TAG_SCORE_END);
2334 status = deserializer(profile->devices+i, (
const unsigned char*)deviceScoreStart, deviceScoreEnd-deviceScoreStart);
2335 if (status != DS_SUCCESS) {
2343 currentPosition = dataEnd+strlen(DS_TAG_DEVICE_END);
2347 if (contentStart!=NULL) RelinquishMagickMemory(contentStart);
2353 static ds_status getNumDeviceWithEmptyScore(ds_profile* profile,
unsigned int* num) {
2355 if (profile == NULL || num==NULL)
2356 return DS_MEMORY_ERROR;
2358 for (i = 0; i < profile->numDevices; i++) {
2359 if (profile->devices[i].score == NULL) {
2372 typedef double AccelerateScoreType;
2374 static ds_status AcceleratePerfEvaluator(ds_device *device,
2375 void *magick_unused(data))
2377 #define ACCELERATE_PERF_DIMEN "2048x1536"
2379 #define ReturnStatus(status) \
2381 if (oldClEnv != (MagickCLEnv) NULL) \
2382 defaultCLEnv=oldClEnv; \
2383 if (clEnv != (MagickCLEnv) NULL) \
2384 (void) RelinquishMagickOpenCLEnv(clEnv); \
2401 magick_unreferenced(data);
2404 ReturnStatus(DS_PERF_EVALUATOR_ERROR);
2406 clEnv=AcquireMagickOpenCLEnv();
2407 exception=AcquireExceptionInfo();
2409 if (device->type == DS_DEVICE_NATIVE_CPU)
2412 MagickBooleanType flag=MagickTrue;
2413 SetMagickOpenCLEnvParamInternal(clEnv,
2414 MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED,
sizeof(MagickBooleanType),
2417 else if (device->type == DS_DEVICE_OPENCL_DEVICE)
2420 SetMagickOpenCLEnvParamInternal(clEnv,MAGICK_OPENCL_ENV_PARAM_DEVICE,
2421 sizeof(cl_device_id),&device->oclDeviceID,exception);
2424 ReturnStatus(DS_PERF_EVALUATOR_ERROR);
2427 clEnv->disableProgramCache = defaultCLEnv->disableProgramCache;
2429 status=InitOpenCLEnvInternal(clEnv,exception);
2430 oldClEnv=defaultCLEnv;
2434 if (status != MagickFalse)
2445 imageInfo=AcquireImageInfo();
2446 CloneString(&imageInfo->size,ACCELERATE_PERF_DIMEN);
2447 CopyMagickString(imageInfo->filename,
"xc:none",MaxTextExtent);
2448 inputImage=ReadImage(imageInfo,exception);
2450 initAccelerateTimer(&timer);
2452 for (i=0; i<=NUM_ITER; i++)
2466 startAccelerateTimer(&timer);
2468 #ifdef MAGICKCORE_CLPERFMARKER
2469 clBeginPerfMarkerAMD(
"PerfEvaluatorRegion",
"");
2472 bluredImage=BlurImage(inputImage,10.0f,3.5f,exception);
2473 unsharpedImage=UnsharpMaskImage(bluredImage,2.0f,2.0f,50.0f,10.0f,
2475 resizedImage=ResizeImage(unsharpedImage,640,480,LanczosFilter,1.0,
2482 if (device->type != DS_DEVICE_NATIVE_CPU)
2484 events=GetOpenCLEvents(resizedImage,&event_count);
2485 if (event_count > 0)
2486 clEnv->library->clWaitForEvents(event_count,events);
2487 events=(cl_event *) RelinquishMagickMemory(events);
2490 #ifdef MAGICKCORE_CLPERFMARKER
2491 clEndPerfMarkerAMD();
2495 stopAccelerateTimer(&timer);
2498 DestroyImage(bluredImage);
2500 DestroyImage(unsharpedImage);
2502 DestroyImage(resizedImage);
2504 DestroyImage(inputImage);
2508 if (device->score == NULL)
2509 device->score= AcquireMagickMemory(
sizeof(AccelerateScoreType));
2511 if (status != MagickFalse)
2512 *(AccelerateScoreType*) device->score=readAccelerateTimer(&timer);
2514 *(AccelerateScoreType*) device->score=42;
2516 ReturnStatus(DS_SUCCESS);
2519 ds_status AccelerateScoreSerializer(ds_device* device,
void** serializedScore,
unsigned int* serializedScoreSize) {
2523 char* s = (
char*) AcquireQuantumMemory(256,
sizeof(
char));
2524 sprintf(s,
"%.4f",*((AccelerateScoreType*)device->score));
2525 *serializedScore = (
void*)s;
2526 *serializedScoreSize = (
unsigned int) strlen(s);
2530 return DS_SCORE_SERIALIZER_ERROR;
2534 ds_status AccelerateScoreDeserializer(ds_device* device,
const unsigned char* serializedScore,
unsigned int serializedScoreSize) {
2537 char* s = (
char*) AcquireQuantumMemory(1,serializedScoreSize+1);
2538 memcpy(s, serializedScore, serializedScoreSize);
2539 s[serializedScoreSize] = (char)
'\0';
2540 device->score = AcquireMagickMemory(
sizeof(AccelerateScoreType));
2541 *((AccelerateScoreType*)device->score) = (AccelerateScoreType)
2542 strtod(s, (
char **) NULL);
2543 RelinquishMagickMemory(s);
2547 return DS_SCORE_DESERIALIZER_ERROR;
2551 ds_status AccelerateScoreRelease(
void* score) {
2553 RelinquishMagickMemory(score);
2558 ds_status canWriteProfileToFile(
const char *path)
2560 FILE* profileFile = fopen(path,
"ab");
2562 if (profileFile==NULL)
2563 return DS_FILE_ERROR;
2565 fclose(profileFile);
2570 #define IMAGEMAGICK_PROFILE_VERSION "ImageMagick Device Selection v0.9"
2571 #define IMAGEMAGICK_PROFILE_FILE "ImagemagickOpenCLDeviceProfile"
2574 MagickBooleanType mStatus = MagickFalse;
2576 ds_profile* profile;
2577 unsigned int numDeviceProfiled = 0;
2579 unsigned int bestDeviceIndex;
2580 AccelerateScoreType bestScore;
2581 char path[MaxTextExtent];
2582 MagickBooleanType flag;
2583 ds_evaluation_type profileType;
2585 LockDefaultOpenCLEnv();
2589 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2590 ,
sizeof(MagickBooleanType), &flag, exception);
2593 OpenCLLib=GetOpenCLLib();
2594 if (OpenCLLib==NULL)
2596 mStatus=InitOpenCLEnvInternal(clEnv, exception);
2600 clEnv->library=OpenCLLib;
2602 status = initDSProfile(&profile, IMAGEMAGICK_PROFILE_VERSION);
2603 if (status!=DS_SUCCESS) {
2604 (void) ThrowMagickException(exception, GetMagickModule(), ModuleFatalError,
"Error when initializing the profile",
"'%s'",
".");
2608 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
2609 ,GetOpenCLCachedFilesDirectory()
2610 ,DirectorySeparator,IMAGEMAGICK_PROFILE_FILE);
2612 if (canWriteProfileToFile(path) != DS_SUCCESS) {
2616 bestDeviceIndex = 0;
2617 for (i = 1; i < profile->numDevices; i++) {
2618 if ((profile->devices[i].type == DS_DEVICE_OPENCL_DEVICE) && (profile->devices[i].oclDeviceType == CL_DEVICE_TYPE_GPU)) {
2619 bestDeviceIndex = i;
2625 if (clEnv->regenerateProfile != MagickFalse) {
2626 profileType = DS_EVALUATE_ALL;
2629 readProfileFromFile(profile, AccelerateScoreDeserializer, path);
2630 profileType = DS_EVALUATE_NEW_ONLY;
2632 status = profileDevices(profile, profileType, AcceleratePerfEvaluator, NULL, &numDeviceProfiled);
2634 if (status!=DS_SUCCESS) {
2635 (void) ThrowMagickException(exception, GetMagickModule(), ModuleFatalError,
"Error when initializing the profile",
"'%s'",
".");
2638 if (numDeviceProfiled > 0) {
2639 status = writeProfileToFile(profile, AccelerateScoreSerializer, path);
2640 if (status!=DS_SUCCESS) {
2641 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"Error when saving the profile into a file",
"'%s'",
".");
2646 bestDeviceIndex = 0;
2647 bestScore = *(AccelerateScoreType*)profile->devices[bestDeviceIndex].score;
2648 for (i = 1; i < profile->numDevices; i++) {
2649 AccelerateScoreType score = *(AccelerateScoreType*)profile->devices[i].score;
2650 if (score < bestScore) {
2651 bestDeviceIndex = i;
2658 if (profile->devices[bestDeviceIndex].type == DS_DEVICE_NATIVE_CPU) {
2661 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2662 ,
sizeof(MagickBooleanType), &flag, exception);
2664 else if (profile->devices[bestDeviceIndex].type == DS_DEVICE_OPENCL_DEVICE) {
2667 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2668 ,
sizeof(MagickBooleanType), &flag, exception);
2669 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2670 ,
sizeof(cl_device_id), &profile->devices[bestDeviceIndex].oclDeviceID,exception);
2673 status = DS_PERF_EVALUATOR_ERROR;
2676 mStatus=InitOpenCLEnvInternal(clEnv, exception);
2678 status = releaseDSProfile(profile, AccelerateScoreRelease);
2679 if (status!=DS_SUCCESS) {
2680 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"Error when releasing the profile",
"'%s'",
".");
2685 UnlockDefaultOpenCLEnv();
2723 MagickExport MagickBooleanType InitImageMagickOpenCL(
2724 ImageMagickOpenCLMode mode,
void *userSelectedDevice,
void *selectedDevice,
2727 MagickBooleanType status = MagickFalse;
2729 MagickBooleanType flag;
2731 clEnv = GetDefaultOpenCLEnv();
2735 case MAGICK_OPENCL_OFF:
2737 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2738 ,
sizeof(MagickBooleanType), &flag, exception);
2739 status = InitOpenCLEnv(clEnv, exception);
2742 *(cl_device_id*)selectedDevice = NULL;
2745 case MAGICK_OPENCL_DEVICE_SELECT_USER:
2747 if (userSelectedDevice == NULL)
2751 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2752 ,
sizeof(MagickBooleanType), &flag, exception);
2754 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2755 ,
sizeof(cl_device_id), userSelectedDevice,exception);
2757 status = InitOpenCLEnv(clEnv, exception);
2758 if (selectedDevice) {
2759 GetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2760 ,
sizeof(cl_device_id), selectedDevice, exception);
2764 case MAGICK_OPENCL_DEVICE_SELECT_AUTO_CLEAR_CACHE:
2766 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED
2767 ,
sizeof(MagickBooleanType), &flag, exception);
2769 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE
2770 ,
sizeof(MagickBooleanType), &flag, exception);
2773 case MAGICK_OPENCL_DEVICE_SELECT_AUTO:
2776 cl_device_id d = NULL;
2778 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2779 ,
sizeof(MagickBooleanType), &flag, exception);
2780 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2781 ,
sizeof(cl_device_id), &d,exception);
2782 status = InitOpenCLEnv(clEnv, exception);
2783 if (selectedDevice) {
2784 GetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2785 ,
sizeof(cl_device_id), selectedDevice, exception);
2796 MagickBooleanType OpenCLThrowMagickException(
ExceptionInfo *exception,
2797 const char *module,
const char *
function,
const size_t line,
2798 const ExceptionType severity,
const char *tag,
const char *format,...) {
2804 status = MagickTrue;
2806 clEnv = GetDefaultOpenCLEnv();
2809 assert(exception->signature == MagickCoreSignature);
2812 cl_device_type dType;
2813 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_TYPE ,
sizeof(cl_device_type),&dType,NULL);
2814 if (dType == CL_DEVICE_TYPE_CPU) {
2815 char buffer[MaxTextExtent];
2816 clEnv->library->clGetPlatformInfo(clEnv->platform, CL_PLATFORM_NAME, MaxTextExtent, buffer, NULL);
2820 if (strncmp(buffer,
"Intel",5) == 0) {
2822 InitImageMagickOpenCL(MAGICK_OPENCL_OFF, NULL, NULL, exception);
2827 #ifdef OPENCLLOG_ENABLED
2831 va_start(operands,format);
2832 status=ThrowMagickExceptionList(exception,module,
function,line,severity,tag, format,operands);
2836 magick_unreferenced(module);
2837 magick_unreferenced(
function);
2838 magick_unreferenced(line);
2839 magick_unreferenced(tag);
2840 magick_unreferenced(format);
2846 char* openclCachedFilesDirectory;
2850 const char* GetOpenCLCachedFilesDirectory() {
2851 if (openclCachedFilesDirectory == NULL) {
2852 if (openclCachedFilesDirectoryLock == NULL)
2854 ActivateSemaphoreInfo(&openclCachedFilesDirectoryLock);
2856 LockSemaphoreInfo(openclCachedFilesDirectoryLock);
2857 if (openclCachedFilesDirectory == NULL) {
2858 char path[MaxTextExtent];
2861 struct stat attributes;
2862 MagickBooleanType status;
2863 int mkdirStatus = 0;
2867 home=GetEnvironmentValue(
"MAGICK_OPENCL_CACHE_DIR");
2868 if (home == (
char *) NULL)
2870 home=GetEnvironmentValue(
"XDG_CACHE_HOME");
2871 #if defined(MAGICKCORE_WINDOWS_SUPPORT) || defined(__MINGW32__)
2872 if (home == (
char *) NULL)
2873 home=GetEnvironmentValue(
"LOCALAPPDATA");
2874 if (home == (
char *) NULL)
2875 home=GetEnvironmentValue(
"APPDATA");
2876 if (home == (
char *) NULL)
2877 home=GetEnvironmentValue(
"USERPROFILE");
2881 if (home != (
char *) NULL)
2884 (void) FormatLocaleString(path,MaxTextExtent,
"%s",home);
2885 status=GetPathAttributes(path,&attributes);
2886 if (status == MagickFalse)
2889 #ifdef MAGICKCORE_WINDOWS_SUPPORT
2890 mkdirStatus = mkdir(path);
2892 mkdirStatus = mkdir(path, 0777);
2899 (void) FormatLocaleString(path,MaxTextExtent,
2900 "%s%sImageMagick",home,DirectorySeparator);
2902 status=GetPathAttributes(path,&attributes);
2903 if (status == MagickFalse)
2905 #ifdef MAGICKCORE_WINDOWS_SUPPORT
2906 mkdirStatus = mkdir(path);
2908 mkdirStatus = mkdir(path, 0777);
2915 temp = (
char*)AcquireCriticalMemory(strlen(path)+1);
2916 CopyMagickString(temp,path,strlen(path)+1);
2918 home=DestroyString(home);
2920 home=GetEnvironmentValue(
"HOME");
2921 if (home != (
char *) NULL)
2927 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s.cache",
2928 home,DirectorySeparator);
2929 status=GetPathAttributes(path,&attributes);
2930 if (status == MagickFalse)
2933 #ifdef MAGICKCORE_WINDOWS_SUPPORT
2934 mkdirStatus = mkdir(path);
2936 mkdirStatus = mkdir(path, 0777);
2943 (void) FormatLocaleString(path,MaxTextExtent,
2944 "%s%s.cache%sImageMagick",home,DirectorySeparator,
2945 DirectorySeparator);
2947 status=GetPathAttributes(path,&attributes);
2948 if (status == MagickFalse)
2950 #ifdef MAGICKCORE_WINDOWS_SUPPORT
2951 mkdirStatus = mkdir(path);
2953 mkdirStatus = mkdir(path, 0777);
2960 temp = (
char*)AcquireCriticalMemory(strlen(path)+1);
2961 CopyMagickString(temp,path,strlen(path)+1);
2963 home=DestroyString(home);
2966 openclCachedFilesDirectory = temp;
2968 UnlockSemaphoreInfo(openclCachedFilesDirectoryLock);
2970 return openclCachedFilesDirectory;
2975 void OpenCLLog(
const char* message) {
2977 #ifdef OPENCLLOG_ENABLED
2978 #define OPENCL_LOG_FILE "ImageMagickOpenCL.log"
2981 if (getenv(
"MAGICK_OCL_LOG"))
2984 char path[MaxTextExtent];
2985 unsigned long allocSize;
2989 clEnv = GetDefaultOpenCLEnv();
2992 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
2993 ,GetOpenCLCachedFilesDirectory()
2994 ,DirectorySeparator,OPENCL_LOG_FILE);
2997 log = fopen(path,
"ab");
2998 if (log == (FILE *) NULL)
3000 fwrite(message,
sizeof(
char), strlen(message), log);
3001 fwrite(
"\n",
sizeof(
char), 1, log);
3003 if (clEnv->OpenCLInitialized && !clEnv->OpenCLDisabled)
3005 allocSize = GetOpenCLDeviceMaxMemAllocSize(clEnv);
3006 fprintf(log,
"Devic Max Memory Alloc Size: %lu\n", allocSize);
3013 magick_unreferenced(message);
3017 MagickPrivate
void OpenCLTerminus()
3020 if (openclCachedFilesDirectory != (
char *) NULL)
3021 openclCachedFilesDirectory=DestroyString(openclCachedFilesDirectory);
3023 DestroySemaphoreInfo(&openclCachedFilesDirectoryLock);
3026 (void) RelinquishMagickOpenCLEnv(defaultCLEnv);
3030 DestroySemaphoreInfo(&defaultCLEnvLock);
3031 if (OpenCLLib != (MagickLibrary *)NULL)
3033 if (OpenCLLib->base != (
void *) NULL)
3034 (
void) lt_dlclose(OpenCLLib->base);
3035 OpenCLLib=(MagickLibrary *)RelinquishMagickMemory(OpenCLLib);
3038 DestroySemaphoreInfo(&OpenCLLibLock);
3044 MagickBooleanType OpenCLInitialized;
3055 MagickExport MagickBooleanType SetMagickOpenCLEnvParam(
3056 MagickCLEnv magick_unused(clEnv),MagickOpenCLEnvParam magick_unused(param),
3057 size_t magick_unused(dataSize),
void *magick_unused(data),
3060 magick_unreferenced(clEnv);
3061 magick_unreferenced(param);
3062 magick_unreferenced(dataSize);
3063 magick_unreferenced(data);
3064 magick_unreferenced(exception);
3065 return(MagickFalse);
3068 MagickExport MagickBooleanType GetMagickOpenCLEnvParam(
3069 MagickCLEnv magick_unused(clEnv),MagickOpenCLEnvParam magick_unused(param),
3070 size_t magick_unused(dataSize),
void *magick_unused(data),
3073 magick_unreferenced(clEnv);
3074 magick_unreferenced(param);
3075 magick_unreferenced(dataSize);
3076 magick_unreferenced(data);
3077 magick_unreferenced(exception);
3078 return(MagickFalse);
3081 MagickExport MagickBooleanType InitOpenCLEnv(
MagickCLEnv magick_unused(clEnv),
3084 magick_unreferenced(clEnv);
3085 magick_unreferenced(exception);
3086 return(MagickFalse);
3089 MagickExport MagickBooleanType InitImageMagickOpenCL(
3090 ImageMagickOpenCLMode magick_unused(mode),
3091 void *magick_unused(userSelectedDevice),
void *magick_unused(selectedDevice),
3094 magick_unreferenced(mode);
3095 magick_unreferenced(userSelectedDevice);
3096 magick_unreferenced(selectedDevice);
3097 magick_unreferenced(exception);
3098 return(MagickFalse);