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"
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;
110static const char *kernelNames[] = {
124 "LocalContrastBlurRow",
125 "LocalContrastBlurApplyColumn",
129 "RandomNumberGenerator",
132 "UnsharpMaskBlurColumn",
138 profileRecords[KERNEL_COUNT];
140typedef struct _AccelerateTimer {
146void 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;
158void 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;
172void resetAccelerateTimer(AccelerateTimer* timer) {
177void initAccelerateTimer(AccelerateTimer* timer) {
179 QueryPerformanceFrequency((LARGE_INTEGER*)&timer->_freq);
181 timer->_freq = (
long long)1.0E3;
183 resetAccelerateTimer(timer);
186double readAccelerateTimer(AccelerateTimer* timer) {
187 return (
double)timer->_clocks/(double)timer->_freq;
190MagickPrivate 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);
223void 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
284void *OsLibraryLoad(
const char *libraryName)
286#ifdef MAGICKCORE_WINDOWS_SUPPORT
287 return (
void *)LoadLibraryA(libraryName);
289 return (
void *)dlopen(libraryName, RTLD_NOW);
294void *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);
365MagickPrivate 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);
396MagickLibrary * OpenCLLib;
400static 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);
455MagickLibrary * 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);
540static void LockDefaultOpenCLEnv() {
541 if (defaultCLEnvLock == NULL)
543 ActivateSemaphoreInfo(&defaultCLEnvLock);
545 LockSemaphoreInfo(defaultCLEnvLock);
548static void UnlockDefaultOpenCLEnv() {
549 if (defaultCLEnvLock == NULL)
551 ActivateSemaphoreInfo(&defaultCLEnvLock);
554 UnlockSemaphoreInfo(defaultCLEnvLock);
584 LockDefaultOpenCLEnv();
585 oldEnv = defaultCLEnv;
586 defaultCLEnv = clEnv;
587 UnlockDefaultOpenCLEnv();
624static 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;
841static 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);
868static 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);
958static 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;
1021static 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();
1067extern 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;
1350static 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;
1388static 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();
1522MagickPrivate 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,
1576MagickPrivate 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;
1769typedef ds_status (*ds_score_release)(
void* score);
1771static 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);
1781static 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);
1799static 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);
1923typedef ds_status (*ds_perf_evaluator)(ds_device* device,
void* data);
1927 ,DS_EVALUATE_NEW_ONLY
1928} ds_evaluation_type;
1930static 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"
1991typedef ds_status (*ds_score_serializer)(ds_device* device,
void** serializedScore,
unsigned int* serializedScoreSize);
1992static 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);
2080static 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);
2124static 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;
2145typedef ds_status (*ds_score_deserializer)(ds_device* device,
const unsigned char* serializedScore,
unsigned int serializedScoreSize);
2146static 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);
2191DisableMSCWarning(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);
2364static 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) {
2383typedef double AccelerateScoreType;
2385static 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);
2461 initAccelerateTimer(&timer);
2463 for (i=0; i<=NUM_ITER; i++)
2477 startAccelerateTimer(&timer);
2479#ifdef MAGICKCORE_CLPERFMARKER
2480 clBeginPerfMarkerAMD(
"PerfEvaluatorRegion",
"");
2483 bluredImage=BlurImage(inputImage,10.0f,3.5f,exception);
2484 unsharpedImage=UnsharpMaskImage(bluredImage,2.0f,2.0f,50.0f,10.0f,
2486 resizedImage=ResizeImage(unsharpedImage,640,480,LanczosFilter,1.0,
2493 if (device->type != DS_DEVICE_NATIVE_CPU)
2495 events=GetOpenCLEvents(resizedImage,&event_count);
2496 if (event_count > 0)
2497 clEnv->library->clWaitForEvents(event_count,events);
2498 events=(cl_event *) RelinquishMagickMemory(events);
2501#ifdef MAGICKCORE_CLPERFMARKER
2502 clEndPerfMarkerAMD();
2506 stopAccelerateTimer(&timer);
2509 DestroyImage(bluredImage);
2511 DestroyImage(unsharpedImage);
2513 DestroyImage(resizedImage);
2515 DestroyImage(inputImage);
2519 if (device->score == NULL)
2520 device->score= AcquireMagickMemory(
sizeof(AccelerateScoreType));
2522 if (status != MagickFalse)
2523 *(AccelerateScoreType*) device->score=readAccelerateTimer(&timer);
2525 *(AccelerateScoreType*) device->score=42;
2527 ReturnStatus(DS_SUCCESS);
2530ds_status AccelerateScoreSerializer(ds_device* device,
void** serializedScore,
unsigned int* serializedScoreSize) {
2534 char* s = (
char*) AcquireQuantumMemory(256,
sizeof(
char));
2535 (void) FormatLocaleString(s,256,
"%.4f",*((AccelerateScoreType*)
2537 *serializedScore = (
void*)s;
2538 *serializedScoreSize = (
unsigned int) strlen(s);
2542 return DS_SCORE_SERIALIZER_ERROR;
2546ds_status AccelerateScoreDeserializer(ds_device* device,
const unsigned char* serializedScore,
unsigned int serializedScoreSize) {
2549 char* s = (
char*) AcquireQuantumMemory(1,serializedScoreSize+1);
2550 memcpy(s, serializedScore, serializedScoreSize);
2551 s[serializedScoreSize] = (char)
'\0';
2552 device->score = AcquireMagickMemory(
sizeof(AccelerateScoreType));
2553 *((AccelerateScoreType*)device->score) = (AccelerateScoreType)
2554 strtod(s, (
char **) NULL);
2555 RelinquishMagickMemory(s);
2559 return DS_SCORE_DESERIALIZER_ERROR;
2563ds_status AccelerateScoreRelease(
void* score) {
2565 RelinquishMagickMemory(score);
2570ds_status canWriteProfileToFile(
const char *path)
2572 FILE* profileFile = fopen(path,
"ab");
2574 if (profileFile==NULL)
2575 return DS_FILE_ERROR;
2577 fclose(profileFile);
2582#define IMAGEMAGICK_PROFILE_VERSION "ImageMagick Device Selection v0.9"
2583#define IMAGEMAGICK_PROFILE_FILE "ImagemagickOpenCLDeviceProfile"
2586 MagickBooleanType mStatus = MagickFalse;
2588 ds_profile* profile;
2589 unsigned int numDeviceProfiled = 0;
2591 unsigned int bestDeviceIndex;
2592 AccelerateScoreType bestScore;
2593 char path[MaxTextExtent];
2594 MagickBooleanType flag;
2595 ds_evaluation_type profileType;
2597 LockDefaultOpenCLEnv();
2601 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2602 ,
sizeof(MagickBooleanType), &flag, exception);
2605 OpenCLLib=GetOpenCLLib();
2606 if (OpenCLLib==NULL)
2608 mStatus=InitOpenCLEnvInternal(clEnv, exception);
2612 clEnv->library=OpenCLLib;
2614 status = initDSProfile(&profile, IMAGEMAGICK_PROFILE_VERSION);
2615 if (status!=DS_SUCCESS) {
2616 (void) ThrowMagickException(exception, GetMagickModule(), ModuleFatalError,
"Error when initializing the profile",
"'%s'",
".");
2620 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
2621 ,GetOpenCLCachedFilesDirectory()
2622 ,DirectorySeparator,IMAGEMAGICK_PROFILE_FILE);
2624 if (canWriteProfileToFile(path) != DS_SUCCESS) {
2628 bestDeviceIndex = 0;
2629 for (i = 1; i < profile->numDevices; i++) {
2630 if ((profile->devices[i].type == DS_DEVICE_OPENCL_DEVICE) && (profile->devices[i].oclDeviceType == CL_DEVICE_TYPE_GPU)) {
2631 bestDeviceIndex = i;
2637 if (clEnv->regenerateProfile != MagickFalse) {
2638 profileType = DS_EVALUATE_ALL;
2641 readProfileFromFile(profile, AccelerateScoreDeserializer, path);
2642 profileType = DS_EVALUATE_NEW_ONLY;
2644 status = profileDevices(profile, profileType, AcceleratePerfEvaluator, NULL, &numDeviceProfiled);
2646 if (status!=DS_SUCCESS) {
2647 (void) ThrowMagickException(exception, GetMagickModule(), ModuleFatalError,
"Error when initializing the profile",
"'%s'",
".");
2650 if (numDeviceProfiled > 0) {
2651 status = writeProfileToFile(profile, AccelerateScoreSerializer, path);
2652 if (status!=DS_SUCCESS) {
2653 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"Error when saving the profile into a file",
"'%s'",
".");
2658 bestDeviceIndex = 0;
2659 bestScore = *(AccelerateScoreType*)profile->devices[bestDeviceIndex].score;
2660 for (i = 1; i < profile->numDevices; i++) {
2661 AccelerateScoreType score = *(AccelerateScoreType*)profile->devices[i].score;
2662 if (score < bestScore) {
2663 bestDeviceIndex = i;
2670 if (profile->devices[bestDeviceIndex].type == DS_DEVICE_NATIVE_CPU) {
2673 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2674 ,
sizeof(MagickBooleanType), &flag, exception);
2676 else if (profile->devices[bestDeviceIndex].type == DS_DEVICE_OPENCL_DEVICE) {
2679 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2680 ,
sizeof(MagickBooleanType), &flag, exception);
2681 SetMagickOpenCLEnvParamInternal(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2682 ,
sizeof(cl_device_id), &profile->devices[bestDeviceIndex].oclDeviceID,exception);
2685 status = DS_PERF_EVALUATOR_ERROR;
2688 mStatus=InitOpenCLEnvInternal(clEnv, exception);
2690 status = releaseDSProfile(profile, AccelerateScoreRelease);
2691 if (status!=DS_SUCCESS) {
2692 (void) ThrowMagickException(exception, GetMagickModule(), ModuleWarning,
"Error when releasing the profile",
"'%s'",
".");
2697 UnlockDefaultOpenCLEnv();
2735MagickExport MagickBooleanType InitImageMagickOpenCL(
2736 ImageMagickOpenCLMode mode,
void *userSelectedDevice,
void *selectedDevice,
2739 MagickBooleanType status = MagickFalse;
2741 MagickBooleanType flag;
2743 clEnv = GetDefaultOpenCLEnv();
2747 case MAGICK_OPENCL_OFF:
2749 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2750 ,
sizeof(MagickBooleanType), &flag, exception);
2751 status = InitOpenCLEnv(clEnv, exception);
2754 *(cl_device_id*)selectedDevice = NULL;
2757 case MAGICK_OPENCL_DEVICE_SELECT_USER:
2759 if (userSelectedDevice == NULL)
2763 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2764 ,
sizeof(MagickBooleanType), &flag, exception);
2766 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2767 ,
sizeof(cl_device_id), userSelectedDevice,exception);
2769 status = InitOpenCLEnv(clEnv, exception);
2770 if (selectedDevice) {
2771 GetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2772 ,
sizeof(cl_device_id), selectedDevice, exception);
2776 case MAGICK_OPENCL_DEVICE_SELECT_AUTO_CLEAR_CACHE:
2778 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_PROGRAM_CACHE_DISABLED
2779 ,
sizeof(MagickBooleanType), &flag, exception);
2781 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_REGENERATE_PROFILE
2782 ,
sizeof(MagickBooleanType), &flag, exception);
2785 case MAGICK_OPENCL_DEVICE_SELECT_AUTO:
2788 cl_device_id d = NULL;
2790 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_OPENCL_DISABLED
2791 ,
sizeof(MagickBooleanType), &flag, exception);
2792 SetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2793 ,
sizeof(cl_device_id), &d,exception);
2794 status = InitOpenCLEnv(clEnv, exception);
2795 if (selectedDevice) {
2796 GetMagickOpenCLEnvParam(clEnv, MAGICK_OPENCL_ENV_PARAM_DEVICE
2797 ,
sizeof(cl_device_id), selectedDevice, exception);
2808MagickBooleanType OpenCLThrowMagickException(
ExceptionInfo *exception,
2809 const char *module,
const char *function,
const size_t line,
2810 const ExceptionType severity,
const char *tag,
const char *format,...) {
2816 status = MagickTrue;
2818 clEnv = GetDefaultOpenCLEnv();
2821 assert(exception->signature == MagickCoreSignature);
2824 cl_device_type dType;
2825 clEnv->library->clGetDeviceInfo(clEnv->device,CL_DEVICE_TYPE ,
sizeof(cl_device_type),&dType,NULL);
2826 if (dType == CL_DEVICE_TYPE_CPU) {
2827 char buffer[MaxTextExtent];
2828 clEnv->library->clGetPlatformInfo(clEnv->platform, CL_PLATFORM_NAME, MaxTextExtent, buffer, NULL);
2832 if (strncmp(buffer,
"Intel",5) == 0) {
2834 InitImageMagickOpenCL(MAGICK_OPENCL_OFF, NULL, NULL, exception);
2839#ifdef OPENCLLOG_ENABLED
2843 va_start(operands,format);
2844 status=ThrowMagickExceptionList(exception,module,function,line,severity,tag, format,operands);
2848 magick_unreferenced(module);
2849 magick_unreferenced(function);
2850 magick_unreferenced(line);
2851 magick_unreferenced(tag);
2852 magick_unreferenced(format);
2858char* openclCachedFilesDirectory;
2862const char* GetOpenCLCachedFilesDirectory() {
2863 if (openclCachedFilesDirectory == NULL) {
2864 if (openclCachedFilesDirectoryLock == NULL)
2866 ActivateSemaphoreInfo(&openclCachedFilesDirectoryLock);
2868 LockSemaphoreInfo(openclCachedFilesDirectoryLock);
2869 if (openclCachedFilesDirectory == NULL) {
2870 char path[MaxTextExtent];
2873 struct stat attributes;
2874 MagickBooleanType status;
2875 int mkdirStatus = 0;
2879 home=GetEnvironmentValue(
"MAGICK_OPENCL_CACHE_DIR");
2880 if (home == (
char *) NULL)
2882 home=GetEnvironmentValue(
"XDG_CACHE_HOME");
2883#if defined(MAGICKCORE_WINDOWS_SUPPORT) || defined(__MINGW32__)
2884 if (home == (
char *) NULL)
2885 home=GetEnvironmentValue(
"LOCALAPPDATA");
2886 if (home == (
char *) NULL)
2887 home=GetEnvironmentValue(
"APPDATA");
2888 if (home == (
char *) NULL)
2889 home=GetEnvironmentValue(
"USERPROFILE");
2893 if (home != (
char *) NULL)
2896 (void) FormatLocaleString(path,MaxTextExtent,
"%s",home);
2897 status=GetPathAttributes(path,&attributes);
2898 if (status == MagickFalse)
2901#ifdef MAGICKCORE_WINDOWS_SUPPORT
2902 mkdirStatus = mkdir(path);
2904 mkdirStatus = mkdir(path, 0777);
2911 (void) FormatLocaleString(path,MaxTextExtent,
2912 "%s%sImageMagick",home,DirectorySeparator);
2914 status=GetPathAttributes(path,&attributes);
2915 if (status == MagickFalse)
2917#ifdef MAGICKCORE_WINDOWS_SUPPORT
2918 mkdirStatus = mkdir(path);
2920 mkdirStatus = mkdir(path, 0777);
2927 temp = (
char*)AcquireCriticalMemory(strlen(path)+1);
2928 CopyMagickString(temp,path,strlen(path)+1);
2930 home=DestroyString(home);
2932 home=GetEnvironmentValue(
"HOME");
2933 if (home != (
char *) NULL)
2939 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s.cache",
2940 home,DirectorySeparator);
2941 status=GetPathAttributes(path,&attributes);
2942 if (status == MagickFalse)
2945#ifdef MAGICKCORE_WINDOWS_SUPPORT
2946 mkdirStatus = mkdir(path);
2948 mkdirStatus = mkdir(path, 0777);
2955 (void) FormatLocaleString(path,MaxTextExtent,
2956 "%s%s.cache%sImageMagick",home,DirectorySeparator,
2957 DirectorySeparator);
2959 status=GetPathAttributes(path,&attributes);
2960 if (status == MagickFalse)
2962#ifdef MAGICKCORE_WINDOWS_SUPPORT
2963 mkdirStatus = mkdir(path);
2965 mkdirStatus = mkdir(path, 0777);
2972 temp = (
char*)AcquireCriticalMemory(strlen(path)+1);
2973 CopyMagickString(temp,path,strlen(path)+1);
2975 home=DestroyString(home);
2978 openclCachedFilesDirectory = temp;
2980 UnlockSemaphoreInfo(openclCachedFilesDirectoryLock);
2982 return openclCachedFilesDirectory;
2987void OpenCLLog(
const char* message) {
2989#ifdef OPENCLLOG_ENABLED
2990#define OPENCL_LOG_FILE "ImageMagickOpenCL.log"
2993 if (getenv(
"MAGICK_OCL_LOG"))
2996 char path[MaxTextExtent];
2997 unsigned long allocSize;
3001 clEnv = GetDefaultOpenCLEnv();
3004 (void) FormatLocaleString(path,MaxTextExtent,
"%s%s%s"
3005 ,GetOpenCLCachedFilesDirectory()
3006 ,DirectorySeparator,OPENCL_LOG_FILE);
3009 log = fopen(path,
"ab");
3010 if (log == (FILE *) NULL)
3012 fwrite(message,
sizeof(
char), strlen(message), log);
3013 fwrite(
"\n",
sizeof(
char), 1, log);
3015 if (clEnv->OpenCLInitialized && !clEnv->OpenCLDisabled)
3017 allocSize = GetOpenCLDeviceMaxMemAllocSize(clEnv);
3018 fprintf(log,
"Devic Max Memory Alloc Size: %lu\n", allocSize);
3025 magick_unreferenced(message);
3029MagickPrivate
void OpenCLTerminus()
3032 if (openclCachedFilesDirectory != (
char *) NULL)
3033 openclCachedFilesDirectory=DestroyString(openclCachedFilesDirectory);
3035 DestroySemaphoreInfo(&openclCachedFilesDirectoryLock);
3038 (void) RelinquishMagickOpenCLEnv(defaultCLEnv);
3042 DestroySemaphoreInfo(&defaultCLEnvLock);
3043 if (OpenCLLib != (MagickLibrary *)NULL)
3045 if (OpenCLLib->base != (
void *) NULL)
3046 (
void) lt_dlclose(OpenCLLib->base);
3047 OpenCLLib=(MagickLibrary *)RelinquishMagickMemory(OpenCLLib);
3050 DestroySemaphoreInfo(&OpenCLLibLock);
3056 MagickBooleanType OpenCLInitialized;
3067MagickExport MagickBooleanType SetMagickOpenCLEnvParam(
3068 MagickCLEnv magick_unused(clEnv),MagickOpenCLEnvParam magick_unused(param),
3069 size_t magick_unused(dataSize),
void *magick_unused(data),
3072 magick_unreferenced(clEnv);
3073 magick_unreferenced(param);
3074 magick_unreferenced(dataSize);
3075 magick_unreferenced(data);
3076 magick_unreferenced(exception);
3077 return(MagickFalse);
3080MagickExport MagickBooleanType GetMagickOpenCLEnvParam(
3081 MagickCLEnv magick_unused(clEnv),MagickOpenCLEnvParam magick_unused(param),
3082 size_t magick_unused(dataSize),
void *magick_unused(data),
3085 magick_unreferenced(clEnv);
3086 magick_unreferenced(param);
3087 magick_unreferenced(dataSize);
3088 magick_unreferenced(data);
3089 magick_unreferenced(exception);
3090 return(MagickFalse);
3093MagickExport MagickBooleanType InitOpenCLEnv(
MagickCLEnv magick_unused(clEnv),
3096 magick_unreferenced(clEnv);
3097 magick_unreferenced(exception);
3098 return(MagickFalse);
3101MagickExport MagickBooleanType InitImageMagickOpenCL(
3102 ImageMagickOpenCLMode magick_unused(mode),
3103 void *magick_unused(userSelectedDevice),
void *magick_unused(selectedDevice),
3106 magick_unreferenced(mode);
3107 magick_unreferenced(userSelectedDevice);
3108 magick_unreferenced(selectedDevice);
3109 magick_unreferenced(exception);
3110 return(MagickFalse);