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" 93 #if defined(MAGICKCORE_OPENCL_SUPPORT) 95 #if defined(MAGICKCORE_LTDL_DELEGATE) 99 #define NUM_CL_RAND_GENERATORS 1024 100 #define PROFILE_OCL_KERNELS 0 108 } KernelProfileRecord;
110 static const char *kernelNames[] = {
124 "LocalContrastBlurRow",
125 "LocalContrastBlurApplyColumn",
129 "RandomNumberGenerator",
132 "UnsharpMaskBlurColumn",
138 profileRecords[KERNEL_COUNT];
140 typedef struct _AccelerateTimer {
146 void startAccelerateTimer(AccelerateTimer* timer) {
148 QueryPerformanceCounter((LARGE_INTEGER*)&timer->_start);
154 timer->_start = (
long long)s.tv_sec * (
long long)1.0E3 + (
long long)s.tv_usec / (
long long)1.0E3;
158 void stopAccelerateTimer(AccelerateTimer* timer) {
161 QueryPerformanceCounter((LARGE_INTEGER*)&(n));
165 n = (
long long)s.tv_sec * (
long long)1.0E3+ (
long long)s.tv_usec / (
long long)1.0E3;
172 void resetAccelerateTimer(AccelerateTimer* timer) {
177 void initAccelerateTimer(AccelerateTimer* timer) {
179 QueryPerformanceFrequency((LARGE_INTEGER*)&timer->_freq);
181 timer->_freq = (
long long)1.0E3;
183 resetAccelerateTimer(timer);
186 double readAccelerateTimer(AccelerateTimer* timer) {
187 return (
double)timer->_clocks/(double)timer->_freq;
190 MagickPrivate MagickBooleanType RecordProfileData(
MagickCLEnv clEnv, ProfiledKernels kernel, cl_event event)
192 #if PROFILE_OCL_KERNELS 196 cl_ulong elapsed = 0;
197 clEnv->library->clWaitForEvents(1, &event);
198 status = clEnv->library->clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_START,
sizeof(cl_ulong), &start, NULL);
199 status &= clEnv->library->clGetEventProfilingInfo(event, CL_PROFILING_COMMAND_END,
sizeof(cl_ulong), &end, NULL);
200 if (status == CL_SUCCESS) {
203 elapsed = end - start;
205 LockSemaphoreInfo(clEnv->commandQueuesLock);
206 if ((elapsed < profileRecords[kernel].min) || (profileRecords[kernel].count == 0))
207 profileRecords[kernel].min = elapsed;
208 if (elapsed > profileRecords[kernel].max)
209 profileRecords[kernel].max = elapsed;
210 profileRecords[kernel].total += elapsed;
211 profileRecords[kernel].count += 1;
212 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
216 magick_unreferenced(clEnv);
217 magick_unreferenced(kernel);
218 magick_unreferenced(event);
223 void DumpProfileData()
225 #if PROFILE_OCL_KERNELS 228 OpenCLLog(
"====================================================");
239 clEnv = GetDefaultOpenCLEnv();
241 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_VENDOR, 2048, buff, NULL);
244 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_NAME, 2048, buff, NULL);
247 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DRIVER_VERSION, 2048, buff, NULL);
251 OpenCLLog(
"====================================================");
252 OpenCLLog(
" ave\tcalls \tmin -> max");
253 OpenCLLog(
" ---\t----- \t----------");
254 for (i = 0; i < KERNEL_COUNT; ++i) {
257 (void) CopyMagickString(indent,
" ",
259 strncpy(indent, kernelNames[i], min(strlen(kernelNames[i]), strlen(indent) - 1));
260 (void) FormatLocaleString(buf,
sizeof(buf),
"%s%d\t(%d calls) \t%d -> %d",
261 indent, profileRecords[i].count > 0 ? (profileRecords[i].total /
262 profileRecords[i].count) : 0, profileRecords[i].count,
263 profileRecords[i].min, profileRecords[i].max);
269 OpenCLLog(
"====================================================");
278 #ifdef MAGICKCORE_WINDOWS_SUPPORT 284 void *OsLibraryLoad(
const char *libraryName)
286 #ifdef MAGICKCORE_WINDOWS_SUPPORT 287 return (
void *)LoadLibraryA(libraryName);
289 return (
void *)dlopen(libraryName, RTLD_NOW);
294 void *OsLibraryGetFunctionAddress(
void *library,
const char *functionName)
296 #ifdef MAGICKCORE_WINDOWS_SUPPORT 297 if (!library || !functionName)
301 return (
void *) GetProcAddress( (HMODULE)library, functionName);
303 if (!library || !functionName)
307 return (
void *)dlsym(library, functionName);
334 clEnv->commandQueuesPos=-1;
335 ActivateSemaphoreInfo(&clEnv->lock);
336 ActivateSemaphoreInfo(&clEnv->commandQueuesLock);
365 MagickPrivate MagickBooleanType RelinquishMagickOpenCLEnv(
MagickCLEnv clEnv)
369 while (clEnv->commandQueuesPos >= 0)
371 clEnv->library->clReleaseCommandQueue(
372 clEnv->commandQueues[clEnv->commandQueuesPos--]);
374 if (clEnv->programs[0] != (cl_program) NULL)
375 (void) clEnv->library->clReleaseProgram(clEnv->programs[0]);
376 if (clEnv->context != (cl_context) NULL)
377 clEnv->library->clReleaseContext(clEnv->context);
378 DestroySemaphoreInfo(&clEnv->lock);
379 DestroySemaphoreInfo(&clEnv->commandQueuesLock);
380 RelinquishMagickMemory(clEnv);
396 MagickLibrary * OpenCLLib;
400 static MagickBooleanType bindOpenCLFunctions(
void* library)
402 #ifdef MAGICKCORE_HAVE_OPENCL_CL_H 403 #define BIND(X) OpenCLLib->X= &X; 406 if ((OpenCLLib->X=(MAGICKpfn_##X)OsLibraryGetFunctionAddress(library,#X)) == NULL)\ 410 BIND(clGetPlatformIDs);
411 BIND(clGetPlatformInfo);
413 BIND(clGetDeviceIDs);
414 BIND(clGetDeviceInfo);
416 BIND(clCreateContext);
417 BIND(clReleaseContext);
419 BIND(clCreateBuffer);
420 BIND(clRetainMemObject);
421 BIND(clReleaseMemObject);
423 BIND(clCreateProgramWithSource);
424 BIND(clCreateProgramWithBinary);
425 BIND(clBuildProgram);
426 BIND(clReleaseProgram);
427 BIND(clGetProgramInfo);
428 BIND(clGetProgramBuildInfo);
430 BIND(clCreateKernel);
431 BIND(clReleaseKernel);
432 BIND(clSetKernelArg);
437 BIND(clEnqueueNDRangeKernel);
438 BIND(clEnqueueReadBuffer);
439 BIND(clEnqueueMapBuffer);
440 BIND(clEnqueueUnmapMemObject);
442 BIND(clCreateCommandQueue);
443 BIND(clReleaseCommandQueue);
445 BIND(clGetEventProfilingInfo);
446 BIND(clGetEventInfo);
447 BIND(clWaitForEvents);
448 BIND(clReleaseEvent);
450 BIND(clSetEventCallback);
455 MagickLibrary * GetOpenCLLib()
457 if (OpenCLLib == NULL)
459 if (OpenCLLibLock == NULL)
461 ActivateSemaphoreInfo(&OpenCLLibLock);
464 LockSemaphoreInfo(OpenCLLibLock);
466 OpenCLLib = (MagickLibrary *) AcquireMagickMemory (
sizeof (MagickLibrary));
468 if (OpenCLLib != NULL)
470 MagickBooleanType status = MagickFalse;
471 void * library = NULL;
473 #ifdef MAGICKCORE_OPENCL_MACOSX 474 status = bindOpenCLFunctions(library);
477 memset(OpenCLLib, 0,
sizeof(MagickLibrary));
478 #ifdef MAGICKCORE_WINDOWS_SUPPORT 479 library = OsLibraryLoad(
"OpenCL.dll");
481 library = OsLibraryLoad(
"libOpenCL.so");
484 status = bindOpenCLFunctions(library);
486 if (status==MagickTrue)
487 OpenCLLib->base=library;
489 OpenCLLib=(MagickLibrary *)RelinquishMagickMemory(OpenCLLib);
493 UnlockSemaphoreInfo(OpenCLLibLock);
526 if (defaultCLEnv == NULL)
528 if (defaultCLEnvLock == NULL)
530 ActivateSemaphoreInfo(&defaultCLEnvLock);
532 LockSemaphoreInfo(defaultCLEnvLock);
533 if (defaultCLEnv == NULL)
534 defaultCLEnv = AcquireMagickOpenCLEnv();
535 UnlockSemaphoreInfo(defaultCLEnvLock);
540 static void LockDefaultOpenCLEnv() {
541 if (defaultCLEnvLock == NULL)
543 ActivateSemaphoreInfo(&defaultCLEnvLock);
545 LockSemaphoreInfo(defaultCLEnvLock);
548 static void UnlockDefaultOpenCLEnv() {
549 if (defaultCLEnvLock == NULL)
551 ActivateSemaphoreInfo(&defaultCLEnvLock);
554 UnlockSemaphoreInfo(defaultCLEnvLock);
584 LockDefaultOpenCLEnv();
585 oldEnv = defaultCLEnv;
586 defaultCLEnv = clEnv;
587 UnlockDefaultOpenCLEnv();
624 static MagickBooleanType SetMagickOpenCLEnvParamInternal(
MagickCLEnv clEnv, MagickOpenCLEnvParam param
627 MagickBooleanType status = MagickFalse;
635 case MAGICK_OPENCL_ENV_PARAM_DEVICE:
636 if (dataSize !=
sizeof(clEnv->device))
638 clEnv->device = *((cl_device_id*)data);
639 clEnv->OpenCLInitialized = MagickFalse;
643 case MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED:
644 if (dataSize !=
sizeof(clEnv->OpenCLDisabled))
646 clEnv->OpenCLDisabled = *((MagickBooleanType*)data);
647 clEnv->OpenCLInitialized = MagickFalse;
651 case MAGICK_OPENCL_ENV_PARAM_OPENCL_INITIALIZED:
652 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"SetMagickOpenCLEnvParm cannot modify the OpenCL initialization state.",
"'%s'",
".");
655 case MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED:
656 if (dataSize !=
sizeof(clEnv->disableProgramCache))
658 clEnv->disableProgramCache = *((MagickBooleanType*)data);
659 clEnv->OpenCLInitialized = MagickFalse;
663 case MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE:
664 if (dataSize !=
sizeof(clEnv->regenerateProfile))
666 clEnv->regenerateProfile = *((MagickBooleanType*)data);
667 clEnv->OpenCLInitialized = MagickFalse;
680 MagickBooleanType SetMagickOpenCLEnvParam(
MagickCLEnv clEnv, MagickOpenCLEnvParam param
682 MagickBooleanType status = MagickFalse;
684 LockSemaphoreInfo(clEnv->lock);
685 status = SetMagickOpenCLEnvParamInternal(clEnv,param,dataSize,data,exception);
686 UnlockSemaphoreInfo(clEnv->lock);
725 MagickBooleanType GetMagickOpenCLEnvParam(
MagickCLEnv clEnv, MagickOpenCLEnvParam param
734 magick_unreferenced(exception);
736 status = MagickFalse;
744 case MAGICK_OPENCL_ENV_PARAM_DEVICE:
745 if (dataSize !=
sizeof(cl_device_id))
747 *((cl_device_id*)data) = clEnv->device;
751 case MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED:
752 if (dataSize !=
sizeof(clEnv->OpenCLDisabled))
754 *((MagickBooleanType*)data) = clEnv->OpenCLDisabled;
758 case MAGICK_OPENCL_ENV_PARAM_OPENCL_INITIALIZED:
759 if (dataSize !=
sizeof(clEnv->OpenCLDisabled))
761 *((MagickBooleanType*)data) = clEnv->OpenCLInitialized;
765 case MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED:
766 if (dataSize !=
sizeof(clEnv->disableProgramCache))
768 *((MagickBooleanType*)data) = clEnv->disableProgramCache;
772 case MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE:
773 if (dataSize !=
sizeof(clEnv->regenerateProfile))
775 *((MagickBooleanType*)data) = clEnv->regenerateProfile;
779 case MAGICK_OPENCL_ENV_PARAM_PLATFORM_VENDOR:
780 if (dataSize !=
sizeof(
char *))
782 clEnv->library->clGetPlatformInfo(clEnv->platform,CL_PLATFORM_VENDOR,0,
784 *((
char **) data)=(
char *) AcquireQuantumMemory(length,
sizeof(
char));
785 clEnv->library->clGetPlatformInfo(clEnv->platform,CL_PLATFORM_VENDOR,
786 length,*((
char **) data),NULL);
790 case MAGICK_OPENCL_ENV_PARAM_DEVICE_NAME:
791 if (dataSize !=
sizeof(
char *))
793 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_NAME,0,NULL,
795 *((
char **) data)=(
char *) AcquireQuantumMemory(length,
sizeof(
char));
796 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_NAME,length,
797 *((
char **) data),NULL);
838 return clEnv->context;
841 static char* getBinaryCLProgramName(
MagickCLEnv clEnv, MagickOpenCLProgram prog,
unsigned int signature)
845 char path[MaxTextExtent];
846 char deviceName[MaxTextExtent];
847 const char* prefix =
"magick_opencl";
848 clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_NAME, MaxTextExtent, deviceName, NULL);
853 if ( *ptr ==
' ' || *ptr ==
'\\' || *ptr ==
'/' || *ptr ==
':' || *ptr ==
'*' 854 || *ptr ==
'?' || *ptr ==
'"' || *ptr ==
'<' || *ptr ==
'>' || *ptr ==
'|')
860 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s_%s_%02d_%08x_%.20g.bin",
861 GetOpenCLCachedFilesDirectory(),DirectorySeparator,prefix,deviceName,
862 (
unsigned int) prog,signature,(
double)
sizeof(
char*)*8);
863 name = (
char*)AcquireMagickMemory(strlen(path)+1);
864 CopyMagickString(name,path,strlen(path)+1);
868 static void saveBinaryCLProgram(
MagickCLEnv clEnv,MagickOpenCLProgram prog,
885 filename=getBinaryCLProgramName(clEnv,prog,signature);
886 status=clEnv->library->clGetProgramInfo(clEnv->programs[prog],
887 CL_PROGRAM_NUM_DEVICES,
sizeof(cl_uint),&num_devices,NULL);
888 if (status != CL_SUCCESS)
890 size=num_devices*
sizeof(*program_sizes);
891 program_sizes=(
size_t*) AcquireQuantumMemory(1,size);
892 if (program_sizes == (
size_t*) NULL)
894 status=clEnv->library->clGetProgramInfo(clEnv->programs[prog],
895 CL_PROGRAM_BINARY_SIZES,size,program_sizes,NULL);
896 if (status == CL_SUCCESS)
904 binary_program_size=num_devices*
sizeof(*binary_program);
905 binary_program=(
unsigned char **) AcquireQuantumMemory(1,
906 binary_program_size);
907 if (binary_program == (
unsigned char **) NULL)
909 program_sizes=(
size_t *) RelinquishMagickMemory(program_sizes);
912 for (i = 0; i < num_devices; i++)
914 binary_program[i]=AcquireQuantumMemory(MagickMax(*(program_sizes+i),1),
915 sizeof(**binary_program));
916 if (binary_program[i] == (
unsigned char *) NULL)
918 status=CL_OUT_OF_HOST_MEMORY;
922 if (status == CL_SUCCESS)
923 status=clEnv->library->clGetProgramInfo(clEnv->programs[prog],
924 CL_PROGRAM_BINARIES,binary_program_size,binary_program,NULL);
925 if (status == CL_SUCCESS)
927 for (i = 0; i < num_devices; i++)
935 program_size=*(program_sizes+i);
936 if (program_size < 1)
938 file=open_utf8(filename,O_WRONLY | O_CREAT | O_BINARY,S_MODE);
941 write(file,binary_program[i],program_size);
945 (
void) ThrowMagickException(exception,GetMagickModule(),
946 DelegateWarning,
"Saving kernel failed.",
"`%s'",filename);
950 for (i = 0; i < num_devices; i++)
951 binary_program[i]=(
unsigned char *) RelinquishMagickMemory(
953 binary_program=(
unsigned char **) RelinquishMagickMemory(binary_program);
955 program_sizes=(
size_t *) RelinquishMagickMemory(program_sizes);
958 static MagickBooleanType loadBinaryCLProgram(
MagickCLEnv clEnv, MagickOpenCLProgram prog,
unsigned int signature)
960 MagickBooleanType loadSuccessful;
961 unsigned char* binaryProgram;
962 char* binaryFileName;
965 #ifdef MAGICKCORE_CLPERFMARKER 966 clBeginPerfMarkerAMD(__FUNCTION__,
"");
969 binaryProgram = NULL;
970 binaryFileName = NULL;
972 loadSuccessful = MagickFalse;
974 binaryFileName = getBinaryCLProgramName(clEnv, prog, signature);
975 fileHandle = fopen(binaryFileName,
"rb");
976 if (fileHandle != NULL)
981 cl_int clBinaryStatus;
985 b_error |= fseek( fileHandle, 0, SEEK_END ) < 0;
986 b_error |= ( length = ftell( fileHandle ) ) <= 0;
987 b_error |= fseek( fileHandle, 0, SEEK_SET ) < 0;
991 binaryProgram = (
unsigned char*)AcquireMagickMemory(length);
992 if (binaryProgram == NULL)
995 memset(binaryProgram, 0, length);
996 b_error |= fread(binaryProgram, 1, length, fileHandle) != length;
998 clEnv->programs[prog] = clEnv->library->clCreateProgramWithBinary(clEnv->context, 1, &clEnv->device, &length, (
const unsigned char**)&binaryProgram, &clBinaryStatus, &clStatus);
999 if (clStatus != CL_SUCCESS
1000 || clBinaryStatus != CL_SUCCESS)
1003 loadSuccessful = MagickTrue;
1007 if (fileHandle != NULL)
1009 if (binaryFileName != NULL)
1010 RelinquishMagickMemory(binaryFileName);
1011 if (binaryProgram != NULL)
1012 RelinquishMagickMemory(binaryProgram);
1014 #ifdef MAGICKCORE_CLPERFMARKER 1015 clEndPerfMarkerAMD();
1018 return loadSuccessful;
1021 static unsigned int stringSignature(
const char*
string)
1023 unsigned int stringLength;
1025 unsigned int signature;
1029 const unsigned int* u;
1032 #ifdef MAGICKCORE_CLPERFMARKER 1033 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1036 stringLength = (
unsigned int) strlen(
string);
1037 signature = stringLength;
1038 n = stringLength/
sizeof(
unsigned int);
1040 for (i = 0; i < n; i++)
1044 if (n *
sizeof(
unsigned int) != stringLength)
1047 j = n *
sizeof(
unsigned int);
1048 for (i = 0; i < 4; i++,j++)
1050 if (j < stringLength)
1059 #ifdef MAGICKCORE_CLPERFMARKER 1060 clEndPerfMarkerAMD();
1067 extern const char *accelerateKernels, *accelerateKernels2;
1071 MagickBooleanType status = MagickFalse;
1074 char* accelerateKernelsBuffer = NULL;
1077 const char* MagickOpenCLProgramStrings[MAGICK_OPENCL_NUM_PROGRAMS];
1079 char options[MaxTextExtent];
1080 unsigned int optionsSignature;
1082 #ifdef MAGICKCORE_CLPERFMARKER 1083 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1087 (void) FormatLocaleString(options, MaxTextExtent, CLOptions, (
float)QuantumRange,
1088 (float)QuantumScale, (
float)CLCharQuantumScale, (float)MagickEpsilon, (
float)MagickPI, (
unsigned int)MaxMap, (
unsigned int)MAGICKCORE_QUANTUM_DEPTH);
1103 optionsSignature = stringSignature(options);
1106 accelerateKernelsBuffer = (
char*) AcquireQuantumMemory(1,strlen(accelerateKernels)+strlen(accelerateKernels2)+1);
1107 FormatLocaleString(accelerateKernelsBuffer,strlen(accelerateKernels)+
1108 strlen(accelerateKernels2)+1,
"%s%s",accelerateKernels,accelerateKernels2);
1109 MagickOpenCLProgramStrings[MAGICK_OPENCL_ACCELERATE] = accelerateKernelsBuffer;
1111 for (i = 0; i < MAGICK_OPENCL_NUM_PROGRAMS; i++)
1113 MagickBooleanType loadSuccessful = MagickFalse;
1114 unsigned int programSignature = stringSignature(MagickOpenCLProgramStrings[i]) ^ optionsSignature;
1117 if (clEnv->disableProgramCache != MagickTrue
1118 && !getenv(
"MAGICK_OCL_REC"))
1119 loadSuccessful = loadBinaryCLProgram(clEnv, (MagickOpenCLProgram)i, programSignature);
1121 if (loadSuccessful == MagickFalse)
1124 size_t programLength = strlen(MagickOpenCLProgramStrings[i]);
1125 clEnv->programs[i] = clEnv->library->clCreateProgramWithSource(clEnv->context, 1, &(MagickOpenCLProgramStrings[i]), &programLength, &clStatus);
1126 if (clStatus!=CL_SUCCESS)
1128 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1129 "clCreateProgramWithSource failed.",
"(%d)", (int)clStatus);
1135 clStatus = clEnv->library->clBuildProgram(clEnv->programs[i], 1, &clEnv->device, options, NULL, NULL);
1136 if (clStatus!=CL_SUCCESS)
1138 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1139 "clBuildProgram failed.",
"(%d)", (int)clStatus);
1141 if (loadSuccessful == MagickFalse)
1143 char path[MaxTextExtent];
1147 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s" 1148 ,GetOpenCLCachedFilesDirectory()
1149 ,DirectorySeparator,
"magick_badcl.cl");
1150 fileHandle = fopen(path,
"wb");
1151 if (fileHandle != NULL)
1153 fwrite(MagickOpenCLProgramStrings[i],
sizeof(
char), strlen(MagickOpenCLProgramStrings[i]), fileHandle);
1161 clEnv->library->clGetProgramBuildInfo(clEnv->programs[i], clEnv->device, CL_PROGRAM_BUILD_LOG, 0, NULL, &logSize);
1162 log = (
char*)AcquireCriticalMemory(logSize);
1163 clEnv->library->clGetProgramBuildInfo(clEnv->programs[i], clEnv->device, CL_PROGRAM_BUILD_LOG, logSize, log, &logSize);
1165 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s" 1166 ,GetOpenCLCachedFilesDirectory()
1167 ,DirectorySeparator,
"magick_badcl_build.log");
1168 fileHandle = fopen(path,
"wb");
1169 if (fileHandle != NULL)
1171 const char* buildOptionsTitle =
"build options: ";
1172 fwrite(buildOptionsTitle,
sizeof(
char), strlen(buildOptionsTitle), fileHandle);
1173 fwrite(options,
sizeof(
char), strlen(options), fileHandle);
1174 fwrite(
"\n",
sizeof(
char), 1, fileHandle);
1175 fwrite(log,
sizeof(
char), logSize, fileHandle);
1178 RelinquishMagickMemory(log);
1184 if (loadSuccessful == MagickFalse)
1187 saveBinaryCLProgram(clEnv, (MagickOpenCLProgram)i, programSignature, exception);
1191 status = MagickTrue;
1195 if (accelerateKernelsBuffer!=NULL) RelinquishMagickMemory(accelerateKernelsBuffer);
1197 #ifdef MAGICKCORE_CLPERFMARKER 1198 clEndPerfMarkerAMD();
1207 cl_uint numPlatforms = 0;
1208 cl_platform_id *platforms = NULL;
1209 char* MAGICK_OCL_DEVICE = NULL;
1210 MagickBooleanType OpenCLAvailable = MagickFalse;
1212 #ifdef MAGICKCORE_CLPERFMARKER 1213 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1217 MAGICK_OCL_DEVICE = getenv(
"MAGICK_OCL_DEVICE");
1218 if (MAGICK_OCL_DEVICE == (
char *) NULL)
1219 return(MagickFalse);
1220 if (strcmp(MAGICK_OCL_DEVICE,
"CPU") == 0)
1221 clEnv->deviceType = CL_DEVICE_TYPE_CPU;
1222 else if (strcmp(MAGICK_OCL_DEVICE,
"GPU") == 0)
1223 clEnv->deviceType = CL_DEVICE_TYPE_GPU;
1224 else if (IsStringTrue(MAGICK_OCL_DEVICE) != MagickFalse)
1226 if (clEnv->deviceType == 0)
1227 clEnv->deviceType = CL_DEVICE_TYPE_ALL;
1230 return(MagickFalse);
1232 if (clEnv->device != NULL)
1234 status = clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_PLATFORM,
sizeof(cl_platform_id), &clEnv->platform, NULL);
1235 if (status != CL_SUCCESS) {
1236 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1237 "Failed to get OpenCL platform from the selected device.",
"(%d)", status);
1241 else if (clEnv->platform != NULL)
1244 platforms = (cl_platform_id *) AcquireQuantumMemory(1,numPlatforms *
sizeof(cl_platform_id));
1245 if (platforms == (cl_platform_id *) NULL)
1247 (void) ThrowMagickException(exception, GetMagickModule(), ResourceLimitError,
1248 "AcquireMagickMemory failed.",
".");
1251 platforms[0] = clEnv->platform;
1255 clEnv->device = NULL;
1258 status = clEnv->library->clGetPlatformIDs(0, NULL, &numPlatforms);
1259 if (status != CL_SUCCESS)
1261 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1262 "clGetplatformIDs failed.",
"(%d)", status);
1267 if (numPlatforms == 0) {
1271 platforms = (cl_platform_id *) AcquireQuantumMemory(1,numPlatforms *
sizeof(cl_platform_id));
1272 if (platforms == (cl_platform_id *) NULL)
1274 (void) ThrowMagickException(exception, GetMagickModule(), ResourceLimitError,
1275 "AcquireMagickMemory failed.",
".");
1279 status = clEnv->library->clGetPlatformIDs(numPlatforms, platforms, NULL);
1280 if (status != CL_SUCCESS)
1282 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1283 "clGetPlatformIDs failed.",
"(%d)", status);
1289 clEnv->device = NULL;
1290 for (j = 0; j < 2; j++)
1293 cl_device_type deviceType;
1294 if (clEnv->deviceType == CL_DEVICE_TYPE_ALL)
1297 deviceType = CL_DEVICE_TYPE_GPU;
1299 deviceType = CL_DEVICE_TYPE_CPU;
1306 deviceType = clEnv->deviceType;
1308 for (i = 0; i < numPlatforms; i++)
1310 char version[MaxTextExtent];
1312 status = clEnv->library->clGetPlatformInfo(clEnv->platform, CL_PLATFORM_VERSION, MaxTextExtent, version, NULL);
1313 if (status != CL_SUCCESS)
1315 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1316 "clGetPlatformInfo failed.",
"(%d)", status);
1319 if (strncmp(version,
"OpenCL 1.0 ",11) == 0)
1321 status = clEnv->library->clGetDeviceIDs(platforms[i], deviceType, 1, &(clEnv->device), &numDevices);
1322 if (status != CL_SUCCESS)
1324 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1325 "clGetDeviceIDs failed.",
"(%d)", status);
1328 if (clEnv->device != NULL)
1330 clEnv->platform = platforms[i];
1337 if (platforms!=NULL)
1338 RelinquishMagickMemory(platforms);
1340 OpenCLAvailable = (clEnv->platform!=NULL
1341 && clEnv->device!=NULL)?MagickTrue:MagickFalse;
1343 #ifdef MAGICKCORE_CLPERFMARKER 1344 clEndPerfMarkerAMD();
1347 return OpenCLAvailable;
1350 static MagickBooleanType EnableOpenCLInternal(
MagickCLEnv clEnv) {
1351 if (clEnv->OpenCLInitialized != MagickFalse
1352 && clEnv->platform != NULL
1353 && clEnv->device != NULL) {
1354 clEnv->OpenCLDisabled = MagickFalse;
1357 clEnv->OpenCLDisabled = MagickTrue;
1388 static void RelinquishCommandQueues(
MagickCLEnv clEnv)
1393 LockSemaphoreInfo(clEnv->commandQueuesLock);
1394 while (clEnv->commandQueuesPos >= 0)
1395 clEnv->library->clReleaseCommandQueue(
1396 clEnv->commandQueues[clEnv->commandQueuesPos--]);
1397 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1402 MagickBooleanType status = MagickTrue;
1404 cl_context_properties cps[3];
1406 #ifdef MAGICKCORE_CLPERFMARKER 1408 int status = clInitializePerfMarkerAMD();
1409 if (status == AP_SUCCESS) {
1414 clEnv->OpenCLInitialized = MagickTrue;
1417 OpenCLLib=GetOpenCLLib();
1420 clEnv->library=OpenCLLib;
1425 MagickBooleanType flag;
1427 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
1428 ,
sizeof(MagickBooleanType), &flag, exception);
1431 if (clEnv->OpenCLDisabled != MagickFalse)
1434 clEnv->OpenCLDisabled = MagickTrue;
1436 status = InitOpenCLPlatformDevice(clEnv, exception);
1437 if (status == MagickFalse) {
1443 cps[0] = CL_CONTEXT_PLATFORM;
1444 cps[1] = (cl_context_properties)clEnv->platform;
1446 clEnv->context = clEnv->library->clCreateContext(cps, 1, &(clEnv->device), NULL, NULL, &clStatus);
1447 if (clStatus != CL_SUCCESS)
1449 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1450 "clCreateContext failed.",
"(%d)", clStatus);
1451 status = MagickFalse;
1455 RelinquishCommandQueues(clEnv);
1457 status = CompileOpenCLKernels(clEnv, exception);
1458 if (status == MagickFalse) {
1459 (void) ThrowMagickException(exception, GetMagickModule(), DelegateWarning,
1460 "clCreateCommandQueue failed.",
"(%d)", status);
1465 status = EnableOpenCLInternal(clEnv);
1474 MagickBooleanType status = MagickFalse;
1476 if ((clEnv == NULL) || (getenv(
"MAGICK_OCL_DEVICE") == (
const char *) NULL))
1479 #ifdef MAGICKCORE_CLPERFMARKER 1480 clBeginPerfMarkerAMD(__FUNCTION__,
"");
1483 LockSemaphoreInfo(clEnv->lock);
1484 if (clEnv->OpenCLInitialized == MagickFalse) {
1485 if (clEnv->device==NULL && clEnv->OpenCLDisabled == MagickFalse)
1486 status = autoSelectDevice(clEnv, exception);
1488 status = InitOpenCLEnvInternal(clEnv, exception);
1490 UnlockSemaphoreInfo(clEnv->lock);
1492 #ifdef MAGICKCORE_CLPERFMARKER 1493 clEndPerfMarkerAMD();
1522 MagickPrivate cl_command_queue AcquireOpenCLCommandQueue(
MagickCLEnv clEnv)
1527 cl_command_queue_properties
1531 return (cl_command_queue) NULL;
1532 LockSemaphoreInfo(clEnv->commandQueuesLock);
1533 if (clEnv->commandQueuesPos >= 0) {
1534 queue=clEnv->commandQueues[clEnv->commandQueuesPos--];
1535 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1538 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1540 #if PROFILE_OCL_KERNELS 1541 properties=CL_QUEUE_PROFILING_ENABLE;
1543 queue=clEnv->library->clCreateCommandQueue(clEnv->context,clEnv->device,
1576 MagickPrivate MagickBooleanType RelinquishOpenCLCommandQueue(
MagickCLEnv clEnv,
1577 cl_command_queue queue)
1583 return(MagickFalse);
1585 LockSemaphoreInfo(clEnv->commandQueuesLock);
1587 if (clEnv->commandQueuesPos >= MAX_COMMAND_QUEUES-1)
1589 clEnv->library->clFinish(queue);
1590 status=(clEnv->library->clReleaseCommandQueue(queue) == CL_SUCCESS) ?
1591 MagickTrue : MagickFalse;
1595 clEnv->library->clFlush(queue);
1596 clEnv->commandQueues[++clEnv->commandQueuesPos]=queue;
1600 UnlockSemaphoreInfo(clEnv->commandQueuesLock);
1634 cl_kernel AcquireOpenCLKernel(
MagickCLEnv clEnv, MagickOpenCLProgram program,
const char* kernelName)
1637 cl_kernel kernel = NULL;
1638 if (clEnv != NULL && kernelName!=NULL)
1640 kernel = clEnv->library->clCreateKernel(clEnv->programs[program], kernelName, &clStatus);
1674 MagickBooleanType RelinquishOpenCLKernel(
MagickCLEnv clEnv, cl_kernel kernel)
1676 MagickBooleanType status = MagickFalse;
1677 if (clEnv != NULL && kernel != NULL)
1679 status = ((clEnv->library->clReleaseKernel(kernel) == CL_SUCCESS)?MagickTrue:MagickFalse);
1709 unsigned long GetOpenCLDeviceLocalMemorySize(
MagickCLEnv clEnv)
1711 cl_ulong localMemorySize;
1712 clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_LOCAL_MEM_SIZE,
sizeof(cl_ulong), &localMemorySize, NULL);
1713 return (
unsigned long)localMemorySize;
1717 unsigned long GetOpenCLDeviceMaxMemAllocSize(
MagickCLEnv clEnv)
1719 cl_ulong maxMemAllocSize;
1720 clEnv->library->clGetDeviceInfo(clEnv->device, CL_DEVICE_MAX_MEM_ALLOC_SIZE,
sizeof(cl_ulong), &maxMemAllocSize, NULL);
1721 return (
unsigned long)maxMemAllocSize;
1732 ,DS_INVALID_PROFILE = 1000
1734 ,DS_INVALID_PERF_EVALUATOR_TYPE
1735 ,DS_INVALID_PERF_EVALUATOR
1736 ,DS_PERF_EVALUATOR_ERROR
1738 ,DS_UNKNOWN_DEVICE_TYPE
1739 ,DS_PROFILE_FILE_ERROR
1740 ,DS_SCORE_SERIALIZER_ERROR
1741 ,DS_SCORE_DESERIALIZER_ERROR
1746 DS_DEVICE_NATIVE_CPU = 0
1747 ,DS_DEVICE_OPENCL_DEVICE
1752 ds_device_type type;
1753 cl_device_type oclDeviceType;
1754 cl_device_id oclDeviceID;
1755 char* oclDeviceName;
1756 char* oclDriverVersion;
1757 cl_uint oclMaxClockFrequency;
1758 cl_uint oclMaxComputeUnits;
1763 unsigned int numDevices;
1765 const char* version;
1769 typedef ds_status (*ds_score_release)(
void* score);
1771 static ds_status releaseDeviceResource(ds_device* device, ds_score_release sr) {
1772 ds_status status = DS_SUCCESS;
1774 if (device->oclDeviceName) RelinquishMagickMemory(device->oclDeviceName);
1775 if (device->oclDriverVersion) RelinquishMagickMemory(device->oclDriverVersion);
1776 if (device->score) status = sr(device->score);
1781 static ds_status releaseDSProfile(ds_profile* profile, ds_score_release sr) {
1782 ds_status status = DS_SUCCESS;
1783 if (profile!=NULL) {
1784 if (profile->devices!=NULL && sr!=NULL) {
1786 for (i = 0; i < profile->numDevices; i++) {
1787 status = releaseDeviceResource(profile->devices+i,sr);
1788 if (status != DS_SUCCESS)
1791 RelinquishMagickMemory(profile->devices);
1793 RelinquishMagickMemory(profile);
1799 static ds_status initDSProfile(ds_profile** p,
const char* version) {
1801 cl_uint numPlatforms = 0;
1802 cl_platform_id* platforms = NULL;
1803 cl_device_id* devices = NULL;
1804 ds_status status = DS_SUCCESS;
1805 ds_profile* profile = NULL;
1806 unsigned int next = 0;
1810 return DS_INVALID_PROFILE;
1812 profile = (ds_profile*) AcquireMagickMemory(
sizeof(ds_profile));
1813 if (profile == NULL)
1814 return DS_MEMORY_ERROR;
1816 memset(profile, 0,
sizeof(ds_profile));
1818 OpenCLLib->clGetPlatformIDs(0, NULL, &numPlatforms);
1819 if (numPlatforms > 0) {
1820 platforms = (cl_platform_id*) AcquireQuantumMemory(numPlatforms,
sizeof(cl_platform_id));
1821 if (platforms == NULL) {
1822 status = DS_MEMORY_ERROR;
1825 OpenCLLib->clGetPlatformIDs(numPlatforms, platforms, NULL);
1826 for (i = 0; i < (
unsigned int)numPlatforms; i++) {
1828 if (OpenCLLib->clGetDeviceIDs(platforms[i], CL_DEVICE_TYPE_CPU | CL_DEVICE_TYPE_GPU, 0, NULL, &num) == CL_SUCCESS)
1833 profile->numDevices = numDevices+1;
1835 profile->devices = (ds_device*) AcquireQuantumMemory(profile->numDevices,
sizeof(ds_device));
1836 if (profile->devices == NULL) {
1837 profile->numDevices = 0;
1838 status = DS_MEMORY_ERROR;
1841 memset(profile->devices, 0, profile->numDevices*
sizeof(ds_device));
1843 if (numDevices > 0) {
1844 devices = (cl_device_id*) AcquireQuantumMemory(numDevices,
sizeof(cl_device_id));
1845 if (devices == NULL) {
1846 status = DS_MEMORY_ERROR;
1849 for (i = 0; i < (
unsigned int)numPlatforms; i++) {
1853 for (d = 0; d < 2; d++) {
1855 cl_device_type deviceType;
1858 deviceType = CL_DEVICE_TYPE_GPU;
1861 deviceType = CL_DEVICE_TYPE_CPU;
1867 if (OpenCLLib->clGetDeviceIDs(platforms[i], deviceType, numDevices, devices, &num) != CL_SUCCESS)
1869 for (j = 0; j < num; j++, next++) {
1872 profile->devices[next].type = DS_DEVICE_OPENCL_DEVICE;
1873 profile->devices[next].oclDeviceID = devices[j];
1875 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_NAME
1876 , 0, NULL, &length);
1877 profile->devices[next].oclDeviceName = (
char*) AcquireQuantumMemory(length,
sizeof(
char));
1878 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_NAME
1879 , length, profile->devices[next].oclDeviceName, NULL);
1881 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DRIVER_VERSION
1882 , 0, NULL, &length);
1883 profile->devices[next].oclDriverVersion = (
char*) AcquireQuantumMemory(length,
sizeof(
char));
1884 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DRIVER_VERSION
1885 , length, profile->devices[next].oclDriverVersion, NULL);
1887 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_MAX_CLOCK_FREQUENCY
1888 ,
sizeof(cl_uint), &profile->devices[next].oclMaxClockFrequency, NULL);
1890 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_MAX_COMPUTE_UNITS
1891 ,
sizeof(cl_uint), &profile->devices[next].oclMaxComputeUnits, NULL);
1893 OpenCLLib->clGetDeviceInfo(profile->devices[next].oclDeviceID, CL_DEVICE_TYPE
1894 ,
sizeof(cl_device_type), &profile->devices[next].oclDeviceType, NULL);
1900 profile->devices[next].type = DS_DEVICE_NATIVE_CPU;
1901 profile->version = version;
1904 if (platforms) RelinquishMagickMemory(platforms);
1905 if (devices) RelinquishMagickMemory(devices);
1906 if (status == DS_SUCCESS) {
1911 if (profile->devices)
1912 RelinquishMagickMemory(profile->devices);
1913 RelinquishMagickMemory(profile);
1923 typedef ds_status (*ds_perf_evaluator)(ds_device* device,
void* data);
1927 ,DS_EVALUATE_NEW_ONLY
1928 } ds_evaluation_type;
1930 static ds_status profileDevices(ds_profile* profile,
const ds_evaluation_type type
1931 ,ds_perf_evaluator evaluator,
void* evaluatorData,
unsigned int* numUpdates) {
1932 ds_status status = DS_SUCCESS;
1934 unsigned int updates = 0;
1936 if (profile == NULL) {
1937 return DS_INVALID_PROFILE;
1939 if (evaluator == NULL) {
1940 return DS_INVALID_PERF_EVALUATOR;
1943 for (i = 0; i < profile->numDevices; i++) {
1944 ds_status evaluatorStatus;
1947 case DS_EVALUATE_NEW_ONLY:
1948 if (profile->devices[i].score != NULL)
1951 case DS_EVALUATE_ALL:
1952 evaluatorStatus = evaluator(profile->devices+i,evaluatorData);
1953 if (evaluatorStatus != DS_SUCCESS) {
1954 status = evaluatorStatus;
1960 return DS_INVALID_PERF_EVALUATOR_TYPE;
1965 *numUpdates = updates;
1970 #define DS_TAG_VERSION "<version>" 1971 #define DS_TAG_VERSION_END "</version>" 1972 #define DS_TAG_DEVICE "<device>" 1973 #define DS_TAG_DEVICE_END "</device>" 1974 #define DS_TAG_SCORE "<score>" 1975 #define DS_TAG_SCORE_END "</score>" 1976 #define DS_TAG_DEVICE_TYPE "<type>" 1977 #define DS_TAG_DEVICE_TYPE_END "</type>" 1978 #define DS_TAG_DEVICE_NAME "<name>" 1979 #define DS_TAG_DEVICE_NAME_END "</name>" 1980 #define DS_TAG_DEVICE_DRIVER_VERSION "<driver>" 1981 #define DS_TAG_DEVICE_DRIVER_VERSION_END "</driver>" 1982 #define DS_TAG_DEVICE_MAX_COMPUTE_UNITS "<max cu>" 1983 #define DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END "</max cu>" 1984 #define DS_TAG_DEVICE_MAX_CLOCK_FREQ "<max clock>" 1985 #define DS_TAG_DEVICE_MAX_CLOCK_FREQ_END "</max clock>" 1987 #define DS_DEVICE_NATIVE_CPU_STRING "native_cpu" 1991 typedef ds_status (*ds_score_serializer)(ds_device* device,
void** serializedScore,
unsigned int* serializedScoreSize);
1992 static ds_status writeProfileToFile(ds_profile* profile, ds_score_serializer serializer,
const char* file) {
1993 ds_status status = DS_SUCCESS;
1994 FILE* profileFile = NULL;
1997 if (profile == NULL)
1998 return DS_INVALID_PROFILE;
2000 profileFile = fopen(file,
"wb");
2001 if (profileFile==NULL) {
2002 status = DS_FILE_ERROR;
2008 fwrite(DS_TAG_VERSION,
sizeof(
char), strlen(DS_TAG_VERSION), profileFile);
2009 fwrite(profile->version,
sizeof(
char), strlen(profile->version), profileFile);
2010 fwrite(DS_TAG_VERSION_END,
sizeof(
char), strlen(DS_TAG_VERSION_END), profileFile);
2011 fwrite(
"\n",
sizeof(
char), 1, profileFile);
2013 for (i = 0; i < profile->numDevices && status == DS_SUCCESS; i++) {
2014 void* serializedScore;
2015 unsigned int serializedScoreSize;
2017 fwrite(DS_TAG_DEVICE,
sizeof(
char), strlen(DS_TAG_DEVICE), profileFile);
2019 fwrite(DS_TAG_DEVICE_TYPE,
sizeof(
char), strlen(DS_TAG_DEVICE_TYPE), profileFile);
2020 fwrite(&profile->devices[i].type,
sizeof(ds_device_type),1, profileFile);
2021 fwrite(DS_TAG_DEVICE_TYPE_END,
sizeof(
char), strlen(DS_TAG_DEVICE_TYPE_END), profileFile);
2023 switch(profile->devices[i].type) {
2024 case DS_DEVICE_NATIVE_CPU:
2034 case DS_DEVICE_OPENCL_DEVICE:
2038 fwrite(DS_TAG_DEVICE_NAME,
sizeof(
char), strlen(DS_TAG_DEVICE_NAME), profileFile);
2039 fwrite(profile->devices[i].oclDeviceName,
sizeof(
char),strlen(profile->devices[i].oclDeviceName), profileFile);
2040 fwrite(DS_TAG_DEVICE_NAME_END,
sizeof(
char), strlen(DS_TAG_DEVICE_NAME_END), profileFile);
2042 fwrite(DS_TAG_DEVICE_DRIVER_VERSION,
sizeof(
char), strlen(DS_TAG_DEVICE_DRIVER_VERSION), profileFile);
2043 fwrite(profile->devices[i].oclDriverVersion,
sizeof(
char),strlen(profile->devices[i].oclDriverVersion), profileFile);
2044 fwrite(DS_TAG_DEVICE_DRIVER_VERSION_END,
sizeof(
char), strlen(DS_TAG_DEVICE_DRIVER_VERSION_END), profileFile);
2046 fwrite(DS_TAG_DEVICE_MAX_COMPUTE_UNITS,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_COMPUTE_UNITS), profileFile);
2047 (void) FormatLocaleString(tmp,
sizeof(tmp),
"%d",
2048 profile->devices[i].oclMaxComputeUnits);
2049 fwrite(tmp,
sizeof(
char),strlen(tmp), profileFile);
2050 fwrite(DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END), profileFile);
2052 fwrite(DS_TAG_DEVICE_MAX_CLOCK_FREQ,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_CLOCK_FREQ), profileFile);
2053 (void) FormatLocaleString(tmp,
sizeof(tmp),
"%d",
2054 profile->devices[i].oclMaxClockFrequency);
2055 fwrite(tmp,
sizeof(
char),strlen(tmp), profileFile);
2056 fwrite(DS_TAG_DEVICE_MAX_CLOCK_FREQ_END,
sizeof(
char), strlen(DS_TAG_DEVICE_MAX_CLOCK_FREQ_END), profileFile);
2060 status = DS_UNKNOWN_DEVICE_TYPE;
2064 fwrite(DS_TAG_SCORE,
sizeof(
char), strlen(DS_TAG_SCORE), profileFile);
2065 status = serializer(profile->devices+i, &serializedScore, &serializedScoreSize);
2066 if (status == DS_SUCCESS && serializedScore!=NULL && serializedScoreSize > 0) {
2067 fwrite(serializedScore,
sizeof(
char), serializedScoreSize, profileFile);
2068 RelinquishMagickMemory(serializedScore);
2070 fwrite(DS_TAG_SCORE_END,
sizeof(
char), strlen(DS_TAG_SCORE_END), profileFile);
2071 fwrite(DS_TAG_DEVICE_END,
sizeof(
char), strlen(DS_TAG_DEVICE_END), profileFile);
2072 fwrite(
"\n",
sizeof(
char),1,profileFile);
2074 fclose(profileFile);
2080 static ds_status readProFile(
const char* fileName,
char** content,
size_t* contentSize) {
2081 ds_status status = DS_SUCCESS;
2082 FILE * input = NULL;
2085 char* binary = NULL;
2090 input = fopen(fileName,
"rb");
2092 return DS_FILE_ERROR;
2095 fseek(input, 0L, SEEK_END);
2096 size = ftell(input);
2098 binary = (
char*) AcquireQuantumMemory(1,size);
2099 if(binary == NULL) {
2100 status = DS_FILE_ERROR;
2103 rsize = fread(binary,
sizeof(
char), size, input);
2106 status = DS_FILE_ERROR;
2109 *contentSize = size;
2113 if (input != NULL) fclose(input);
2114 if (status != DS_SUCCESS
2115 && binary != NULL) {
2116 RelinquishMagickMemory(binary);
2124 static const char* findString(
const char* contentStart,
const char* contentEnd,
const char*
string) {
2125 size_t stringLength;
2126 const char* currentPosition;
2129 stringLength = strlen(
string);
2130 currentPosition = contentStart;
2131 for(currentPosition = contentStart; currentPosition < contentEnd; currentPosition++) {
2132 if (*currentPosition ==
string[0]) {
2133 if (currentPosition+stringLength < contentEnd) {
2134 if (strncmp(currentPosition,
string, stringLength) == 0) {
2135 found = currentPosition;
2145 typedef ds_status (*ds_score_deserializer)(ds_device* device,
const unsigned char* serializedScore,
unsigned int serializedScoreSize);
2146 static ds_status readProfileFromFile(ds_profile* profile, ds_score_deserializer deserializer,
const char* file) {
2148 ds_status status = DS_SUCCESS;
2149 char* contentStart = NULL;
2150 const char* contentEnd = NULL;
2154 return DS_INVALID_PROFILE;
2156 status = readProFile(file, &contentStart, &contentSize);
2157 if (status == DS_SUCCESS) {
2158 const char* currentPosition;
2159 const char* dataStart;
2160 const char* dataEnd;
2161 size_t versionStringLength;
2163 contentEnd = contentStart + contentSize;
2164 currentPosition = contentStart;
2168 dataStart = findString(currentPosition, contentEnd, DS_TAG_VERSION);
2169 if (dataStart == NULL) {
2170 status = DS_PROFILE_FILE_ERROR;
2173 dataStart += strlen(DS_TAG_VERSION);
2175 dataEnd = findString(dataStart, contentEnd, DS_TAG_VERSION_END);
2176 if (dataEnd==NULL) {
2177 status = DS_PROFILE_FILE_ERROR;
2181 versionStringLength = strlen(profile->version);
2182 if (versionStringLength!=(
size_t)(dataEnd-dataStart)
2183 || strncmp(profile->version, dataStart, versionStringLength)!=(
int)0) {
2185 status = DS_PROFILE_FILE_ERROR;
2188 currentPosition = dataEnd+strlen(DS_TAG_VERSION_END);
2191 DisableMSCWarning(4127)
2196 const char* deviceTypeStart;
2197 const char* deviceTypeEnd;
2198 ds_device_type deviceType;
2200 const char* deviceNameStart;
2201 const char* deviceNameEnd;
2203 const char* deviceScoreStart;
2204 const char* deviceScoreEnd;
2206 const char* deviceDriverStart;
2207 const char* deviceDriverEnd;
2209 const char* tmpStart;
2213 cl_uint maxClockFrequency;
2214 cl_uint maxComputeUnits;
2216 dataStart = findString(currentPosition, contentEnd, DS_TAG_DEVICE);
2217 if (dataStart==NULL) {
2221 dataStart+=strlen(DS_TAG_DEVICE);
2222 dataEnd = findString(dataStart, contentEnd, DS_TAG_DEVICE_END);
2223 if (dataEnd==NULL) {
2224 status = DS_PROFILE_FILE_ERROR;
2229 deviceTypeStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_TYPE);
2230 if (deviceTypeStart==NULL) {
2231 status = DS_PROFILE_FILE_ERROR;
2234 deviceTypeStart+=strlen(DS_TAG_DEVICE_TYPE);
2235 deviceTypeEnd = findString(deviceTypeStart, contentEnd, DS_TAG_DEVICE_TYPE_END);
2236 if (deviceTypeEnd==NULL) {
2237 status = DS_PROFILE_FILE_ERROR;
2240 memcpy(&deviceType, deviceTypeStart,
sizeof(ds_device_type));
2244 if (deviceType == DS_DEVICE_OPENCL_DEVICE) {
2246 deviceNameStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_NAME);
2247 if (deviceNameStart==NULL) {
2248 status = DS_PROFILE_FILE_ERROR;
2251 deviceNameStart+=strlen(DS_TAG_DEVICE_NAME);
2252 deviceNameEnd = findString(deviceNameStart, contentEnd, DS_TAG_DEVICE_NAME_END);
2253 if (deviceNameEnd==NULL) {
2254 status = DS_PROFILE_FILE_ERROR;
2259 deviceDriverStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_DRIVER_VERSION);
2260 if (deviceDriverStart==NULL) {
2261 status = DS_PROFILE_FILE_ERROR;
2264 deviceDriverStart+=strlen(DS_TAG_DEVICE_DRIVER_VERSION);
2265 deviceDriverEnd = findString(deviceDriverStart, contentEnd, DS_TAG_DEVICE_DRIVER_VERSION_END);
2266 if (deviceDriverEnd ==NULL) {
2267 status = DS_PROFILE_FILE_ERROR;
2272 tmpStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_MAX_COMPUTE_UNITS);
2273 if (tmpStart==NULL) {
2274 status = DS_PROFILE_FILE_ERROR;
2277 tmpStart+=strlen(DS_TAG_DEVICE_MAX_COMPUTE_UNITS);
2278 tmpEnd = findString(tmpStart, contentEnd, DS_TAG_DEVICE_MAX_COMPUTE_UNITS_END);
2279 if (tmpEnd ==NULL) {
2280 status = DS_PROFILE_FILE_ERROR;
2283 memcpy(tmp,tmpStart,tmpEnd-tmpStart);
2284 tmp[tmpEnd-tmpStart] =
'\0';
2285 maxComputeUnits = strtol(tmp,(
char **) NULL,10);
2288 tmpStart = findString(dataStart, contentEnd, DS_TAG_DEVICE_MAX_CLOCK_FREQ);
2289 if (tmpStart==NULL) {
2290 status = DS_PROFILE_FILE_ERROR;
2293 tmpStart+=strlen(DS_TAG_DEVICE_MAX_CLOCK_FREQ);
2294 tmpEnd = findString(tmpStart, contentEnd, DS_TAG_DEVICE_MAX_CLOCK_FREQ_END);
2295 if (tmpEnd ==NULL) {
2296 status = DS_PROFILE_FILE_ERROR;
2299 memcpy(tmp,tmpStart,tmpEnd-tmpStart);
2300 tmp[tmpEnd-tmpStart] =
'\0';
2301 maxClockFrequency = strtol(tmp,(
char **) NULL,10);
2305 for (i = 0; i < profile->numDevices; i++) {
2306 if (profile->devices[i].type == DS_DEVICE_OPENCL_DEVICE) {
2307 size_t actualDeviceNameLength;
2308 size_t driverVersionLength;
2310 actualDeviceNameLength = strlen(profile->devices[i].oclDeviceName);
2311 driverVersionLength = strlen(profile->devices[i].oclDriverVersion);
2312 if (actualDeviceNameLength == (
size_t)(deviceNameEnd - deviceNameStart)
2313 && driverVersionLength == (
size_t)(deviceDriverEnd - deviceDriverStart)
2314 && maxComputeUnits == profile->devices[i].oclMaxComputeUnits
2315 && maxClockFrequency == profile->devices[i].oclMaxClockFrequency
2316 && strncmp(profile->devices[i].oclDeviceName, deviceNameStart, actualDeviceNameLength)==(
int)0
2317 && strncmp(profile->devices[i].oclDriverVersion, deviceDriverStart, driverVersionLength)==(
int)0) {
2319 deviceScoreStart = findString(dataStart, contentEnd, DS_TAG_SCORE);
2320 if (deviceNameStart==NULL) {
2321 status = DS_PROFILE_FILE_ERROR;
2324 deviceScoreStart+=strlen(DS_TAG_SCORE);
2325 deviceScoreEnd = findString(deviceScoreStart, contentEnd, DS_TAG_SCORE_END);
2326 status = deserializer(profile->devices+i, (
const unsigned char*)deviceScoreStart, deviceScoreEnd-deviceScoreStart);
2327 if (status != DS_SUCCESS) {
2335 else if (deviceType == DS_DEVICE_NATIVE_CPU) {
2336 for (i = 0; i < profile->numDevices; i++) {
2337 if (profile->devices[i].type == DS_DEVICE_NATIVE_CPU) {
2338 deviceScoreStart = findString(dataStart, contentEnd, DS_TAG_SCORE);
2339 if (deviceScoreStart==NULL) {
2340 status = DS_PROFILE_FILE_ERROR;
2343 deviceScoreStart+=strlen(DS_TAG_SCORE);
2344 deviceScoreEnd = findString(deviceScoreStart, contentEnd, DS_TAG_SCORE_END);
2345 status = deserializer(profile->devices+i, (
const unsigned char*)deviceScoreStart, deviceScoreEnd-deviceScoreStart);
2346 if (status != DS_SUCCESS) {
2354 currentPosition = dataEnd+strlen(DS_TAG_DEVICE_END);
2358 if (contentStart!=NULL) RelinquishMagickMemory(contentStart);
2364 static ds_status getNumDeviceWithEmptyScore(ds_profile* profile,
unsigned int* num) {
2366 if (profile == NULL || num==NULL)
2367 return DS_MEMORY_ERROR;
2369 for (i = 0; i < profile->numDevices; i++) {
2370 if (profile->devices[i].score == NULL) {
2383 typedef double AccelerateScoreType;
2385 static ds_status AcceleratePerfEvaluator(ds_device *device,
2386 void *magick_unused(data))
2388 #define ACCELERATE_PERF_DIMEN "2048x1536" 2390 #define ReturnStatus(status) \ 2392 if (oldClEnv != (MagickCLEnv) NULL) \ 2393 defaultCLEnv=oldClEnv; \ 2394 if (clEnv != (MagickCLEnv) NULL) \ 2395 (void) RelinquishMagickOpenCLEnv(clEnv); \ 2412 magick_unreferenced(data);
2415 ReturnStatus(DS_PERF_EVALUATOR_ERROR);
2417 clEnv=AcquireMagickOpenCLEnv();
2418 exception=AcquireExceptionInfo();
2420 if (device->type == DS_DEVICE_NATIVE_CPU)
2423 MagickBooleanType flag=MagickTrue;
2424 SetMagickOpenCLEnvParamInternal(clEnv,
2425 MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED,
sizeof(MagickBooleanType),
2428 else if (device->type == DS_DEVICE_OPENCL_DEVICE)
2431 SetMagickOpenCLEnvParamInternal(clEnv,MAGICK_OPENCL_ENV_PARAM_DEVICE,
2432 sizeof(cl_device_id),&device->oclDeviceID,exception);
2435 ReturnStatus(DS_PERF_EVALUATOR_ERROR);
2438 clEnv->disableProgramCache = defaultCLEnv->disableProgramCache;
2440 status=InitOpenCLEnvInternal(clEnv,exception);
2441 oldClEnv=defaultCLEnv;
2445 if (status != MagickFalse)
2456 imageInfo=AcquireImageInfo();
2457 CloneString(&imageInfo->size,ACCELERATE_PERF_DIMEN);
2458 CopyMagickString(imageInfo->filename,
"xc:none",MaxTextExtent);
2459 inputImage=ReadImage(imageInfo,exception);
2460 if (inputImage == (
Image *) NULL)
2461 ReturnStatus(DS_PERF_EVALUATOR_ERROR);
2463 initAccelerateTimer(&timer);
2465 for (i=0; i<=NUM_ITER; i++)
2479 startAccelerateTimer(&timer);
2481 #ifdef MAGICKCORE_CLPERFMARKER 2482 clBeginPerfMarkerAMD(
"PerfEvaluatorRegion",
"");
2485 bluredImage=BlurImage(inputImage,10.0f,3.5f,exception);
2486 unsharpedImage=UnsharpMaskImage(bluredImage,2.0f,2.0f,50.0f,10.0f,
2488 resizedImage=ResizeImage(unsharpedImage,640,480,LanczosFilter,1.0,
2495 if (device->type != DS_DEVICE_NATIVE_CPU)
2497 events=GetOpenCLEvents(resizedImage,&event_count);
2498 if (event_count > 0)
2499 clEnv->library->clWaitForEvents(event_count,events);
2500 events=(cl_event *) RelinquishMagickMemory(events);
2503 #ifdef MAGICKCORE_CLPERFMARKER 2504 clEndPerfMarkerAMD();
2508 stopAccelerateTimer(&timer);
2511 DestroyImage(bluredImage);
2513 DestroyImage(unsharpedImage);
2515 DestroyImage(resizedImage);
2517 DestroyImage(inputImage);
2521 if (device->score == NULL)
2522 device->score= AcquireMagickMemory(
sizeof(AccelerateScoreType));
2524 if (status != MagickFalse)
2525 *(AccelerateScoreType*) device->score=readAccelerateTimer(&timer);
2527 *(AccelerateScoreType*) device->score=42;
2529 ReturnStatus(DS_SUCCESS);
2532 ds_status AccelerateScoreSerializer(ds_device* device,
void** serializedScore,
unsigned int* serializedScoreSize) {
2536 char* s = (
char*) AcquireQuantumMemory(256,
sizeof(
char));
2537 (void) FormatLocaleString(s,256,
"%.4f",*((AccelerateScoreType*)
2539 *serializedScore = (
void*)s;
2540 *serializedScoreSize = (
unsigned int) strlen(s);
2544 return DS_SCORE_SERIALIZER_ERROR;
2548 ds_status AccelerateScoreDeserializer(ds_device* device,
const unsigned char* serializedScore,
unsigned int serializedScoreSize) {
2551 char* s = (
char*) AcquireQuantumMemory(1,serializedScoreSize+1);
2552 memcpy(s, serializedScore, serializedScoreSize);
2553 s[serializedScoreSize] = (char)
'\0';
2554 device->score = AcquireMagickMemory(
sizeof(AccelerateScoreType));
2555 *((AccelerateScoreType*)device->score) = (AccelerateScoreType)
2556 strtod(s, (
char **) NULL);
2557 RelinquishMagickMemory(s);
2561 return DS_SCORE_DESERIALIZER_ERROR;
2565 ds_status AccelerateScoreRelease(
void* score) {
2567 RelinquishMagickMemory(score);
2572 ds_status canWriteProfileToFile(
const char *path)
2574 FILE* profileFile = fopen(path,
"ab");
2576 if (profileFile==NULL)
2577 return DS_FILE_ERROR;
2579 fclose(profileFile);
2584 #define IMAGEMAGICK_PROFILE_VERSION "ImageMagick Device Selection v0.9" 2585 #define IMAGEMAGICK_PROFILE_FILE "ImagemagickOpenCLDeviceProfile" 2588 MagickBooleanType mStatus = MagickFalse;
2590 ds_profile* profile;
2591 unsigned int numDeviceProfiled = 0;
2593 unsigned int bestDeviceIndex;
2594 AccelerateScoreType bestScore;
2595 char path[MaxTextExtent];
2596 MagickBooleanType flag;
2597 ds_evaluation_type profileType;
2599 LockDefaultOpenCLEnv();
2603 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2604 ,
sizeof(MagickBooleanType), &flag, exception);
2607 OpenCLLib=GetOpenCLLib();
2608 if (OpenCLLib==NULL)
2610 mStatus=InitOpenCLEnvInternal(clEnv, exception);
2614 clEnv->library=OpenCLLib;
2616 status = initDSProfile(&profile, IMAGEMAGICK_PROFILE_VERSION);
2617 if (status!=DS_SUCCESS) {
2618 (void) ThrowMagickException(exception, GetMagickModule(), ModuleFatalError,
"Error when initializing the profile",
"'%s'",
".");
2622 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s" 2623 ,GetOpenCLCachedFilesDirectory()
2624 ,DirectorySeparator,IMAGEMAGICK_PROFILE_FILE);
2626 if (canWriteProfileToFile(path) != DS_SUCCESS) {
2630 bestDeviceIndex = 0;
2631 for (i = 1; i < profile->numDevices; i++) {
2632 if ((profile->devices[i].type == DS_DEVICE_OPENCL_DEVICE) && (profile->devices[i].oclDeviceType == CL_DEVICE_TYPE_GPU)) {
2633 bestDeviceIndex = i;
2639 if (clEnv->regenerateProfile != MagickFalse) {
2640 profileType = DS_EVALUATE_ALL;
2643 readProfileFromFile(profile, AccelerateScoreDeserializer, path);
2644 profileType = DS_EVALUATE_NEW_ONLY;
2646 status = profileDevices(profile, profileType, AcceleratePerfEvaluator, NULL, &numDeviceProfiled);
2648 if (status!=DS_SUCCESS) {
2649 (void) ThrowMagickException(exception, GetMagickModule(), ModuleFatalError,
"Error when initializing the profile",
"'%s'",
".");
2652 if (numDeviceProfiled > 0) {
2653 status = writeProfileToFile(profile, AccelerateScoreSerializer, path);
2654 if (status!=DS_SUCCESS) {
2655 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"Error when saving the profile into a file",
"'%s'",
".");
2660 bestDeviceIndex = 0;
2661 bestScore = *(AccelerateScoreType*)profile->devices[bestDeviceIndex].score;
2662 for (i = 1; i < profile->numDevices; i++) {
2663 AccelerateScoreType score = *(AccelerateScoreType*)profile->devices[i].score;
2664 if (score < bestScore) {
2665 bestDeviceIndex = i;
2672 if (profile->devices[bestDeviceIndex].type == DS_DEVICE_NATIVE_CPU) {
2675 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2676 ,
sizeof(MagickBooleanType), &flag, exception);
2678 else if (profile->devices[bestDeviceIndex].type == DS_DEVICE_OPENCL_DEVICE) {
2681 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2682 ,
sizeof(MagickBooleanType), &flag, exception);
2683 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2684 ,
sizeof(cl_device_id), &profile->devices[bestDeviceIndex].oclDeviceID,exception);
2687 status = DS_PERF_EVALUATOR_ERROR;
2690 mStatus=InitOpenCLEnvInternal(clEnv, exception);
2692 status = releaseDSProfile(profile, AccelerateScoreRelease);
2693 if (status!=DS_SUCCESS) {
2694 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"Error when releasing the profile",
"'%s'",
".");
2699 UnlockDefaultOpenCLEnv();
2737 MagickExport MagickBooleanType InitImageMagickOpenCL(
2738 ImageMagickOpenCLMode mode,
void *userSelectedDevice,
void *selectedDevice,
2741 MagickBooleanType status = MagickFalse;
2743 MagickBooleanType flag;
2745 clEnv = GetDefaultOpenCLEnv();
2749 case MAGICK_OPENCL_OFF:
2751 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2752 ,
sizeof(MagickBooleanType), &flag, exception);
2753 status = InitOpenCLEnv(clEnv, exception);
2756 *(cl_device_id*)selectedDevice = NULL;
2759 case MAGICK_OPENCL_DEVICE_SELECT_USER:
2761 if (userSelectedDevice == NULL)
2765 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2766 ,
sizeof(MagickBooleanType), &flag, exception);
2768 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2769 ,
sizeof(cl_device_id), userSelectedDevice,exception);
2771 status = InitOpenCLEnv(clEnv, exception);
2772 if (selectedDevice) {
2773 GetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2774 ,
sizeof(cl_device_id), selectedDevice, exception);
2778 case MAGICK_OPENCL_DEVICE_SELECT_AUTO_CLEAR_CACHE:
2780 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED
2781 ,
sizeof(MagickBooleanType), &flag, exception);
2783 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE
2784 ,
sizeof(MagickBooleanType), &flag, exception);
2787 case MAGICK_OPENCL_DEVICE_SELECT_AUTO:
2790 cl_device_id d = NULL;
2792 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2793 ,
sizeof(MagickBooleanType), &flag, exception);
2794 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2795 ,
sizeof(cl_device_id), &d,exception);
2796 status = InitOpenCLEnv(clEnv, exception);
2797 if (selectedDevice) {
2798 GetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2799 ,
sizeof(cl_device_id), selectedDevice, exception);
2810 MagickBooleanType OpenCLThrowMagickException(
ExceptionInfo *exception,
2811 const char *module,
const char *
function,
const size_t line,
2812 const ExceptionType severity,
const char *tag,
const char *format,...) {
2818 status = MagickTrue;
2820 clEnv = GetDefaultOpenCLEnv();
2823 assert(exception->signature == MagickCoreSignature);
2826 cl_device_type dType;
2827 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_TYPE ,
sizeof(cl_device_type),&dType,NULL);
2828 if (dType == CL_DEVICE_TYPE_CPU) {
2829 char buffer[MaxTextExtent];
2830 clEnv->library->clGetPlatformInfo(clEnv->platform, CL_PLATFORM_NAME, MaxTextExtent, buffer, NULL);
2834 if (strncmp(buffer,
"Intel",5) == 0) {
2836 InitImageMagickOpenCL(MAGICK_OPENCL_OFF, NULL, NULL, exception);
2841 #ifdef OPENCLLOG_ENABLED 2845 va_start(operands,format);
2846 status=ThrowMagickExceptionList(exception,module,
function,line,severity,tag, format,operands);
2850 magick_unreferenced(module);
2851 magick_unreferenced(
function);
2852 magick_unreferenced(line);
2853 magick_unreferenced(tag);
2854 magick_unreferenced(format);
2860 char* openclCachedFilesDirectory;
2864 const char* GetOpenCLCachedFilesDirectory() {
2865 if (openclCachedFilesDirectory == NULL) {
2866 if (openclCachedFilesDirectoryLock == NULL)
2868 ActivateSemaphoreInfo(&openclCachedFilesDirectoryLock);
2870 LockSemaphoreInfo(openclCachedFilesDirectoryLock);
2871 if (openclCachedFilesDirectory == NULL) {
2872 char path[MaxTextExtent];
2875 struct stat attributes;
2876 MagickBooleanType status;
2877 int mkdirStatus = 0;
2881 home=GetEnvironmentValue(
"MAGICK_OPENCL_CACHE_DIR");
2882 if (home == (
char *) NULL)
2884 home=GetEnvironmentValue(
"XDG_CACHE_HOME");
2885 #if defined(MAGICKCORE_WINDOWS_SUPPORT) || defined(__MINGW32__) 2886 if (home == (
char *) NULL)
2887 home=GetEnvironmentValue(
"LOCALAPPDATA");
2888 if (home == (
char *) NULL)
2889 home=GetEnvironmentValue(
"APPDATA");
2890 if (home == (
char *) NULL)
2891 home=GetEnvironmentValue(
"USERPROFILE");
2895 if (home != (
char *) NULL)
2898 (void) FormatLocaleString(path,MaxTextExtent,
"%s",home);
2899 status=GetPathAttributes(path,&attributes);
2900 if (status == MagickFalse)
2903 #ifdef MAGICKCORE_WINDOWS_SUPPORT 2904 mkdirStatus = mkdir(path);
2906 mkdirStatus = mkdir(path, 0777);
2913 (void) FormatLocaleString(path,MaxTextExtent,
2914 "%s%sImageMagick",home,DirectorySeparator);
2916 status=GetPathAttributes(path,&attributes);
2917 if (status == MagickFalse)
2919 #ifdef MAGICKCORE_WINDOWS_SUPPORT 2920 mkdirStatus = mkdir(path);
2922 mkdirStatus = mkdir(path, 0777);
2929 temp = (
char*)AcquireCriticalMemory(strlen(path)+1);
2930 CopyMagickString(temp,path,strlen(path)+1);
2932 home=DestroyString(home);
2934 home=GetEnvironmentValue(
"HOME");
2935 if (home != (
char *) NULL)
2941 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s.cache",
2942 home,DirectorySeparator);
2943 status=GetPathAttributes(path,&attributes);
2944 if (status == MagickFalse)
2947 #ifdef MAGICKCORE_WINDOWS_SUPPORT 2948 mkdirStatus = mkdir(path);
2950 mkdirStatus = mkdir(path, 0777);
2957 (void) FormatLocaleString(path,MaxTextExtent,
2958 "%s%s.cache%sImageMagick",home,DirectorySeparator,
2959 DirectorySeparator);
2961 status=GetPathAttributes(path,&attributes);
2962 if (status == MagickFalse)
2964 #ifdef MAGICKCORE_WINDOWS_SUPPORT 2965 mkdirStatus = mkdir(path);
2967 mkdirStatus = mkdir(path, 0777);
2974 temp = (
char*)AcquireCriticalMemory(strlen(path)+1);
2975 CopyMagickString(temp,path,strlen(path)+1);
2977 home=DestroyString(home);
2980 openclCachedFilesDirectory = temp;
2982 UnlockSemaphoreInfo(openclCachedFilesDirectoryLock);
2984 return openclCachedFilesDirectory;
2989 void OpenCLLog(
const char* message) {
2991 #ifdef OPENCLLOG_ENABLED 2992 #define OPENCL_LOG_FILE "ImageMagickOpenCL.log" 2995 if (getenv(
"MAGICK_OCL_LOG"))
2998 char path[MaxTextExtent];
2999 unsigned long allocSize;
3003 clEnv = GetDefaultOpenCLEnv();
3006 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s" 3007 ,GetOpenCLCachedFilesDirectory()
3008 ,DirectorySeparator,OPENCL_LOG_FILE);
3011 log = fopen(path,
"ab");
3012 if (log == (FILE *) NULL)
3014 fwrite(message,
sizeof(
char), strlen(message), log);
3015 fwrite(
"\n",
sizeof(
char), 1, log);
3017 if (clEnv->OpenCLInitialized && !clEnv->OpenCLDisabled)
3019 allocSize = GetOpenCLDeviceMaxMemAllocSize(clEnv);
3020 fprintf(log,
"Devic Max Memory Alloc Size: %lu\n", allocSize);
3027 magick_unreferenced(message);
3031 MagickPrivate
void OpenCLTerminus()
3034 if (openclCachedFilesDirectory != (
char *) NULL)
3035 openclCachedFilesDirectory=DestroyString(openclCachedFilesDirectory);
3037 DestroySemaphoreInfo(&openclCachedFilesDirectoryLock);
3040 (void) RelinquishMagickOpenCLEnv(defaultCLEnv);
3044 DestroySemaphoreInfo(&defaultCLEnvLock);
3045 if (OpenCLLib != (MagickLibrary *)NULL)
3047 if (OpenCLLib->base != (
void *) NULL)
3048 (
void) lt_dlclose(OpenCLLib->base);
3049 OpenCLLib=(MagickLibrary *)RelinquishMagickMemory(OpenCLLib);
3052 DestroySemaphoreInfo(&OpenCLLibLock);
3058 MagickBooleanType OpenCLInitialized;
3069 MagickExport MagickBooleanType SetMagickOpenCLEnvParam(
3070 MagickCLEnv magick_unused(clEnv),MagickOpenCLEnvParam magick_unused(param),
3071 size_t magick_unused(dataSize),
void *magick_unused(data),
3074 magick_unreferenced(clEnv);
3075 magick_unreferenced(param);
3076 magick_unreferenced(dataSize);
3077 magick_unreferenced(data);
3078 magick_unreferenced(exception);
3079 return(MagickFalse);
3082 MagickExport MagickBooleanType GetMagickOpenCLEnvParam(
3083 MagickCLEnv magick_unused(clEnv),MagickOpenCLEnvParam magick_unused(param),
3084 size_t magick_unused(dataSize),
void *magick_unused(data),
3087 magick_unreferenced(clEnv);
3088 magick_unreferenced(param);
3089 magick_unreferenced(dataSize);
3090 magick_unreferenced(data);
3091 magick_unreferenced(exception);
3092 return(MagickFalse);
3095 MagickExport MagickBooleanType InitOpenCLEnv(
MagickCLEnv magick_unused(clEnv),
3098 magick_unreferenced(clEnv);
3099 magick_unreferenced(exception);
3100 return(MagickFalse);
3103 MagickExport MagickBooleanType InitImageMagickOpenCL(
3104 ImageMagickOpenCLMode magick_unused(mode),
3105 void *magick_unused(userSelectedDevice),
void *magick_unused(selectedDevice),
3108 magick_unreferenced(mode);
3109 magick_unreferenced(userSelectedDevice);
3110 magick_unreferenced(selectedDevice);
3111 magick_unreferenced(exception);
3112 return(MagickFalse);