CLIJ deconvolution

Hi everyone!

I have been following this thread about CLIJ decon, and a couple of weeks ago I decided to give it a go with a set of large stacks I’ve been working on for quite a while. I managed to adapt this script to run a deconvolution with a small (41 x 41 x 15) measured PSF, and also to do the blocking (+PSF padding) & re-stitching trick to be able to process my stacks, which are about 9-16 times larger than the chunks my laptop GPU can chew at a time (the stacks are 2057 x 2208 x 22-11 slices (16 bit), 190Mb average each). I finally was able to cobble together a script which processed about 5.5 Gb of raw data in less than three and a half hours. I’m really happy with the results! So, first of all, thanks @haesleinhuepf and @bnorthan!

Now comes the issue part:

I just realized I had a small problem with one of the stacks, which had a couple of blocks over the size limit (the number of blocks is currently hard-coded). When I tried to re-run the same script, I get an ImportError with this line:

from net.haesleinhuepf.clijx.plugins import DeconvolveFFT

Was DeconvolveFTT moved / renamed? I believe there was a recent CLIJ-related update.

Thanks again!

Cheers,
Nico

1 Like

Hey @NicoDF,

glad you like it :slightly_smiling_face:

Yes, we renamed it to name the underlying algorithm. Also its usage changed a bit. Please check the new example script:

Let us know if this solves the problem! And thanks for the feedback :slightly_smiling_face:

Cheers,
Robert

3 Likes

Hi Robert @haesleinhuepf,

I attempted to run the macro script on a Windows machine and I’m getting the same error that @aarandaramos had a while back. It still works fine on my Mac including your recent updates. I’m running the latest version of Fiji with everything current from the update sites. The script was actually working before some update I clicked without thinking and I’ve tried with fresh copies of Fiji and am getting the same error. Were you able to figure out what causes this?

Thanks,
Ben

Hi @Bem,

I think I just fixed this issue. Would you mind updating your Fiji and trying it again?

Thanks for reporting it btw.!

Cheers,
Robert

1 Like

@haesleinhuepf,

It’s working now. Thanks Robert!

Ben

1 Like

Hello Robert! @haesleinhuepf

I am trying the Richardson Lucy decon using the ImageJ macro example (although I’ve also tried the Jython version), however I get an error similar to Ben @Bem above.

I did a fresh FIJI install and have only the two Clij and Clij2 update sites checked in addition to the stock FIJI ones.

Looks like its missing some dependency, could you please fix this?

Also, on a side note, is it possible to implement some residual value cutoff parameter, to stop the iterations early if it is reached, similar to DeconvolutionLab2 “residual” and the Huygens “quality” parameter?

Thank you!
-Julia

(Fiji Is Just) ImageJ 2.1.0/1.53c; Java 1.8.0_172 [64-bit]; Windows 10 10.0; 100MB of 5959MB (1%)

java.lang.RuntimeException: java.lang.UnsatisfiedLinkError: C:\Users\Julia\Documents\Fiji.app\lib\win64\jniclij2fftWrapper.dll: Can't find dependent libraries
   at net.haesleinhuepf.clij.clearcl.util.ElapsedTime.measure(ElapsedTime.java:117)
   at net.haesleinhuepf.clij.clearcl.util.ElapsedTime.measure(ElapsedTime.java:28)
   at net.haesleinhuepf.clij.macro.CLIJHandler.handleExtension(CLIJHandler.java:53)
   at ij.macro.ExtensionDescriptor.dispatch(ExtensionDescriptor.java:288)
   at ij.macro.Functions.doExt(Functions.java:4967)
   at ij.macro.Functions.getStringFunction(Functions.java:278)
   at ij.macro.Interpreter.getStringTerm(Interpreter.java:1475)
   at ij.macro.Interpreter.getString(Interpreter.java:1453)
   at ij.macro.Interpreter.doStatement(Interpreter.java:333)
   at ij.macro.Interpreter.doStatements(Interpreter.java:264)
   at ij.macro.Interpreter.run(Interpreter.java:160)
   at ij.macro.Interpreter.run(Interpreter.java:93)
   at ij.macro.Interpreter.run(Interpreter.java:104)
   at ij.plugin.Macro_Runner.runMacro(Macro_Runner.java:161)
   at ij.IJ.runMacro(IJ.java:153)
   at ij.IJ.runMacro(IJ.java:142)
   at net.imagej.legacy.IJ1Helper$3.call(IJ1Helper.java:1148)
   at net.imagej.legacy.IJ1Helper$3.call(IJ1Helper.java:1144)
   at net.imagej.legacy.IJ1Helper.runMacroFriendly(IJ1Helper.java:1095)
   at net.imagej.legacy.IJ1Helper.runMacro(IJ1Helper.java:1144)
   at net.imagej.legacy.plugin.IJ1MacroEngine.eval(IJ1MacroEngine.java:145)
   at org.scijava.script.ScriptModule.run(ScriptModule.java:157)
   at org.scijava.module.ModuleRunner.run(ModuleRunner.java:165)
   at org.scijava.module.ModuleRunner.call(ModuleRunner.java:124)
   at org.scijava.module.ModuleRunner.call(ModuleRunner.java:63)
   at org.scijava.thread.DefaultThreadService.lambda$wrap$2(DefaultThreadService.java:225)
   at java.util.concurrent.FutureTask.run(FutureTask.java:266)
   at java.util.concurrent.ThreadPoolExecutor.runWorker(ThreadPoolExecutor.java:1149)
   at java.util.concurrent.ThreadPoolExecutor$Worker.run(ThreadPoolExecutor.java:624)
   at java.lang.Thread.run(Thread.java:748)
Caused by: java.lang.UnsatisfiedLinkError: C:\Users\Julia\Documents\Fiji.app\lib\win64\jniclij2fftWrapper.dll: Can't find dependent libraries
   at java.lang.ClassLoader$NativeLibrary.load(Native Method)
   at java.lang.ClassLoader.loadLibrary0(ClassLoader.java:1941)
   at java.lang.ClassLoader.loadLibrary(ClassLoader.java:1857)
   at java.lang.Runtime.loadLibrary0(Runtime.java:870)
   at java.lang.System.loadLibrary(System.java:1122)
   at org.bytedeco.javacpp.Loader.loadLibrary(Loader.java:1302)
   at org.bytedeco.javacpp.Loader.load(Loader.java:1043)
   at org.bytedeco.javacpp.Loader.load(Loader.java:935)
   at net.haesleinhuepf.clijx.plugins.clij2fftWrapper.<clinit>(clij2fftWrapper.java:27)
   at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.runDecon(DeconvolveRichardsonLucyFFT.java:153)
   at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveFFT(DeconvolveRichardsonLucyFFT.java:110)
   at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveRichardsonLucyFFT(DeconvolveRichardsonLucyFFT.java:77)
   at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.executeCL(DeconvolveRichardsonLucyFFT.java:42)
   at net.haesleinhuepf.clij.macro.CLIJHandler.lambda$handleExtension$0(CLIJHandler.java:163)
   at net.haesleinhuepf.clij.clearcl.util.ElapsedTime.measure(ElapsedTime.java:97)
   ... 29 more

2 Likes

Hi Julia @l-jaye,

I must admit, I’m speculating a bit, but would you mind downloading the C++ redistributable package (64-bit) “vc_redist.x64.exe” from here:
https://support.microsoft.com/en-us/help/2977003/the-latest-supported-visual-c-downloads

If this helps, I would add that link to the documentation. Please let me know!

Cheers,
Robert

4 Likes

Hi Robert @haesleinhuepf,

It works wooooohoooo! So awesome!

Just for comparison, DeconvolutionLab2 RL with 25 iterations on 512x512x19 stack took 87.7 seconds (including opening the images, etc), while CLIJ decon on my laptop GPU (integrated Intel UHD 620) took 14.5 seconds, so ~6-fold speed up. On my i7-8550U laptop CPU, the same CLIJ decon took 28.5 seconds.

Thank you so much!
-Julia

3 Likes

Hey Julia,

glad that it works. Let me forward the props to @bnorthan, he did all the work.:slightly_smiling_face:

Thanks for letting us know!

Cheers,
Robert

4 Likes

That a good long term goal. A shorter term goal is to refactor the code to make something like this easier for other developers. The interface needs be changed so there is a parameter for “starting point”. That way one could run the deconvolution for a few iterations, check the result, and if the stopping criteria hasn’t been reached, run it again, passing in the previous result as “starting point” thus restarting the deconvolution from it left off.

Brian

3 Likes

Hello Everyone,

Figured I would try out the CLIJ Richardson-Lucy Deconvolution (using cIFFT, experimental) in CLIJx. On a Red Hat Linux Server System 7.6, Fiji, and GPU is a Tesla V100-PCIE-32GB. Run into the following error.

Thanks in advance,
Kevin

(Fiji Is Just) ImageJ 2.1.0/1.53c; Java 1.8.0_172 [64-bit]; Linux 3.10.0-957.el7.x86_64; 2698MB of 200043MB (1%)

java.lang.UnsatisfiedLinkError: /home2/kdean/Desktop/Applications/Fiji/lib/linux64/libjniclij2fftWrapper.so: /lib64/libstdc++.so.6: version `CXXABI_1.3.8’ not found (required by /home2/kdean/Desktop/Applications/Fiji/lib/linux64/libjniclij2fftWrapper.so)
at java.lang.ClassLoader$NativeLibrary.load(Native Method)
at java.lang.ClassLoader.loadLibrary0(ClassLoader.java:1941)
at java.lang.ClassLoader.loadLibrary(ClassLoader.java:1857)
at java.lang.Runtime.loadLibrary0(Runtime.java:870)
at java.lang.System.loadLibrary(System.java:1122)
at org.bytedeco.javacpp.Loader.loadLibrary(Loader.java:1302)
at org.bytedeco.javacpp.Loader.load(Loader.java:1043)
at org.bytedeco.javacpp.Loader.load(Loader.java:935)
at net.haesleinhuepf.clijx.plugins.clij2fftWrapper.(clij2fftWrapper.java:27)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.runDecon(DeconvolveRichardsonLucyFFT.java:153)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveFFT(DeconvolveRichardsonLucyFFT.java:110)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveRichardsonLucyFFT(DeconvolveRichardsonLucyFFT.java:77)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.executeCL(DeconvolveRichardsonLucyFFT.java:42)
at net.haesleinhuepf.clij.macro.AbstractCLIJPlugin.run(AbstractCLIJPlugin.java:478)
at ij.plugin.filter.PlugInFilterRunner.processOneImage(PlugInFilterRunner.java:265)
at ij.plugin.filter.PlugInFilterRunner.(PlugInFilterRunner.java:114)
at ij.IJ.runUserPlugIn(IJ.java:237)
at ij.IJ.runPlugIn(IJ.java:198)
at ij.Executer.runCommand(Executer.java:150)
at ij.Executer.run(Executer.java:68)
at java.lang.Thread.run(Thread.java:748)

1 Like

Hey @bnorthan ,

to me the error above looks like something is not installed or something should be shipped via the clij ImageJ update site… Do you have a better guess?

@k-dean thanks for reporting! That looks solvable :slightly_smiling_face:

Thanks!

Cheers,
Robert

Is sounds similar to this person’s issue. Do you know what version of gcc is on your system? Maybe updating to version 8 of gcc will fix the problem.

Hello Brian,

Good call. I have to explicitly load gcc, and I didn’t think this would need it. Unfortunately, after loading gcc/8.3.0, I get a different error… This one is a bit long, so I apologize in advance.

Exception

(Fiji Is Just) ImageJ 2.1.0/1.53c; Java 1.8.0_172 [64-bit]; Linux 3.10.0-957.el7.x86_64; 1206MB of 200043MB (<1%)

java.lang.NullPointerException
at net.haesleinhuepf.clij2.CLIJ2.execute(CLIJ2.java:424)
at net.haesleinhuepf.clij2.plugins.Crop3D.crop(Crop3D.java:66)
at net.haesleinhuepf.clij2.CLIJ2Ops.crop(CLIJ2Ops.java:3624)
at net.haesleinhuepf.clijx.plugins.OpenCLFFTUtility.cropExtended(OpenCLFFTUtility.java:260)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveFFT(DeconvolveRichardsonLucyFFT.java:112)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveRichardsonLucyFFT(DeconvolveRichardsonLucyFFT.java:77)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.executeCL(DeconvolveRichardsonLucyFFT.java:42)
at net.haesleinhuepf.clij.macro.AbstractCLIJPlugin.run(AbstractCLIJPlugin.java:478)
at ij.plugin.filter.PlugInFilterRunner.processOneImage(PlugInFilterRunner.java:265)
at ij.plugin.filter.PlugInFilterRunner.(PlugInFilterRunner.java:114)
at ij.IJ.runUserPlugIn(IJ.java:237)
at ij.IJ.runPlugIn(IJ.java:198)
at ij.Executer.runCommand(Executer.java:150)
at ij.Executer.run(Executer.java:68)
at java.lang.Thread.run(Thread.java:748)

Console Output

package net.haesleinhuepf.clijx.plugins, clij2-fft, version 0.0

//###########################################################################
// Preamble:
#pragma OPENCL EXTENSION cl_khr_3d_image_writes : enable

#pragma OPENCL EXTENSION cl_amd_printf : enable

#pragma OPENCL EXTENSION cl_khr_byte_addressable_store : enable

#ifndef M_PI
#define M_PI 3.14159265358979323846f /* pi */
#endif

#ifndef M_LOG2E
#define M_LOG2E 1.4426950408889634074f /* log_2 e */
#endif

#ifndef M_LOG10E
#define M_LOG10E 0.43429448190325182765f /* log_10 e */
#endif

#ifndef M_LN2
#define M_LN2 0.69314718055994530942f /* log_e 2 */
#endif

#ifndef M_LN10
#define M_LN10 2.30258509299404568402f /* log_e 10 */
#endif

#ifndef BUFFER_READ_WRITE
#define BUFFER_READ_WRITE 1

#define MINMAX_TYPE long

inline char2 read_buffer4dc(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global char * buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, position.w};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
// todo: correct pos.w
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height + pos.w * read_buffer_width * read_buffer_height * read_buffer_depth ;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) { // todo: check pos.w
return (char2){0, 0};
}
return (char2){buffer_var[pos_in_buffer],0};
}

inline uchar2 read_buffer4duc(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global uchar * buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, position.w};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
// todo: correct pos.w
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height + pos.w * read_buffer_width * read_buffer_height * read_buffer_depth;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) { // todo: check pos.w
return (uchar2){0, 0};
}
return (uchar2){buffer_var[pos_in_buffer],0};
}

inline short2 read_buffer4di(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global short * buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, position.w};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
// todo: correct pos.w
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height + pos.w * read_buffer_width * read_buffer_height * read_buffer_depth;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) { // todo: check pos.w
return (short2){0, 0};
}
return (short2){buffer_var[pos_in_buffer],0};
}

inline ushort2 read_buffer4dui(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global ushort * buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, position.w};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
// todo: correct pos.w
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height + pos.w * read_buffer_width * read_buffer_height * read_buffer_depth;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) { // todo: check pos.w
return (ushort2){0, 0};
}
return (ushort2){buffer_var[pos_in_buffer],0};
}

inline float2 read_buffer4df(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global float* buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, position.w};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
// todo: correct pos.w
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height + pos.w * read_buffer_width * read_buffer_height * read_buffer_depth;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) { // todo: check pos.w
return (float2){0, 0};
}
return (float2){buffer_var[pos_in_buffer],0};
}

inline void write_buffer4dc(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global char * buffer_var, int4 pos, char value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height + pos.w * write_buffer_width * write_buffer_height * write_buffer_depth;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) { // todo: check pos.w
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer4duc(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global uchar * buffer_var, int4 pos, uchar value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height + pos.w * write_buffer_width * write_buffer_height * write_buffer_depth;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) { // todo: check pos.w
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer4di(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global short * buffer_var, int4 pos, short value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height + pos.w * write_buffer_width * write_buffer_height * write_buffer_depth;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) { // todo: check pos.w
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer4dui(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global ushort * buffer_var, int4 pos, ushort value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height + pos.w * write_buffer_width * write_buffer_height * write_buffer_depth;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) { // todo: check pos.w
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer4df(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global float* buffer_var, int4 pos, float value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height + pos.w * write_buffer_width * write_buffer_height * write_buffer_depth;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) { // todo: check pos.w
return;
}
buffer_var[pos_in_buffer] = value;
}

inline char2 read_buffer3dc(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global char * buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, 0};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) {
return (char2){0, 0};
}
return (char2){buffer_var[pos_in_buffer],0};
}

inline uchar2 read_buffer3duc(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global uchar * buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, 0};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) {
return (uchar2){0, 0};
}
return (uchar2){buffer_var[pos_in_buffer],0};
}

inline short2 read_buffer3di(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global short * buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, 0};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) {
return (short2){0, 0};
}
return (short2){buffer_var[pos_in_buffer],0};
}

inline ushort2 read_buffer3dui(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global ushort * buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, 0};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) {
return (ushort2){0, 0};
}
return (ushort2){buffer_var[pos_in_buffer],0};
}

inline float2 read_buffer3df(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global float* buffer_var, sampler_t sampler, int4 position )
{
int4 pos = (int4){position.x, position.y, position.z, 0};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.z = max((MINMAX_TYPE)pos.z, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
pos.z = min((MINMAX_TYPE)pos.z, (MINMAX_TYPE)read_buffer_depth - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width + (long)pos.z * read_buffer_width * read_buffer_height;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height || pos.z < 0 || (long)pos.z >= read_buffer_depth) {
return (float2){0, 0};
}
return (float2){buffer_var[pos_in_buffer],0};
}

inline void write_buffer3dc(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global char * buffer_var, int4 pos, char value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer3duc(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global uchar * buffer_var, int4 pos, uchar value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer3di(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global short * buffer_var, int4 pos, short value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer3dui(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global ushort * buffer_var, int4 pos, ushort value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer3df(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global float* buffer_var, int4 pos, float value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width + (long)pos.z * write_buffer_width * write_buffer_height;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height || pos.z < 0 || (long)pos.z >= write_buffer_depth) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline char2 read_buffer2dc(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global char * buffer_var, sampler_t sampler, int2 position )
{
int2 pos = (int2){position.x, position.y};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height) {
return (char2){0, 0};
}
return (char2){buffer_var[pos_in_buffer],0};
}

inline uchar2 read_buffer2duc(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global uchar * buffer_var, sampler_t sampler, int2 position )
{
int2 pos = (int2){position.x, position.y};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height) {
return (uchar2){0, 0};
}
return (uchar2){buffer_var[pos_in_buffer],0};
}

inline short2 read_buffer2di(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global short * buffer_var, sampler_t sampler, int2 position )
{
int2 pos = (int2){position.x, position.y};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height) {
return (short2){0, 0};
}
return (short2){buffer_var[pos_in_buffer],0};
}

inline ushort2 read_buffer2dui(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global ushort * buffer_var, sampler_t sampler, int2 position )
{
int2 pos = (int2){position.x, position.y};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height) {
return (ushort2){0, 0};
}
return (ushort2){buffer_var[pos_in_buffer],0};
}

inline float2 read_buffer2df(long read_buffer_width, long read_buffer_height, long read_buffer_depth, __global float* buffer_var, sampler_t sampler, int2 position )
{
int2 pos = (int2){position.x, position.y};
if (true) { // if (CLK_ADDRESS_CLAMP_TO_EDGE & sampler) {
pos.x = max((MINMAX_TYPE)pos.x, (MINMAX_TYPE)0);
pos.y = max((MINMAX_TYPE)pos.y, (MINMAX_TYPE)0);
pos.x = min((MINMAX_TYPE)pos.x, (MINMAX_TYPE)read_buffer_width - 1);
pos.y = min((MINMAX_TYPE)pos.y, (MINMAX_TYPE)read_buffer_height - 1);
}
long pos_in_buffer = (long)pos.x + (long)pos.y * read_buffer_width;
if (pos.x < 0 || (long)pos.x >= read_buffer_width || pos.y < 0 || (long)pos.y >= read_buffer_height) {
return (float2){0, 0};
}
return (float2){buffer_var[pos_in_buffer],0};
}

inline void write_buffer2dc(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global char * buffer_var, int2 pos, char value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer2duc(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global uchar * buffer_var, int2 pos, uchar value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer2di(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global short * buffer_var, int2 pos, short value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer2dui(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global ushort * buffer_var, int2 pos, ushort value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline void write_buffer2df(long write_buffer_width, long write_buffer_height, long write_buffer_depth, __global float* buffer_var, int2 pos, float value )
{
long pos_in_buffer = (long)pos.x + (long)pos.y * write_buffer_width;
if (pos.x < 0 || (long)pos.x >= write_buffer_width || pos.y < 0 || (long)pos.y >= write_buffer_height) {
return;
}
buffer_var[pos_in_buffer] = value;
}

inline uchar clij_convert_uchar_sat(float value) {
if (value > 255) {
return 255;
}
if (value < 0) {
return 0;
}
return (uchar)value;
}

inline char clij_convert_char_sat(float value) {
if (value > 127) {
return 127;
}
if (value < -128) {
return -128;
}
return (char)value;
}

inline ushort clij_convert_ushort_sat(float value) {
if (value > 65535) {
return 65535;
}
if (value < 0) {
return 0;
}
return (ushort)value;
}

inline short clij_convert_short_sat(float value) {
if (value > 32767) {
return 32767;
}
if (value < -32768) {
return -32768;
}
return (short)value;
}

inline uint clij_convert_uint_sat(float value) {
if (value > 4294967295) {
return 4294967295;
}
if (value < 0) {
return 0;
}
return (uint)value;
}

inline int clij_convert_int_sat(float value) {
if (value > 2147483647) {
return 2147483647;
}
if (value < -2147483648) {
return -2147483648;
}
return (int)value;
}

inline float clij_convert_float_sat(float value) {
return value;
}

#define READ_IMAGE(a,b,c) READ_ ## a ## IMAGE(a,b,c)
#define WRITE_IMAGE(a,b,c) WRITE
## a ## _IMAGE(a,b,c)

#endif

//###########################################################################
// Defines:
#define WRITE_dst_IMAGE(a,b,c) write_buffer3df(GET_IMAGE_WIDTH(a),GET_IMAGE_HEIGHT(a),GET_IMAGE_DEPTH(a),a,b,c)
#define IMAGE_src_PIXEL_TYPE ushort
#define CONVERT_dst_PIXEL_TYPE clij_convert_float_sat
#define POS_src_INSTANCE(pos0, pos1, pos2, pos3) ((int4)(pos0, pos1, pos2, pos3))
#define GET_IMAGE_HEIGHT(image_key) image_size_ ## image_key ## height
#define READ_src_IMAGE(a,b,c) read_buffer3dui(GET_IMAGE_WIDTH(a),GET_IMAGE_HEIGHT(a),GET_IMAGE_DEPTH(a),a,b,c)
#define POS_dst_INSTANCE(pos0, pos1, pos2, pos3) ((int4)(pos0, pos1, pos2, pos3))
#define POS_src_TYPE int4
#define GET_IMAGE_DEPTH(image_key) image_size
## image_key ## _depth
#define IMAGE_dst_TYPE long image_size_dst_width, long image_size_dst_height, long image_size_dst_depth, _global float*
#define MAX_ARRAY_SIZE 1000
#define CONVERT_src_PIXEL_TYPE clij_convert_ushort_sat
#define WRITE_src_IMAGE(a,b,c) write_buffer3dui(GET_IMAGE_WIDTH(a),GET_IMAGE_HEIGHT(a),GET_IMAGE_DEPTH(a),a,b,c)
#define IMAGE_dst_PIXEL_TYPE float
#define GET_IMAGE_WIDTH(image_key) image_size
## image_key ## _width
#define IMAGE_src_TYPE long image_size_src_width, long image_size_src_height, long image_size_src_depth, __global ushort*
#define READ_dst_IMAGE(a,b,c) read_buffer3df(GET_IMAGE_WIDTH(a),GET_IMAGE_HEIGHT(a),GET_IMAGE_DEPTH(a),a,b,c)
#define POS_dst_TYPE int4

//###########################################################################
// Source: ‘crop_3d_x.cl’ relative to Crop3D

__kernel void crop_3d(
IMAGE_dst_TYPE dst,
IMAGE_src_TYPE src,
int start_x,
int start_y,
int start_z
) {
const sampler_t sampler = CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_CLAMP_TO_EDGE | CLK_FILTER_NEAREST;

const int dx = get_global_id(0);
const int dy = get_global_id(1);
const int dz = get_global_id(2);

const int sx = start_x + dx;
const int sy = start_y + dy;
const int sz = start_z + dz;

const POS_dst_TYPE dpos = POS_dst_INSTANCE(dx, dy, dz, 0);
const POS_src_TYPE spos = POS_src_INSTANCE(sx, sy, sz, 0);

const float out = READ_src_IMAGE(src,sampler,spos).x;
WRITE_dst_IMAGE(dst,dpos, CONVERT_dst_PIXEL_TYPE(out));
}

Error when trying to create kernel crop_3d
net.haesleinhuepf.clij.clearcl.exceptions.OpenCLException: OpenCL error: -45 → CL_INVALID_PROGRAM_EXECUTABLE
at net.haesleinhuepf.clij.clearcl.backend.BackendUtils.checkOpenCLErrorCode(BackendUtils.java:352)
at net.haesleinhuepf.clij.clearcl.backend.jocl.ClearCLBackendJOCL.lambda$getKernelPeerPointer$19(ClearCLBackendJOCL.java:601)
at net.haesleinhuepf.clij.clearcl.backend.BackendUtils.checkExceptions(BackendUtils.java:156)
at net.haesleinhuepf.clij.clearcl.backend.jocl.ClearCLBackendJOCL.getKernelPeerPointer(ClearCLBackendJOCL.java:593)
at net.haesleinhuepf.clij.clearcl.ClearCLCompiledProgram.createKernel(ClearCLCompiledProgram.java:137)
at net.haesleinhuepf.clij.clearcl.ClearCLProgram.createKernel(ClearCLProgram.java:685)
at net.haesleinhuepf.clij.clearcl.util.CLKernelExecutor.getKernel(CLKernelExecutor.java:353)
at net.haesleinhuepf.clij.clearcl.util.CLKernelExecutor.enqueue(CLKernelExecutor.java:229)
at net.haesleinhuepf.clij2.CLIJ2.lambda$executeSubsequently$0(CLIJ2.java:466)
at net.haesleinhuepf.clij.clearcl.util.ElapsedTime.measure(ElapsedTime.java:97)
at net.haesleinhuepf.clij.clearcl.util.ElapsedTime.measure(ElapsedTime.java:28)
at net.haesleinhuepf.clij2.CLIJ2.executeSubsequently(CLIJ2.java:456)
at net.haesleinhuepf.clij2.CLIJ2.executeSubsequently(CLIJ2.java:443)
at net.haesleinhuepf.clij2.CLIJ2.executeSubsequently(CLIJ2.java:438)
at net.haesleinhuepf.clij2.CLIJ2.execute(CLIJ2.java:423)
at net.haesleinhuepf.clij2.plugins.Crop3D.crop(Crop3D.java:66)
at net.haesleinhuepf.clij2.CLIJ2Ops.crop(CLIJ2Ops.java:3624)
at net.haesleinhuepf.clijx.plugins.OpenCLFFTUtility.cropExtended(OpenCLFFTUtility.java:260)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveFFT(DeconvolveRichardsonLucyFFT.java:112)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveRichardsonLucyFFT(DeconvolveRichardsonLucyFFT.java:77)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.executeCL(DeconvolveRichardsonLucyFFT.java:42)
at net.haesleinhuepf.clij.macro.AbstractCLIJPlugin.run(AbstractCLIJPlugin.java:478)
at ij.plugin.filter.PlugInFilterRunner.processOneImage(PlugInFilterRunner.java:265)
at ij.plugin.filter.PlugInFilterRunner.(PlugInFilterRunner.java:114)
at ij.IJ.runUserPlugIn(IJ.java:237)
at ij.IJ.runPlugIn(IJ.java:198)
at ij.Executer.runCommand(Executer.java:150)
at ij.Executer.run(Executer.java:68)
at java.lang.Thread.run(Thread.java:748)

It looks like it is failing almost at the end where it is cropping the image back to the original size (it is padded for deconvolution). Do you know what size your image and PSF are? I am wondering if something went wrong with the calculation of the padded size and the unpadded region to crop back to, if you let me know size of the image and PSF it could provide a clue.

Hello Brian,

The PSF is 80x80x21 pixels.

The image is 814x866x511 pixels.

To test if it has to do with an odd number, I first removed one image plane from the image that I am deconvolving and I got the same error. Tried the same with the PSF, ran into an error where OpenCL was out of resources.

Exception

(Fiji Is Just) ImageJ 2.1.0/1.53c; Java 1.8.0_172 [64-bit]; Linux 3.10.0-957.el7.x86_64; 1569MB of 200043MB (<1%)

net.haesleinhuepf.clij.clearcl.exceptions.OpenCLException: OpenCL error: -5 → CL_OUT_OF_RESOURCES
at net.haesleinhuepf.clij.clearcl.backend.BackendUtils.checkOpenCLError(BackendUtils.java:346)
at net.haesleinhuepf.clij.clearcl.backend.jocl.ClearCLBackendJOCL.lambda$enqueueWriteToBuffer$25(ClearCLBackendJOCL.java:813)
at net.haesleinhuepf.clij.clearcl.backend.BackendUtils.checkExceptions(BackendUtils.java:171)
at net.haesleinhuepf.clij.clearcl.backend.jocl.ClearCLBackendJOCL.enqueueWriteToBuffer(ClearCLBackendJOCL.java:811)
at net.haesleinhuepf.clij.clearcl.ClearCLBuffer.readFrom(ClearCLBuffer.java:511)
at net.haesleinhuepf.clij.clearcl.ClearCLBuffer.readFrom(ClearCLBuffer.java:477)
at net.haesleinhuepf.clij.converters.implementations.ImagePlusToClearCLBufferConverter.convert(ImagePlusToClearCLBufferConverter.java:126)
at net.haesleinhuepf.clij.converters.implementations.ImagePlusToClearCLBufferConverter.convert(ImagePlusToClearCLBufferConverter.java:24)
at net.haesleinhuepf.clij.CLIJ.convert(CLIJ.java:475)
at net.haesleinhuepf.clij.CLIJ.push(CLIJ.java:406)
at net.haesleinhuepf.clij.macro.CLIJHandler.pushToGPU(CLIJHandler.java:267)
at net.haesleinhuepf.clij.macro.AbstractCLIJPlugin.run(AbstractCLIJPlugin.java:415)
at ij.plugin.filter.PlugInFilterRunner.processOneImage(PlugInFilterRunner.java:265)
at ij.plugin.filter.PlugInFilterRunner.(PlugInFilterRunner.java:114)
at ij.IJ.runUserPlugIn(IJ.java:237)
at ij.IJ.runPlugIn(IJ.java:198)
at ij.Executer.runCommand(Executer.java:150)
at ij.Executer.run(Executer.java:68)
at java.lang.Thread.run(Thread.java:748)

Rebooted Fiji. Try again with even sized PSF (80x80x20) and image (814x866x510), and I get the same Crop3D error.

Exception

(Fiji Is Just) ImageJ 2.1.0/1.53c; Java 1.8.0_172 [64-bit]; Linux 3.10.0-957.el7.x86_64; 2361MB of 200043MB (1%)

java.lang.NullPointerException
at net.haesleinhuepf.clij2.CLIJ2.execute(CLIJ2.java:424)
at net.haesleinhuepf.clij2.plugins.Crop3D.crop(Crop3D.java:66)
at net.haesleinhuepf.clij2.CLIJ2Ops.crop(CLIJ2Ops.java:3624)
at net.haesleinhuepf.clijx.plugins.OpenCLFFTUtility.cropExtended(OpenCLFFTUtility.java:260)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveFFT(DeconvolveRichardsonLucyFFT.java:112)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.deconvolveRichardsonLucyFFT(DeconvolveRichardsonLucyFFT.java:77)
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.executeCL(DeconvolveRichardsonLucyFFT.java:42)
at net.haesleinhuepf.clij.macro.AbstractCLIJPlugin.run(AbstractCLIJPlugin.java:478)
at ij.plugin.filter.PlugInFilterRunner.processOneImage(PlugInFilterRunner.java:265)
at ij.plugin.filter.PlugInFilterRunner.(PlugInFilterRunner.java:114)
at ij.IJ.runUserPlugIn(IJ.java:237)
at ij.IJ.runPlugIn(IJ.java:198)
at ij.Executer.runCommand(Executer.java:150)
at ij.Executer.run(Executer.java:65)
at java.lang.Thread.run(Thread.java:748)

Thank you,
Kevin

Hi Kevin

Assuming the 32G is all on the same card, which I think is the case with V100 that image should work (if your setup is made of multiple smaller cards you will have to crop a smaller region, we don’ti yet have the functionality to divide into sub-vols and recombine).

I recently fixed an issue that occurred when accessing memory locations > 4G. You could try copying the files in the link below to your installation (copy from Linux directory to your Fiji.app, and keep the same folder structure). Also start up Fiji from a terminal, there are a bunch of extra diagnostic messages that are sent to the terminal, and they might provide some clues.

I wasn’t sure if I should I remove a selection of the existing components from the Fiji/lib/linux64 directory (libblosc.so and libclFFT.so)? I also assumed that I should remove the older FFT plugin (clij2-fft_-2.1.3.1-BETA.jar).

In both cases, I did get an error for the full size image.

Exception

(Fiji Is Just) ImageJ 2.1.0/1.53c; Java 1.8.0_172 [64-bit]; Linux 3.10.0-957.el7.x86_64; 1441MB of 200043MB (<1%)

java.lang.ClassCastException: java.lang.Double cannot be cast to java.lang.Boolean
at net.haesleinhuepf.clijx.plugins.DeconvolveRichardsonLucyFFT.executeCL(DeconvolveRichardsonLucyFFT.java:65)
at net.haesleinhuepf.clij.macro.AbstractCLIJPlugin.run(AbstractCLIJPlugin.java:478)
at ij.plugin.filter.PlugInFilterRunner.processOneImage(PlugInFilterRunner.java:265)
at ij.plugin.filter.PlugInFilterRunner.(PlugInFilterRunner.java:114)
at ij.IJ.runUserPlugIn(IJ.java:237)
at ij.IJ.runPlugIn(IJ.java:198)
at ij.Executer.runCommand(Executer.java:150)
at ij.Executer.run(Executer.java:68)
at java.lang.Thread.run(Thread.java:748)

Console Output

package net.haesleinhuepf.clijx.plugins, clijx_, version 0.0[/details]

Terminal Output

[kdean@NucleusC040 ~]$ clear; module add gcc/8.3.0; ‘/home2/kdean/Desktop/ImageJ-linux64’

Java HotSpot™ 64-Bit Server VM warning: ignoring option PermSize=128m; support was removed in 8.0
Java HotSpot™ 64-Bit Server VM warning: Using incremental CMS is deprecated and will likely be removed in a future release
[WARNING] Not overwriting extension ‘py’:
proposed = net.haesleinhuepf.clijx.te_oki.TeOkiLanguage [file:/home2/kdean/Desktop/Applications/Fiji/plugins/clijx-assistant_-0.4.2.16.jar
existing = org.scijava.plugins.scripting.jython.JythonScriptLanguage [file:/home2/kdean/Desktop/Applications/Fiji/jars/scripting-jython-1.0.0.jar
[WARNING] Not overwriting extension ‘ijm’:
proposed = net.clesperanto.macro.interpreter.ClEsperantoMacroLanguage [file:/home2/kdean/Desktop/Applications/Fiji/plugins/clijx-assistant_-0.4.2.16.jar
existing = net.imagej.legacy.plugin.IJ1MacroLanguage [file:/home2/kdean/Desktop/Applications/Fiji/jars/imagej-legacy-0.37.4.jar
[INFO] Overriding Ice ; identifier: script:LUTs/CMOcean/Ice.ijm; jar: file:/home2/kdean/Desktop/Applications/Fiji/jars/scijava-common-2.83.3.jar
[INFO] Overriding Visualise vector field (experimental); identifier: command:net.haesleinhuepf.clijx.piv.visualisation.VisualiseVectorFieldsPlugin; jar: file:/home2/kdean/Desktop/Applications/Fiji/plugins/clijx_-0.30.1.16.jar
package net.haesleinhuepf.clijx.plugins, clijx_, version 0.0

I also tried it with the libblosc.so and libclFFT.so files in the lib/linux64 directory, and got a similar result.

And lastly, I cropped down the image substantially (219x232x510, 16-bit, 49MB), and still had the same issue.

Thank you,
Kevin

Should it be helpful, I’ve placed the cropped image, and the PSF, here: BioHPC File Exchange

OK. That error was just a GUI thing where it did not parse the non-Circulant checkbox properly. I just updated clij2-fft_-2.2.0.10 on dropbox.

You’ll also have to change the regularization factor from 2.0 to a number below 0,005 (change it to 0 to turn regularization off). You can turn ‘nonCirculant’ on or off. Shouldn’t matter too much. If you have extended objects in your image turning nonCirculant on might help a bit.