← Back to branch summary

~ubuntu-branches/ubuntu/saucy/darktable/saucy

« back to all changes in this revision

Viewing changes to src/common/opencl.c

  • Committer: Bazaar Package Importer
  • Author(s): David Bremner
  • Date: 2011-07-12 09:36:46 UTC
  • mfrom: (1.1.1 upstream)
  • Revision ID: james.westby@ubuntu.com-20110712093646-yp9dbxan44dmw15h
Tags: 0.9-1
* New upstream release.
* Remove all patches now upstream; only patch for
  -Wno-error=unused-but-set-variable remains.
* Bump Standards-Version to 3.9.2 (no changes)

Show diffs side-by-side

added added

removed removed

Lines of Context:
src/common/opencl.c
15
15
    You should have received a copy of the GNU General Public License
16
16
    along with darktable.  If not, see <http://www.gnu.org/licenses/>.
17
17
*/
18
 
#include <strings.h>
 
18
 
 
19
#ifdef HAVE_OPENCL
 
20
 
 
21
#include <string.h>
19
22
#include <stdio.h>
20
23
 
21
24
#include "common/darktable.h"
22
25
#include "common/opencl.h"
 
26
#include "common/dlopencl.h"
 
27
#include "control/conf.h"
23
28
 
24
29
void dt_opencl_init(dt_opencl_t *cl, const int argc, char *argv[])
25
30
{
26
31
  dt_pthread_mutex_init(&cl->lock, NULL);
27
32
  cl->inited = 0;
 
33
  cl->enabled = 0;
 
34
  cl->dlocl = NULL;
 
35
  int exclude_opencl = 0;
 
36
 
 
37
  // preliminary disable opencl in prefs. We will re-set it to previous state later if possible
 
38
  // Will remain disabled if initialization fails.
 
39
  const int prefs = dt_conf_get_bool("opencl");
 
40
  dt_conf_set_bool("opencl", FALSE);
 
41
 
28
42
  for(int k=0; k<argc; k++) if(!strcmp(argv[k], "--disable-opencl"))
29
43
    {
30
 
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] not using opencl by explicit request\n");
 
44
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] do not try to find and use an opencl runtime library due to explicit user request\n");
 
45
      exclude_opencl = 1;
 
46
    }
 
47
 
 
48
  if(exclude_opencl) return;
 
49
 
 
50
 
 
51
  // look for explicit definition of opencl_runtime library in preferences
 
52
  const char *library = dt_conf_get_string("opencl_runtime");
 
53
  dt_print(DT_DEBUG_OPENCL, "[opencl_init] trying to load opencl library: %s\n", library && strlen(library) != 0 ? library : "<system default>");
 
54
 
 
55
  // dynamically load opencl runtime
 
56
  if(!dt_dlopencl_init(library, &cl->dlocl))
 
57
    {
 
58
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] no working opencl library found. Continue with opencl disabled\n");
31
59
      return;
32
60
    }
 
61
    else
 
62
    {
 
63
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] opencl library %s found on your system and loaded\n", cl->dlocl->library);
 
64
    }
 
65
 
33
66
  cl_int err;
34
67
  cl_platform_id all_platforms[5];
35
68
  cl_platform_id platform = NULL;
36
69
  cl_uint num_platforms = 5;
37
 
  err = clGetPlatformIDs (5, all_platforms, &num_platforms);
 
70
  err = (cl->dlocl->symbols->dt_clGetPlatformIDs) (5, all_platforms, &num_platforms);
38
71
  if(err != CL_SUCCESS)
39
72
  {
40
73
    dt_print(DT_DEBUG_OPENCL, "[opencl_init] could not get platforms: %d\n", err);
45
78
  // get the number of GPU devices available to the platform
46
79
  // the other common option is CL_DEVICE_TYPE_GPU/CPU (but the latter doesn't work with the nvidia drivers)
47
80
  cl_uint num_devices = 0;
48
 
  err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &num_devices);
 
81
  err = (cl->dlocl->symbols->dt_clGetDeviceIDs)(platform, CL_DEVICE_TYPE_ALL, 0, NULL, &num_devices);
49
82
  if(err != CL_SUCCESS)
50
83
  {
51
84
    dt_print(DT_DEBUG_OPENCL, "[opencl_init] could not get device id size: %d\n", err);
53
86
  }
54
87
 
55
88
  // create the device list
56
 
  cl->num_devs = num_devices;
57
89
  cl->dev = (dt_opencl_device_t *)malloc(sizeof(dt_opencl_device_t)*num_devices);
58
90
  cl_device_id *devices = (cl_device_id *)malloc(sizeof(cl_device_id)*num_devices);
59
 
  err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, num_devices, devices, NULL);
 
91
  err = (cl->dlocl->symbols->dt_clGetDeviceIDs)(platform, CL_DEVICE_TYPE_ALL, num_devices, devices, NULL);
60
92
  if(err != CL_SUCCESS)
61
93
  {
62
94
    dt_print(DT_DEBUG_OPENCL, "[opencl_init] could not get devices list: %d\n", err);
63
95
    return;
64
96
  }
65
97
  dt_print(DT_DEBUG_OPENCL, "[opencl_init] found %d devices\n", num_devices);
 
98
  int dev = 0;
66
99
  for(int k=0; k<num_devices; k++)
67
100
  {
68
 
    memset(cl->dev[k].program_used, 0x0, sizeof(int)*DT_OPENCL_MAX_PROGRAMS);
69
 
    memset(cl->dev[k].kernel_used,  0x0, sizeof(int)*DT_OPENCL_MAX_KERNELS);
70
 
    cl->dev[k].devid = devices[k];
 
101
    memset(cl->dev[dev].program_used, 0x0, sizeof(int)*DT_OPENCL_MAX_PROGRAMS);
 
102
    memset(cl->dev[dev].kernel_used,  0x0, sizeof(int)*DT_OPENCL_MAX_KERNELS);
 
103
    cl_device_id devid = cl->dev[dev].devid = devices[k];
71
104
 
72
105
    char infostr[1024];
73
106
    size_t infoint;
74
107
    size_t infointtab[1024];
75
108
    cl_bool image_support = 0;
76
109
    cl_ulong max_global_mem = 0;
77
 
    size_t image_width = 0, image_height = 0;
 
110
    //size_t image_width = 0, image_height = 0;
78
111
 
79
112
    // test 1GB mem and image support:
80
 
    clGetDeviceInfo(cl->dev[k].devid, CL_DEVICE_NAME, sizeof(infostr), &infostr, NULL);
81
 
    clGetDeviceInfo(cl->dev[k].devid, CL_DEVICE_IMAGE_SUPPORT, sizeof(cl_bool), &image_support, NULL);
82
 
    clGetDeviceInfo(cl->dev[k].devid, CL_DEVICE_IMAGE2D_MAX_HEIGHT, sizeof(size_t), &image_height, NULL);
83
 
    clGetDeviceInfo(cl->dev[k].devid, CL_DEVICE_IMAGE2D_MAX_WIDTH,  sizeof(size_t), &image_width,  NULL);
 
113
    (cl->dlocl->symbols->dt_clGetDeviceInfo)(devid, CL_DEVICE_NAME, sizeof(infostr), &infostr, NULL);
 
114
    (cl->dlocl->symbols->dt_clGetDeviceInfo)(devid, CL_DEVICE_IMAGE_SUPPORT, sizeof(cl_bool), &image_support, NULL);
 
115
    (cl->dlocl->symbols->dt_clGetDeviceInfo)(devid, CL_DEVICE_IMAGE2D_MAX_HEIGHT, sizeof(size_t), &(cl->dev[dev].max_image_height), NULL);
 
116
    (cl->dlocl->symbols->dt_clGetDeviceInfo)(devid, CL_DEVICE_IMAGE2D_MAX_WIDTH,  sizeof(size_t), &(cl->dev[dev].max_image_width),  NULL);
84
117
    if(!image_support)
85
118
    {
86
 
      cl->num_devs = --num_devices;
87
119
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] discarding device %d `%s' due to missing image support.\n", k, infostr);
88
120
      continue;
89
121
    }
90
 
    if(image_height < 8192 || image_width < 8192)
91
 
    {
92
 
      fprintf(stderr, "[opencl_init] WARNING: your card only supports image sizes of %zd x %zd\n", image_width, image_height);
93
 
      fprintf(stderr, "[opencl_init] WARNING: expect random crashes, especially with images larger than that.\n");
94
 
    }
 
122
    //if(image_height < 8192 || image_width < 8192)
 
123
    //{
 
124
    //  fprintf(stderr, "[opencl_init] WARNING: your card only supports image sizes of %zd x %zd\n", image_width, image_height);
 
125
    //  fprintf(stderr, "[opencl_init] WARNING: expect random crashes, especially with images larger than that.\n");
 
126
    //}
95
127
 
96
 
    clGetDeviceInfo(cl->dev[k].devid, CL_DEVICE_GLOBAL_MEM_SIZE, sizeof(cl_ulong), &max_global_mem, NULL);
 
128
    (cl->dlocl->symbols->dt_clGetDeviceInfo)(devid, CL_DEVICE_GLOBAL_MEM_SIZE, sizeof(cl_ulong), &max_global_mem, NULL);
97
129
    if(max_global_mem < 1000000000ul)
98
130
    {
99
 
      cl->num_devs = --num_devices;
100
 
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] discarding device %d `%s' due to insufficient global memory.\n", k, infostr);
 
131
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] discarding device %d `%s' due to insufficient global memory (%luMB).\n", k, infostr, max_global_mem/1024/1024);
101
132
      continue;
102
133
    }
103
134
    if(darktable.unmuted & DT_DEBUG_OPENCL)
104
135
    {
105
136
      printf("[opencl_init] device %d: %s \n", k, infostr);
106
 
      clGetDeviceInfo(cl->dev[k].devid, CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(infoint), &infoint, NULL);
 
137
      (cl->dlocl->symbols->dt_clGetDeviceInfo)(devid, CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(infoint), &infoint, NULL);
107
138
      printf("     MAX_WORK_GROUP_SIZE:      %zd\n", infoint);
108
 
      clGetDeviceInfo(cl->dev[k].devid, CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS, sizeof(infoint), &infoint, NULL);
 
139
      (cl->dlocl->symbols->dt_clGetDeviceInfo)(devid, CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS, sizeof(infoint), &infoint, NULL);
109
140
      printf("     MAX_WORK_ITEM_DIMENSIONS: %zd\n", infoint);
110
141
      printf("     MAX_WORK_ITEM_SIZES:      [ ");
111
 
      clGetDeviceInfo(cl->dev[k].devid, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(infointtab), infointtab, NULL);
 
142
      (cl->dlocl->symbols->dt_clGetDeviceInfo)(devid, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(infointtab), infointtab, NULL);
112
143
      for (int i=0; i<infoint; i++) printf("%zd ", infointtab[i]);
113
144
      printf("]\n");
114
145
    }
115
 
    dt_pthread_mutex_init(&cl->dev[k].lock, NULL);
 
146
    dt_pthread_mutex_init(&cl->dev[dev].lock, NULL);
116
147
 
117
 
    cl->dev[k].context = clCreateContext(0, 1, &cl->dev[k].devid, NULL, NULL, &err);
 
148
    cl->dev[dev].context = (cl->dlocl->symbols->dt_clCreateContext)(0, 1, &devid, NULL, NULL, &err);
118
149
    if(err != CL_SUCCESS)
119
150
    {
120
151
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] could not create context for device %d: %d\n", k, err);
121
152
      return;
122
153
    }
123
154
    // create a command queue for first device the context reported
124
 
    cl->dev[k].cmd_queue = clCreateCommandQueue(cl->dev[k].context, cl->dev[k].devid, 0, &err);
 
155
    cl->dev[dev].cmd_queue = (cl->dlocl->symbols->dt_clCreateCommandQueue)(cl->dev[dev].context, devid, 0, &err);
125
156
    if(err != CL_SUCCESS)
126
157
    {
127
158
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] could not create command queue for device %d: %d\n", k, err);
151
182
        dt_print(DT_DEBUG_OPENCL, "[opencl_init] compiling program `%s' ..\n", programname);
152
183
        const int prog = dt_opencl_load_program(cl, k, filename);
153
184
        if(dt_opencl_build_program(cl, k, prog))
 
185
        {
154
186
          dt_print(DT_DEBUG_OPENCL, "[opencl_init] failed to compile program `%s'!\n", programname);
 
187
          return;
 
188
        }
155
189
      }
156
190
      fclose(f);
157
191
    }
160
194
      dt_print(DT_DEBUG_OPENCL, "[opencl_init] could not open `%s'!\n", filename);
161
195
      return;
162
196
    }
 
197
    ++dev;
163
198
  }
164
199
  free(devices);
165
 
  if(num_devices > 0)
 
200
  if(dev > 0)
166
201
  {
167
202
    dt_print(DT_DEBUG_OPENCL, "[opencl_init] successfully initialized.\n");
 
203
    dt_print(DT_DEBUG_OPENCL, "[opencl_init] initial status of opencl enabled flag is %s.\n", prefs ? "ON" : "OFF");
 
204
    cl->num_devs = dev;
168
205
    cl->inited = 1;
 
206
    cl->enabled = prefs;
 
207
    // set preferences to saved state
 
208
    dt_conf_set_bool("opencl", prefs);
169
209
  }
170
210
  else
171
211
  {
179
219
  if(cl->inited) for(int i=0; i<cl->num_devs; i++)
180
220
    {
181
221
      dt_pthread_mutex_destroy(&cl->dev[i].lock);
182
 
      for(int k=0; k<DT_OPENCL_MAX_KERNELS; k++) if(cl->dev[i].kernel_used [k]) clReleaseKernel (cl->dev[i].kernel [k]);
183
 
      for(int k=0; k<DT_OPENCL_MAX_PROGRAMS; k++) if(cl->dev[i].program_used[k]) clReleaseProgram(cl->dev[i].program[k]);
184
 
      clReleaseCommandQueue(cl->dev[i].cmd_queue);
185
 
      clReleaseContext(cl->dev[i].context);
 
222
      for(int k=0; k<DT_OPENCL_MAX_KERNELS; k++) if(cl->dev[i].kernel_used [k]) (cl->dlocl->symbols->dt_clReleaseKernel) (cl->dev[i].kernel [k]);
 
223
      for(int k=0; k<DT_OPENCL_MAX_PROGRAMS; k++) if(cl->dev[i].program_used[k]) (cl->dlocl->symbols->dt_clReleaseProgram)(cl->dev[i].program[k]);
 
224
      (cl->dlocl->symbols->dt_clReleaseCommandQueue)(cl->dev[i].cmd_queue);
 
225
      (cl->dlocl->symbols->dt_clReleaseContext)(cl->dev[i].context);
186
226
    }
 
227
 
 
228
  if(cl->dlocl) {
 
229
    free(cl->dlocl->symbols);
 
230
    free(cl->dlocl);
 
231
  }
 
232
 
187
233
  dt_pthread_mutex_destroy(&cl->lock);
188
234
}
189
235
 
 
236
void dt_opencl_finish(cl_command_queue q)
 
237
{
 
238
  if(!darktable.opencl->inited) return;
 
239
  (darktable.opencl->dlocl->symbols->dt_clFinish)(q);
 
240
}
 
241
 
190
242
int dt_opencl_lock_device(dt_opencl_t *cl, const int _dev)
191
243
{
192
244
  if(!cl->inited) return -1;
254
306
  for(; k<DT_OPENCL_MAX_PROGRAMS; k++) if(!cl->dev[dev].program_used[k])
255
307
    {
256
308
      cl->dev[dev].program_used[k] = 1;
257
 
      cl->dev[dev].program[k] = clCreateProgramWithSource(cl->dev[dev].context, lines, sptr, lengths, &err);
 
309
      cl->dev[dev].program[k] = (cl->dlocl->symbols->dt_clCreateProgramWithSource)(cl->dev[dev].context, lines, sptr, lengths, &err);
258
310
      if(err != CL_SUCCESS)
259
311
      {
260
312
        dt_print(DT_DEBUG_OPENCL, "[opencl_load_program] could not create program from file `%s'! (%d)\n", filename, err);
280
332
  if(prog < 0 || prog >= DT_OPENCL_MAX_PROGRAMS) return -1;
281
333
  cl_program program = cl->dev[dev].program[prog];
282
334
  cl_int err;
283
 
  err = clBuildProgram(program, 1, &cl->dev[dev].devid, "-cl-fast-relaxed-math -cl-strict-aliasing", 0, 0);
 
335
  err = (cl->dlocl->symbols->dt_clBuildProgram)(program, 1, &cl->dev[dev].devid, "-cl-fast-relaxed-math -cl-strict-aliasing", 0, 0);
284
336
  if(err != CL_SUCCESS)
285
337
  {
286
338
    dt_print(DT_DEBUG_OPENCL, "[opencl_build_program] could not build program: %d\n", err);
287
339
    cl_build_status build_status;
288
 
    clGetProgramBuildInfo(program, cl->dev[dev].devid, CL_PROGRAM_BUILD_STATUS, sizeof(cl_build_status), &build_status, NULL);
 
340
    (cl->dlocl->symbols->dt_clGetProgramBuildInfo)(program, cl->dev[dev].devid, CL_PROGRAM_BUILD_STATUS, sizeof(cl_build_status), &build_status, NULL);
289
341
    if (build_status != CL_SUCCESS)
290
342
    {
291
343
      char *build_log;
292
344
      size_t ret_val_size;
293
 
      clGetProgramBuildInfo(program, cl->dev[dev].devid, CL_PROGRAM_BUILD_LOG, 0, NULL, &ret_val_size);
 
345
      (cl->dlocl->symbols->dt_clGetProgramBuildInfo)(program, cl->dev[dev].devid, CL_PROGRAM_BUILD_LOG, 0, NULL, &ret_val_size);
294
346
      build_log = (char *)malloc(sizeof(char)*(ret_val_size+1));
295
 
      clGetProgramBuildInfo(program, cl->dev[dev].devid, CL_PROGRAM_BUILD_LOG, ret_val_size, build_log, NULL);
 
347
      (cl->dlocl->symbols->dt_clGetProgramBuildInfo)(program, cl->dev[dev].devid, CL_PROGRAM_BUILD_LOG, ret_val_size, build_log, NULL);
296
348
 
297
349
      build_log[ret_val_size] = '\0';
298
350
 
302
354
      free(build_log);
303
355
    }
304
356
  }
305
 
  dt_print(DT_DEBUG_OPENCL, "[opencl_build_program] successfully built program\n");
 
357
  else
 
358
  {
 
359
    dt_print(DT_DEBUG_OPENCL, "[opencl_build_program] successfully built program\n");
 
360
  }
306
361
  return err;
307
362
}
308
363
 
318
373
    for(; k<DT_OPENCL_MAX_KERNELS; k++) if(!cl->dev[dev].kernel_used[k])
319
374
      {
320
375
        cl->dev[dev].kernel_used[k] = 1;
321
 
        cl->dev[dev].kernel[k] = clCreateKernel(cl->dev[dev].program[prog], name, &err);
 
376
        cl->dev[dev].kernel[k] = (cl->dlocl->symbols->dt_clCreateKernel)(cl->dev[dev].program[prog], name, &err);
322
377
        if(err != CL_SUCCESS)
323
378
        {
324
379
          dt_print(DT_DEBUG_OPENCL, "[opencl_create_kernel] could not create kernel `%s'! (%d)\n", name, err);
352
407
  for(int dev=0; dev<cl->num_devs; dev++)
353
408
  {
354
409
    cl->dev[dev].kernel_used [kernel] = 0;
355
 
    clReleaseKernel (cl->dev[dev].kernel [kernel]);
 
410
    (cl->dlocl->symbols->dt_clReleaseKernel) (cl->dev[dev].kernel [kernel]);
356
411
  }
357
412
  dt_pthread_mutex_unlock(&cl->lock);
358
413
}
360
415
int dt_opencl_get_max_work_item_sizes(dt_opencl_t *cl, const int dev, size_t *sizes)
361
416
{
362
417
  if(!cl->inited) return -1;
363
 
  return clGetDeviceInfo(cl->dev[dev].devid, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(size_t)*3, sizes, NULL);
 
418
  return (cl->dlocl->symbols->dt_clGetDeviceInfo)(cl->dev[dev].devid, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(size_t)*3, sizes, NULL);
364
419
}
365
420
 
366
421
int dt_opencl_set_kernel_arg(dt_opencl_t *cl, const int dev, const int kernel, const int num, const size_t size, const void *arg)
367
422
{
368
423
  if(!cl->inited) return -1;
369
424
  if(kernel < 0 || kernel >= DT_OPENCL_MAX_KERNELS) return -1;
370
 
  return clSetKernelArg(cl->dev[dev].kernel[kernel], num, size, arg);
 
425
  return (cl->dlocl->symbols->dt_clSetKernelArg)(cl->dev[dev].kernel[kernel], num, size, arg);
371
426
}
372
427
 
373
428
int dt_opencl_enqueue_kernel_2d(dt_opencl_t *cl, const int dev, const int kernel, const size_t *sizes)
377
432
  // const size_t local[2] = {16, 16};
378
433
  // let the driver choose:
379
434
  const size_t *local = NULL;
380
 
  return clEnqueueNDRangeKernel(cl->dev[dev].cmd_queue, cl->dev[dev].kernel[kernel], 2, NULL, sizes, local, 0, NULL, NULL);
 
435
  return (cl->dlocl->symbols->dt_clEnqueueNDRangeKernel)(cl->dev[dev].cmd_queue, cl->dev[dev].kernel[kernel], 2, NULL, sizes, local, 0, NULL, NULL);
381
436
}
382
437
 
383
 
void dt_opencl_copy_device_to_host(void *host, void *device, const int width, const int height, const int devid, const int bpp)
 
438
int dt_opencl_copy_device_to_host(void *host, void *device, const int width, const int height, const int devid, const int bpp)
384
439
{
385
 
  if(!darktable.opencl->inited) return;
 
440
  if(!darktable.opencl->inited) return -1;
386
441
  size_t origin[] = {0, 0, 0};
387
442
  size_t region[] = {width, height, 1};
388
443
  // blocking.
389
 
  clEnqueueReadImage(darktable.opencl->dev[devid].cmd_queue, device, CL_TRUE, origin, region, region[0]*bpp, 0, host, 0, NULL, NULL);
 
444
  return (darktable.opencl->dlocl->symbols->dt_clEnqueueReadImage)(darktable.opencl->dev[devid].cmd_queue, device, CL_TRUE, origin, region, region[0]*bpp, 0, host, 0, NULL, NULL);
 
445
}
 
446
 
 
447
int dt_opencl_enqueue_copy_image(cl_command_queue q, cl_mem src, cl_mem dst, size_t *orig_src, size_t *orig_dst, size_t *region, int events, cl_event *wait, cl_event *event)
 
448
{
 
449
  if(!darktable.opencl->inited) return -1;
 
450
  cl_int err;
 
451
  err = (darktable.opencl->dlocl->symbols->dt_clEnqueueCopyImage)(q, src, dst, orig_src, orig_dst, region, events, wait, event);
 
452
  if(err != CL_SUCCESS) dt_print(DT_DEBUG_OPENCL, "[opencl copy_image] could not copy image: %d\n", err);
 
453
  return err;
390
454
}
391
455
 
392
456
void* dt_opencl_copy_host_to_device_constant(const int size, const int devid, void *host)
393
457
{
394
458
  if(!darktable.opencl->inited) return NULL;
395
459
  cl_int err;
396
 
  cl_mem dev = clCreateBuffer (darktable.opencl->dev[devid].context,
 
460
  cl_mem dev = (darktable.opencl->dlocl->symbols->dt_clCreateBuffer) (darktable.opencl->dev[devid].context,
397
461
                               CL_MEM_READ_ONLY|CL_MEM_COPY_HOST_PTR,
398
462
                               size,
399
463
                               host, &err);
400
 
  if(err != CL_SUCCESS) fprintf(stderr, "[opencl alloc_device] could not alloc img buffer on device %d: %d\n", devid, err);
 
464
  if(err != CL_SUCCESS) dt_print(DT_DEBUG_OPENCL, "[opencl copy_host_to_device_constant] could not alloc buffer on device %d: %d\n", devid, err);
401
465
  return dev;
402
466
}
403
467
 
425
489
  else return NULL;
426
490
 
427
491
  // TODO: if fmt = uint16_t, blow up to 4xuint16_t and copy manually!
428
 
  cl_mem dev = clCreateImage2D (darktable.opencl->dev[devid].context,
 
492
  cl_mem dev = (darktable.opencl->dlocl->symbols->dt_clCreateImage2D) (darktable.opencl->dev[devid].context,
429
493
                                CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR,
430
494
                                &fmt,
431
495
                                width, height, 0,
432
496
                                host, &err);
433
 
  if(err != CL_SUCCESS) fprintf(stderr, "[opencl copy_host_to_device] could not alloc/copy img buffer onto device %d: %d\n", devid, err);
 
497
  if(err != CL_SUCCESS) dt_print(DT_DEBUG_OPENCL, "[opencl copy_host_to_device] could not alloc/copy img buffer onto device %d: %d\n", devid, err);
434
498
  return dev;
435
499
}
436
500
 
 
501
 
 
502
void dt_opencl_release_mem_object(void *mem)
 
503
{
 
504
  if (!darktable.opencl->inited) return;
 
505
  (darktable.opencl->dlocl->symbols->dt_clReleaseMemObject)(mem);
 
506
}
 
507
 
 
508
 
437
509
void* dt_opencl_alloc_device(const int width, const int height, const int devid, const int bpp)
438
510
{
439
511
  if(!darktable.opencl->inited) return NULL;
457
529
  };
458
530
  else return NULL;
459
531
 
460
 
  cl_mem dev = clCreateImage2D (darktable.opencl->dev[devid].context,
 
532
  cl_mem dev = (darktable.opencl->dlocl->symbols->dt_clCreateImage2D) (darktable.opencl->dev[devid].context,
461
533
                                CL_MEM_READ_WRITE,
462
534
                                &fmt,
463
535
                                width, height, 0,
464
536
                                NULL, &err);
465
 
  if(err != CL_SUCCESS) fprintf(stderr, "[opencl alloc_device] could not alloc img buffer on device %d: %d\n", devid, err);
 
537
  if(err != CL_SUCCESS) dt_print(DT_DEBUG_OPENCL, "[opencl alloc_device] could not alloc img buffer on device %d: %d\n", devid, err);
466
538
  return dev;
467
539
}
468
540
 
 
541
 
 
542
/** check if image size fit into limits given by OpenCL runtime */
 
543
int dt_opencl_image_fits_device(const int devid, const size_t width, const size_t height)
 
544
{
 
545
  if(!darktable.opencl->inited) return FALSE;
 
546
 
 
547
  return (darktable.opencl->dev[devid].max_image_width >= width && darktable.opencl->dev[devid].max_image_height >= height);
 
548
}
 
549
 
 
550
 
 
551
/** check if opencl is inited */
 
552
int dt_opencl_is_inited(void)
 
553
{
 
554
  return darktable.opencl->inited;
 
555
}
 
556
 
 
557
 
 
558
/** check if opencl is enabled */
 
559
int dt_opencl_is_enabled(void)
 
560
{
 
561
  if(!darktable.opencl->inited) return FALSE;
 
562
  return darktable.opencl->enabled;
 
563
}
 
564
 
 
565
 
 
566
/** disable opencl */
 
567
void dt_opencl_disable(void)
 
568
{
 
569
  if(!darktable.opencl->inited) return;
 
570
  darktable.opencl->enabled = FALSE;
 
571
  dt_conf_set_bool("opencl", FALSE);
 
572
}
 
573
 
 
574
 
 
575
/** update enabled flag with value from preferences */
 
576
void dt_opencl_update_enabled(void)
 
577
{
 
578
  if(!darktable.opencl->inited) return;
 
579
  const int prefs = dt_conf_get_bool("opencl");
 
580
 
 
581
  //printf("[opencl_update_enabled] preferences is set to %d\n", prefs);
 
582
 
 
583
  if (darktable.opencl->enabled != prefs)
 
584
  {
 
585
    darktable.opencl->enabled = prefs;
 
586
    dt_print(DT_DEBUG_OPENCL, "[opencl_update_enabled] enabled flag set to %s\n", prefs ? "ON" : "OFF");    
 
587
  }
 
588
  return;
 
589
}
 
590
 
 
591
#endif
 
592