1 /*
2 This file is part of darktable,
3 Copyright (C) 2010-2020 darktable developers.
4
5 darktable is free software: you can redistribute it and/or modify
6 it under the terms of the GNU General Public License as published by
7 the Free Software Foundation, either version 3 of the License, or
8 (at your option) any later version.
9
10 darktable is distributed in the hope that it will be useful,
11 but WITHOUT ANY WARRANTY; without even the implied warranty of
12 MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
13 GNU General Public License for more details.
14
15 You should have received a copy of the GNU General Public License
16 along with darktable. If not, see <http://www.gnu.org/licenses/>.
17 */
18 #ifdef HAVE_CONFIG_H
19 #include "config.h"
20 #endif
21 #include <stdlib.h>
22 #if defined(__SSE__)
23 #include <xmmintrin.h>
24 #endif
25 #include <cairo.h>
26
27 #include "common/opencl.h"
28 #include "common/imagebuf.h"
29 #include "common/iop_profile.h"
30 #include "control/control.h"
31 #include "develop/develop.h"
32 #include "develop/imageop.h"
33 #include "develop/imageop_math.h"
34 #include "gui/accelerators.h"
35 #include "iop/iop_api.h"
36
37 DT_MODULE(3)
38
39 typedef enum dt_iop_overexposed_colorscheme_t
40 {
41 DT_IOP_OVEREXPOSED_BLACKWHITE = 0,
42 DT_IOP_OVEREXPOSED_REDBLUE = 1,
43 DT_IOP_OVEREXPOSED_PURPLEGREEN = 2
44 } dt_iop_overexposed_colorscheme_t;
45
46 static const float DT_ALIGNED_ARRAY dt_iop_overexposed_colors[][2][4]
47 = { {
48 { 0.0f, 0.0f, 0.0f, 1.0f }, // black
49 { 1.0f, 1.0f, 1.0f, 1.0f } // white
50 },
51 {
52 { 1.0f, 0.0f, 0.0f, 1.0f }, // red
53 { 0.0f, 0.0f, 1.0f, 1.0f } // blue
54 },
55 {
56 { 0.371f, 0.434f, 0.934f, 1.0f }, // purple (#5f6fef)
57 { 0.512f, 0.934f, 0.371f, 1.0f } // green (#83ef5f)
58 } };
59
60 typedef struct dt_iop_overexposed_global_data_t
61 {
62 int kernel_overexposed;
63 } dt_iop_overexposed_global_data_t;
64
65 typedef struct dt_iop_overexposed_t
66 {
67 int dummy;
68 } dt_iop_overexposed_t;
69
name()70 const char *name()
71 {
72 return _("overexposed");
73 }
74
default_group()75 int default_group()
76 {
77 return IOP_GROUP_BASIC | IOP_GROUP_TECHNICAL;
78 }
79
flags()80 int flags()
81 {
82 return IOP_FLAGS_ALLOW_TILING | IOP_FLAGS_HIDDEN | IOP_FLAGS_ONE_INSTANCE | IOP_FLAGS_NO_HISTORY_STACK;
83 }
84
default_colorspace(dt_iop_module_t * self,dt_dev_pixelpipe_t * pipe,dt_dev_pixelpipe_iop_t * piece)85 int default_colorspace(dt_iop_module_t *self, dt_dev_pixelpipe_t *pipe, dt_dev_pixelpipe_iop_t *piece)
86 {
87 return iop_cs_rgb;
88 }
89
90
legacy_params(dt_iop_module_t * self,const void * const old_params,const int old_version,void * new_params,const int new_version)91 int legacy_params(dt_iop_module_t *self, const void *const old_params, const int old_version,
92 void *new_params, const int new_version)
93 {
94 // we do no longer have module params in here and just ignore any legacy entries
95 return 0;
96 }
97
98
99 // void init_key_accels(dt_iop_module_so_t *self)
100 // {
101 // dt_accel_register_slider_iop(self, FALSE, NC_("accel", "lower threshold"));
102 // dt_accel_register_slider_iop(self, FALSE, NC_("accel", "upper threshold"));
103 // dt_accel_register_slider_iop(self, FALSE, NC_("accel", "color scheme"));
104 // }
105 //
106 // void connect_key_accels(dt_iop_module_t *self)
107 // {
108 // dt_iop_overexposed_gui_data_t *g =
109 // (dt_iop_overexposed_gui_data_t*)self->gui_data;
110 //
111 // dt_accel_connect_slider_iop(self, "lower threshold", GTK_WIDGET(g->lower));
112 // dt_accel_connect_slider_iop(self, "upper threshold", GTK_WIDGET(g->upper));
113 // dt_accel_connect_slider_iop(self, "color scheme", GTK_WIDGET(g->colorscheme));
114 // }
115
process(struct dt_iop_module_t * self,dt_dev_pixelpipe_iop_t * piece,const void * const ivoid,void * const ovoid,const dt_iop_roi_t * const roi_in,const dt_iop_roi_t * const roi_out)116 void process(struct dt_iop_module_t *self, dt_dev_pixelpipe_iop_t *piece, const void *const ivoid,
117 void *const ovoid, const dt_iop_roi_t *const roi_in, const dt_iop_roi_t *const roi_out)
118 {
119 if (!dt_iop_have_required_input_format(4 /*we need full-color pixels*/, piece->module, piece->colors,
120 ivoid, ovoid, roi_in, roi_out))
121 return; // image has been copied through to output and module's trouble flag has been updated
122
123 dt_develop_t *dev = self->dev;
124
125 const int ch = 4;
126
127 float *restrict img_tmp = NULL;
128 if (!dt_iop_alloc_image_buffers(self, roi_in, roi_out, ch, &img_tmp, 0))
129 {
130 dt_iop_copy_image_roi(ovoid, ivoid, ch, roi_in, roi_out, TRUE);
131 dt_control_log(_("module overexposed failed in buffer allocation"));
132 return;
133 }
134
135 const float lower = exp2f(fminf(dev->overexposed.lower, -4.f)); // in EV
136 const float upper = dev->overexposed.upper / 100.0f; // in %
137
138 const int colorscheme = dev->overexposed.colorscheme;
139 const float *const upper_color = dt_iop_overexposed_colors[colorscheme][0];
140 const float *const lower_color = dt_iop_overexposed_colors[colorscheme][1];
141
142 const float *const restrict in = __builtin_assume_aligned((const float *const restrict)ivoid, 64);
143 float *const restrict out = __builtin_assume_aligned((float *const restrict)ovoid, 64);
144
145 const dt_iop_order_iccprofile_info_t *const current_profile = dt_ioppr_get_pipe_current_profile_info(self, piece->pipe);
146 const dt_iop_order_iccprofile_info_t *const work_profile = dt_ioppr_get_histogram_profile_info(dev);
147
148 // display mask using histogram profile as output
149 // FIXME: the histogram already does this work -- use that data instead?
150 if(current_profile && work_profile)
151 dt_ioppr_transform_image_colorspace_rgb(in, img_tmp, roi_out->width, roi_out->height, current_profile,
152 work_profile, self->op);
153 else
154 {
155 fprintf(stderr, "[overexposed process] can't create transform profile\n");
156 dt_iop_copy_image_roi(ovoid, ivoid, ch, roi_in, roi_out, TRUE);
157 dt_control_log(_("module overexposed failed in color conversion"));
158 goto process_finish;
159 }
160
161
162 #ifdef __SSE2__
163 // flush denormals to zero to avoid performance penalty if there are a lot of near-zero values
164 const unsigned int oldMode = _MM_GET_FLUSH_ZERO_MODE();
165 _MM_SET_FLUSH_ZERO_MODE(_MM_FLUSH_ZERO_ON);
166 #endif
167
168 if(dev->overexposed.mode == DT_CLIPPING_PREVIEW_ANYRGB)
169 {
170 // Any of the RGB channels is out of bounds
171 #ifdef _OPENMP
172 #pragma omp parallel for default(none) \
173 dt_omp_firstprivate(ch, img_tmp, in, lower, lower_color, out, roi_out, \
174 upper, upper_color) \
175 schedule(static)
176 #endif
177 for(size_t k = 0; k < (size_t)ch * roi_out->width * roi_out->height; k += ch)
178 {
179 if(img_tmp[k + 0] >= upper || img_tmp[k + 1] >= upper || img_tmp[k + 2] >= upper)
180 {
181 copy_pixel(out + k, upper_color);
182 }
183 else if(img_tmp[k + 0] <= lower && img_tmp[k + 1] <= lower && img_tmp[k + 2] <= lower)
184 {
185 copy_pixel(out + k, lower_color);
186 }
187 else
188 {
189 copy_pixel(out + k, in + k);
190 }
191 }
192 }
193
194 else if(dev->overexposed.mode == DT_CLIPPING_PREVIEW_GAMUT && work_profile)
195 {
196 // Gamut is out of bounds
197 #ifdef _OPENMP
198 #pragma omp parallel for default(none) \
199 dt_omp_firstprivate(ch, img_tmp, in, lower, lower_color, out, roi_out, \
200 upper, upper_color, work_profile) \
201 schedule(static)
202 #endif
203 for(size_t k = 0; k < (size_t)ch * roi_out->width * roi_out->height; k += ch)
204 {
205 const float luminance = dt_ioppr_get_rgb_matrix_luminance(img_tmp + k,
206 work_profile->matrix_in, work_profile->lut_in,
207 work_profile->unbounded_coeffs_in,
208 work_profile->lutsize, work_profile->nonlinearlut);
209
210 // luminance is out of bounds
211 if(luminance >= upper)
212 {
213 copy_pixel(out + k, upper_color);
214 }
215 else if(luminance <= lower)
216 {
217 copy_pixel(out + k, lower_color);
218 }
219 // luminance is ok, so check for saturation
220 else
221 {
222 dt_aligned_pixel_t saturation = { 0.f };
223
224 for_each_channel(c,aligned(saturation, img_tmp : 64))
225 {
226 saturation[c] = (img_tmp[k + c] - luminance);
227 saturation[c] = sqrtf(saturation[c] * saturation[c] / (luminance * luminance + img_tmp[k + c] * img_tmp[k + c]));
228 }
229
230 // we got over-saturation, relatively to luminance or absolutely over RGB
231 if(saturation[0] > upper || saturation[1] > upper || saturation[2] > upper ||
232 img_tmp[k + 0] >= upper || img_tmp[k + 1] >= upper || img_tmp[k + 2] >= upper)
233 {
234 copy_pixel(out + k, upper_color);
235 }
236
237 // saturation is fine but we got out-of-bounds RGB
238 else if(img_tmp[k + 0] <= lower && img_tmp[k + 1] <= lower && img_tmp[k + 2] <= lower)
239 {
240 copy_pixel(out + k, lower_color);
241 }
242
243 // evererything is fine
244 else
245 {
246 copy_pixel(out + k, in + k);
247 }
248 }
249 }
250 }
251
252 else if(dev->overexposed.mode == DT_CLIPPING_PREVIEW_LUMINANCE && work_profile)
253 {
254 // Luminance channel is out of bounds
255 #ifdef _OPENMP
256 #pragma omp parallel for default(none) \
257 dt_omp_firstprivate(ch, img_tmp, in, lower, lower_color, out, roi_out, \
258 upper, upper_color, work_profile) \
259 schedule(static)
260 #endif
261 for(size_t k = 0; k < (size_t)ch * roi_out->width * roi_out->height; k += ch)
262 {
263 const float luminance = dt_ioppr_get_rgb_matrix_luminance(img_tmp + k,
264 work_profile->matrix_in, work_profile->lut_in,
265 work_profile->unbounded_coeffs_in,
266 work_profile->lutsize, work_profile->nonlinearlut);
267
268 if(luminance >= upper)
269 {
270 copy_pixel(out + k, upper_color);
271 }
272
273 else if(luminance <= lower)
274 {
275 copy_pixel(out + k, lower_color);
276 }
277 else
278 {
279 copy_pixel(out + k, in + k);
280 }
281 }
282 }
283
284 else if(dev->overexposed.mode == DT_CLIPPING_PREVIEW_SATURATION && work_profile)
285 {
286 // Show saturation out of bounds where luminance is valid
287 #ifdef _OPENMP
288 #pragma omp parallel for default(none) \
289 dt_omp_firstprivate(ch, img_tmp, in, lower, lower_color, out, roi_out, \
290 upper, upper_color, work_profile) \
291 schedule(static)
292 #endif
293 for(size_t k = 0; k < (size_t)ch * roi_out->width * roi_out->height; k += ch)
294 {
295 const float luminance = dt_ioppr_get_rgb_matrix_luminance(img_tmp + k,
296 work_profile->matrix_in, work_profile->lut_in,
297 work_profile->unbounded_coeffs_in,
298 work_profile->lutsize, work_profile->nonlinearlut);
299 if(luminance < upper && luminance > lower)
300 {
301 dt_aligned_pixel_t saturation = { 0.f };
302
303 for_each_channel(c,aligned(saturation, img_tmp : 64))
304 {
305 saturation[c] = (img_tmp[k + c] - luminance);
306 saturation[c] = sqrtf(saturation[c] * saturation[c] / (luminance * luminance + img_tmp[k + c] * img_tmp[k + c]));
307 }
308
309 // we got over-saturation, relatively to luminance or absolutely over RGB
310 if(saturation[0] > upper || saturation[1] > upper || saturation[2] > upper ||
311 img_tmp[k + 0] >= upper || img_tmp[k + 1] >= upper || img_tmp[k + 2] >= upper)
312 {
313 copy_pixel(out + k, upper_color);
314 }
315 else if(img_tmp[k + 0] <= lower && img_tmp[k + 1] <= lower && img_tmp[k + 2] <= lower)
316 {
317 copy_pixel(out + k, lower_color);
318 }
319 else
320 {
321 copy_pixel(out + k, in + k);
322 }
323 }
324
325 else
326 {
327 copy_pixel(out + k, in + k);
328 }
329 }
330 }
331
332 #ifdef __SSE2__
333 _MM_SET_FLUSH_ZERO_MODE(oldMode);
334 #endif
335
336 if(piece->pipe->mask_display & DT_DEV_PIXELPIPE_DISPLAY_MASK)
337 dt_iop_alpha_copy(ivoid, ovoid, roi_out->width, roi_out->height);
338
339 process_finish:
340 dt_free_align(img_tmp);
341 }
342
343 #ifdef HAVE_OPENCL
process_cl(struct dt_iop_module_t * self,dt_dev_pixelpipe_iop_t * piece,cl_mem dev_in,cl_mem dev_out,const dt_iop_roi_t * const roi_in,const dt_iop_roi_t * const roi_out)344 int process_cl(struct dt_iop_module_t *self, dt_dev_pixelpipe_iop_t *piece, cl_mem dev_in, cl_mem dev_out,
345 const dt_iop_roi_t *const roi_in, const dt_iop_roi_t *const roi_out)
346 {
347 dt_develop_t *dev = self->dev;
348 dt_iop_overexposed_global_data_t *gd = (dt_iop_overexposed_global_data_t *)self->global_data;
349
350 cl_int err = -999;
351 const int devid = piece->pipe->devid;
352
353 const int ch = piece->colors;
354 cl_mem dev_tmp = NULL;
355
356 const int width = roi_out->width;
357 const int height = roi_out->height;
358
359 const dt_iop_order_iccprofile_info_t *const current_profile = dt_ioppr_get_pipe_current_profile_info(self, piece->pipe);
360 const dt_iop_order_iccprofile_info_t *const work_profile = dt_ioppr_get_histogram_profile_info(dev);
361
362 // display mask using histogram profile as output
363 dev_tmp = dt_opencl_alloc_device(devid, width, height, sizeof(float) * ch);
364 if(dev_tmp == NULL)
365 {
366 err = CL_MEM_OBJECT_ALLOCATION_FAILURE;
367 fprintf(stderr, "[overexposed process_cl] error allocating memory for color transformation\n");
368 dt_control_log(_("module overexposed failed in buffer allocation"));
369 goto error;
370 }
371
372 if(current_profile && work_profile)
373 dt_ioppr_transform_image_colorspace_rgb_cl(devid, dev_in, dev_tmp, roi_out->width, roi_out->height,
374 current_profile, work_profile, self->op);
375 else
376 {
377 fprintf(stderr, "[overexposed process_cl] can't create transform profile\n");
378 dt_control_log(_("module overexposed failed in color conversion"));
379 goto error;
380 }
381
382 const int use_work_profile = (work_profile == NULL) ? 0 : 1;
383 cl_mem dev_profile_info = NULL;
384 cl_mem dev_profile_lut = NULL;
385 dt_colorspaces_iccprofile_info_cl_t *profile_info_cl;
386 cl_float *profile_lut_cl = NULL;
387
388 err = dt_ioppr_build_iccprofile_params_cl(work_profile, devid, &profile_info_cl, &profile_lut_cl,
389 &dev_profile_info, &dev_profile_lut);
390 if(err != CL_SUCCESS) goto error;
391
392 const float lower = exp2f(fminf(dev->overexposed.lower, -4.f)); // in EV
393 const float upper = dev->overexposed.upper / 100.0f; // in %
394 const int colorscheme = dev->overexposed.colorscheme;
395
396 const float *upper_color = dt_iop_overexposed_colors[colorscheme][0];
397 const float *lower_color = dt_iop_overexposed_colors[colorscheme][1];
398 const int mode = dev->overexposed.mode;
399
400 size_t sizes[2] = { ROUNDUPWD(width), ROUNDUPHT(height) };
401 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 0, sizeof(cl_mem), &dev_in);
402 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 1, sizeof(cl_mem), &dev_out);
403 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 2, sizeof(cl_mem), &dev_tmp);
404 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 3, sizeof(int), &width);
405 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 4, sizeof(int), &height);
406 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 5, sizeof(float), &lower);
407 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 6, sizeof(float), &upper);
408 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 7, 4 * sizeof(float), lower_color);
409 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 8, 4 * sizeof(float), upper_color);
410 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 9, sizeof(cl_mem), (void *)&dev_profile_info);
411 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 10, sizeof(cl_mem), (void *)&dev_profile_lut);
412 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 11, sizeof(int), (void *)&use_work_profile);
413 dt_opencl_set_kernel_arg(devid, gd->kernel_overexposed, 12, sizeof(int), (void *)&mode);
414 err = dt_opencl_enqueue_kernel_2d(devid, gd->kernel_overexposed, sizes);
415 if(err != CL_SUCCESS) goto error;
416 if(dev_tmp) dt_opencl_release_mem_object(dev_tmp);
417 return TRUE;
418
419 error:
420 if(dev_tmp) dt_opencl_release_mem_object(dev_tmp);
421 dt_print(DT_DEBUG_OPENCL, "[opencl_overexposed] couldn't enqueue kernel! %d\n", err);
422 return FALSE;
423 }
424 #endif
425
426
init_global(dt_iop_module_so_t * module)427 void init_global(dt_iop_module_so_t *module)
428 {
429 const int program = 2; // basic.cl from programs.conf
430 dt_iop_overexposed_global_data_t *gd
431 = (dt_iop_overexposed_global_data_t *)malloc(sizeof(dt_iop_overexposed_global_data_t));
432 module->data = gd;
433 gd->kernel_overexposed = dt_opencl_create_kernel(program, "overexposed");
434 }
435
436
cleanup_global(dt_iop_module_so_t * module)437 void cleanup_global(dt_iop_module_so_t *module)
438 {
439 dt_iop_overexposed_global_data_t *gd = (dt_iop_overexposed_global_data_t *)module->data;
440 dt_opencl_free_kernel(gd->kernel_overexposed);
441 free(module->data);
442 module->data = NULL;
443 }
444
commit_params(struct dt_iop_module_t * self,dt_iop_params_t * p1,dt_dev_pixelpipe_t * pipe,dt_dev_pixelpipe_iop_t * piece)445 void commit_params(struct dt_iop_module_t *self, dt_iop_params_t *p1, dt_dev_pixelpipe_t *pipe,
446 dt_dev_pixelpipe_iop_t *piece)
447 {
448 if(pipe->type != DT_DEV_PIXELPIPE_FULL || !self->dev->overexposed.enabled || !self->dev->gui_attached)
449 piece->enabled = 0;
450 }
451
init_pipe(struct dt_iop_module_t * self,dt_dev_pixelpipe_t * pipe,dt_dev_pixelpipe_iop_t * piece)452 void init_pipe(struct dt_iop_module_t *self, dt_dev_pixelpipe_t *pipe, dt_dev_pixelpipe_iop_t *piece)
453 {
454 piece->data = NULL;
455 }
456
cleanup_pipe(struct dt_iop_module_t * self,dt_dev_pixelpipe_t * pipe,dt_dev_pixelpipe_iop_t * piece)457 void cleanup_pipe(struct dt_iop_module_t *self, dt_dev_pixelpipe_t *pipe, dt_dev_pixelpipe_iop_t *piece)
458 {
459 }
460
init(dt_iop_module_t * module)461 void init(dt_iop_module_t *module)
462 {
463 module->params = calloc(1, sizeof(dt_iop_overexposed_t));
464 module->default_params = calloc(1, sizeof(dt_iop_overexposed_t));
465 module->hide_enable_button = 1;
466 module->default_enabled = 1;
467 module->params_size = sizeof(dt_iop_overexposed_t);
468 module->gui_data = NULL;
469 }
470
471 // modelines: These editor modelines have been set for all relevant files by tools/update_modelines.sh
472 // vim: shiftwidth=2 expandtab tabstop=2 cindent
473 // kate: tab-indents: off; indent-width 2; replace-tabs on; indent-mode cstyle; remove-trailing-spaces modified;
474