[ImageJ-devel] OpenCL plugin
Rick Lentz
rwlentz at wisc.edu
Fri Jan 14 10:56:49 CST 2011
Hi Pol,
The Open Source community has been a great for me regarding collaboration
and support. For me, understanding the nature of OpenCL came from previous
work done with CUDA. It helped me to review free online videos of academic
lectures (iTunesU - GPU computing) and articles pertaining to the device
architecture (optimization) since this understanding help me organize a plan
to implement an algorithm for a specific platform.
Here is a 32 bit version of the sobel.cl that should run fine on the GTS
240 (that has also been committed in the repository in place of the old
'double' version):
__kernel void sobel( __global float* input,
__global float* output,
int width,
int height )
{
int x = get_global_id(0);
int y = get_global_id(1);
int offset = y * width + x;
float p0, p1, p2, p3, p5, p6, p7, p8 = 0;
if( x < 1 || y < 1 || x > width - 2 || y > height - 2 )
{
output[offset] = 0;
}
else
{
p0 = input[offset - width - 1] ;
p1 = input[offset - width] ;
p2 = input[offset - width + 1] ;
p3 = input[offset - 1] ;
p5 = input[offset + 1] ;
p6 = input[offset + width - 1] ;
p7 = input[offset + width] ;
p8 = input[offset + width + 1] ;
float sum1 = p0 + 2*p1 + p2 - p6 - 2*p7 - p8; //GY
float sum2 = p0 + 2*p3 + p6 - p2 - 2*p5 - p8; //GX
output[offset] = sqrt( sum1*sum1 + sum2*sum2 );
}
}
For your last question, you can write OpenCL directly from Java if that is
easier for you. Here is an example from a unit test written by Michael
Bien:
/**
* This test was authored by Michael Bien to help identify a platform
specific bug
* identified on OSX (specifically 10.6)
* @author Michael Bien
*
*/
public class ProgramTest {
@BeforeClass
public synchronized static void setUpClass() throws Exception {
out.println("OS: " + System.getProperty("os.name"));
out.println("ARCH: " + System.getProperty("os.arch"));
out.println("VM: " + System.getProperty("java.vm.name"));
out.println("lib path: " + System.getProperty("java.library.path"));
}
// NOTE THE OPENCL PROGRAM SOURCE HAS BEEN ADDED INSIDE THE JAVA CLASS
DECLARATION.
private final static String programSource =""
// + " #pragma OPENCL EXTENSION cl_khr_fp64: enable \n"
+" __kernel void sobel( __global float* input, __global
float* output, int width, int height ) { \n"
+" int x = get_global_id(0); \n"
+" int y = get_global_id(1); \n"
+" int offset = y * width + x; \n"
+"
\n"
+" float p0, p1, p2, p3, p5, p6, p7, p8 = 0;
\n"
+"
\n"
+"
\n"
+" if( x < 1 || y < 1 || x > width - 2 || y >
height - 2 ) \n"
+" { \n"
+" output[offset] = 0; \n"
+" } \n"
+" else \n"
+" { \n"
+" p0 = input[offset - width - 1] ; \n"
+" p1 = input[offset - width] ; \n"
+" p2 = input[offset - width + 1] ; \n"
+" p3 = input[offset - 1] ; \n"
+" p5 = input[offset + 1] ; \n"
+" p6 = input[offset + width - 1] ; \n"
+" p7 = input[offset + width] ; \n"
+" p8 = input[offset + width + 1] ; \n"
+" \n"
+" float sum1 = p0 + 2*p1 + p2 - p6 - 2*p7 -
p8; //GY \n"
+" float sum2 = p0 + 2*p3 + p6 - p2 - 2*p5 -
p8; //GX \n"
+" \n"
+" output[offset] = sqrt( sum1*sum1 +
sum2*sum2 ); \n"
+" } \n"
+" } ";
@Test
public synchronized void buildProgramTest() {
CLContext context = CLContext.create();
try {
System.out.println(context);
System.out.println(context.getPlatform().getVersion());
long contextID = context.ID;
CL cl = CLPlatform.getLowLevelCLInterface();
PointerBuffer buffer = (PointerBuffer)
PointerBuffer.allocateDirect(1).put(programSource.length());
String[] srcArray = new String[]{programSource};
IntBuffer uploadStatus = Buffers.newDirectIntBuffer(1);
final long programID = cl.clCreateProgramWithSource(contextID,
1, srcArray, buffer, uploadStatus);
checkError("on clCreateProgramWithSource", uploadStatus.get(0));
// Build the program
int buildStatus = cl.clBuildProgram(programID, 0, null, null,
null);
System.out.println("please ignore "+srcArray+ "" + buffer); //
please ignore, just a artificial reference lock
System.out.println("src: " + getProgramInfoString(cl, programID,
CL.CL_PROGRAM_SOURCE));
checkError("on clBuildProgram", buildStatus);
} finally {
context.release();
System.out.println("-> success");
}
}
private synchronized String getProgramInfoString(CL cl, long program,
int flag) {
PointerBuffer size = PointerBuffer.allocateDirect(1);
int ret = cl.clGetProgramInfo(program, flag, 0, null, size);
checkError("on clGetProgramInfo", ret);
ByteBuffer buffer = Buffers.newDirectByteBuffer((int)size.get(0));
ret = cl.clGetProgramInfo(program, flag, buffer.capacity(), buffer,
null);
checkError("on clGetProgramInfo", ret);
return CLUtil.clString2JavaString(buffer, (int)size.get(0));
}
private synchronized void checkError(String msg, int ret) {
if(ret != CL.CL_SUCCESS)
throw CLException.newException(ret, msg);
}
}
For me, I prefer to place the OpenCL code in separate files because it is
easier for me to reuse them. When I first started using OpenCL from Java - I
liked to see the OpenCL inside the Java class. Over time reusability, use
of runtime resource loading, and readability caused me to separate the CL
code from the Java code.
ImageJ plugins can also access GPU resources in other languages other than
OpenCL. CUDA is specific to NVidia's hardware and they have produced over
1B devices. Depending on the users of your ImageJ plugin - one might be
preferable CUDA over OpenCL. For the imagejdev.org group, the argument was
distilled to performance vs hardware scope - and having more hardware
coverage was a higher priority for our users.
Once a GPU implementation has been optimized for a specific hardware -
porting between CUDA and OpenCL is relatively easy. If your interested in
working with CUDA through Java, an example of the bindings can be found at
http://www.jcuda.de/ImageJ/ImageJHowTo.html. Further you may find
discussions like this one interesting in understanding the tradeoffs:
http://stackoverflow.com/questions/2633483/best-approach-for-gpgpu-cuda-opencl-in-java
.
Let me know if you have any problems running the updated float version of
sobel filter.
Sincerely,
Rick Lentz
On Fri, Jan 14, 2011 at 2:07 AM, Pol kennel <pol.kennel at gmail.com> wrote:
> Hi Rick,
>
> You perfectly understood what I need, this will lead to apply "heavy"
> texture features extraction on large image aiming segmentation (the method
> will be published soon).
> I hope to have time in future to create a clear plugin for IJ that could be
> shared with users and contribuate to IJ community :) So thanks for your
> help, I'll try your last tutorial today, with hopes to get started with
> OpenCL quick... I have a last (at least) question : are we constrained to
> write gpu code in Cl ? could we use directly java codes ?
>
> Best regards,
> Pol
>
> 2011/1/13 Rick Lentz <rwlentz at wisc.edu>
>
> Hi Pol,
>>
>> I am thinking that the material under http://www.imagejdev.org/OpenCLwas not going to be as useful for you as a getting started tutorial. I have
>> removed the publication related material from our website. I created a Getting
>> Started with OpenCL in ImageJ Tutorial<https://docs.google.com/document/d/12aiL7urwcve7Mi-YyuOtAVlTZ5P6outcay75YiQlqyQ/edit?hl=en> with
>> the goal of helping ImageJ users get started quickly with OpenCL.
>>
>> Back to your issues, for the research publication I compiled the jogamp
>> files - but for your application, I do not think you need to do this. The
>> compilation steps listed were specific to Ubuntu 9.10 and NVidia 3.1. If I
>> am understanding your ImageJ need correctly, I think you are looking to run
>> a few Java examples that help you get into writing OpenCL plugins for
>> ImageJ. I hope the attached draft copy of the tutorial will help you get
>> there quickly.
>>
>> Please let me know if you any troubles with the tutorial - and have any
>> feedback regarding the tutorial.
>>
>> Best Regards,
>>
>> Rick Lentz
>>
>>
>> Hi Pol,
>>
>> Can you do a fresh pull from our repository - You should not have to
>> build these unless you really want to.
>>
>> You can download the needed components directly from:
>> http://jogamp.org/deployment/webstart/
>>
>> I have also updated the ImageJ plugin demo source to include the most
>> recent Jogamp OpenCL binaries and native libs.
>>
>> Sincerely,
>>
>> Rick
>>
>> On Thu, Jan 13, 2011 at 9:55 AM, Pol kennel <pol.kennel at gmail.com> wrote:
>>
>>> In fact my *jocl* building also crash during Junit tests, with error
>>> stacks like :
>>>
>>> [junit] Testcase: createContextTest(com.jogamp.opencl.gl.CLGLTest): Caused
>>> an ERROR
>>> [junit] test timed out after 5000 milliseconds
>>> [junit] java.lang.Exception: test timed out after 5000 milliseconds
>>> [junit] at java.lang.Object.wait(Native Method)
>>> [junit] at java.lang.Object.wait(Object.java:485)
>>> [junit] at
>>> com.jogamp.opengl.impl.SharedResourceRunner.doAndWait(SharedResourceRunner.java:152)
>>> [junit] at
>>> com.jogamp.opengl.impl.SharedResourceRunner.getOrCreateShared(SharedResourceRunner.java:100)
>>> [junit] at
>>> com.jogamp.opengl.impl.x11.glx.X11GLXDrawableFactory.getOrCreateSharedContextImpl(X11GLXDrawableFactory.java:253)
>>> [junit] at
>>> javax.media.opengl.GLDrawableFactory.getOrCreateSharedContext(GLDrawableFactory.java:275)
>>> [junit] at
>>> javax.media.opengl.GLDrawableFactory.getIsSharedContextAvailable(GLDrawableFactory.java:250)
>>> [junit] at
>>> javax.media.opengl.GLProfile.initProfilesForDeviceImpl(GLProfile.java:1251)
>>> [junit] at
>>> javax.media.opengl.GLProfile.initProfilesForDevice(GLProfile.java:1224)
>>> [junit] at
>>> javax.media.opengl.GLProfile.initProfilesForDefaultDevices(GLProfile.java:1192)
>>> [junit] at
>>> javax.media.opengl.GLProfile.access$000(GLProfile.java:66)
>>> [junit] at javax.media.opengl.GLProfile$1.run(GLProfile.java:112)
>>> [junit] at java.security.AccessController.doPrivileged(Native
>>> Method)
>>> [junit] at
>>> javax.media.opengl.GLProfile.initSingleton(GLProfile.java:110)
>>> [junit] at com.jogamp.opencl.gl.CLGLTest.init(CLGLTest.java:70)
>>> [junit] at
>>> com.jogamp.opencl.gl.CLGLTest.createContextTest(CLGLTest.java:102)
>>>
>>> So there is another problem I don t identify...
>>>
>>> 2011/1/13 Pol kennel <pol.kennel at gmail.com>
>>>
>>> Hi Rick,
>>>>
>>>> Thank a lot for your detailed answers an to pay attention to my problems
>>>> !
>>>>
>>>> I finally resolved my problem with the OpenGL lib with your previous
>>>> mail : a link was not set correctly to the libGL.so file. I now have the
>>>> same output with *ldd /usr/lib/libGL.so *command. So I build correctly
>>>> C and OpenCL project, demos work well.
>>>>
>>>> I also build correctly *gluegen*, *jogl*, *jocl* and *jocl-demos*projects (note that lines "git clone
>>>> http://github.com/sgothel/gluegen.git gluegen" and "git clone
>>>> http://github.com/sgothel/jogl-demos.git jogl-demos" are missing on
>>>> the http://www.imagejdev.org/setting-jocl-jogl-and-gluegen page of your
>>>> tutorial).
>>>>
>>>> However, I stoped in building *joal* and *jogl-demos. *You probably
>>>> omit to precise the building of the *joal *project (but it could be
>>>> intended ?) ; when building *joal* with *ant* in the */joal/make/*directory I get an error :
>>>> "*/home/pol/joal/make/build.xml:369: taskdef class
>>>> com.sun.gluegen.ant.GlueGenTask cannot be found*
>>>> * using the classloader
>>>> AntClassLoader[/home/pol/gluegen/build/gluegen.jar:/home/pol/gluegen/build/antlr.jar:/home/pol/gluegen/make/lib/antlr.jar]
>>>> *"
>>>> (gluegen.jar and antlr.jar still in the right directory, is there
>>>> classpath to set for gluegen ?). In the *joal* readme file, I found
>>>> that OpenAL lib are need, so i downloaded libopenal-dev/openal1 packets
>>>> from synaptic, and follow instructions saying to copy gluegen.properties and
>>>> joal.properties and jogl.properties into home directory but don't know how
>>>> to set correctly properties into this files.
>>>> *joal* building stills crash.
>>>>
>>>> Regarding to *jogl-demos *when* *building with* ant *in *
>>>> ~/jogl-demos/make/* i get 27 errors like :
>>>> [javac] Compiling 169 source files to /home/pol/jogl-demos/build/classes
>>>> [javac] /home/pol/jogl-demos/src/demos/applets/GearsJOALApplet.java:11:
>>>> package com.jogamp.openal.util does not exist
>>>> [javac] import com.jogamp.openal.util.ALut;"
>>>> [...]
>>>> I suppose it s due to joal, OpenAL again....
>>>>
>>>> (How) Did you install it ?
>>>>
>>>>
>>>> On the other side, i have update the decon project from svn.
>>>> When running *SobelFilterExample.java *i get :
>>>>
>>>> *Retrieving test image... *
>>>> *Starting iteration... 0*
>>>> *Local work size dimensions are max array size of*
>>>> *unavailable functions: [clCreateEventFromGLsyncKHR,
>>>> clIcdGetPlatformIDsKHR]*
>>>> *Discovered NVIDIA CUDA*
>>>> *com.jogamp.opencl.CLException$CLInvalidBinaryException: *
>>>> *CLDevice [id: 140626625730048 name: GeForce GTS 240 type: GPU profile:
>>>> FULL_PROFILE] build log:*
>>>> *ptxas application ptx input, line 104; error : Instruction 'cvt'
>>>> requires SM 1.3 or higher, or map_f64_to_f32 directive*
>>>> *ptxas application ptx input, line 105; error : Instruction 'cvt'
>>>> requires SM 1.3 or higher, or map_f64_to_f32 directive*
>>>> *ptxas application ptx input, line 106; error : Instruction 'mul'
>>>> requires SM 1.3 or higher, or map_f64_to_f32 directive*
>>>> *ptxas application ptx input, line 107; error : Instruction 'mul'
>>>> requires SM 1.3 or higher, or map_f64_to_f32 directive*
>>>> *ptxas application ptx input, line 108; error : Instruction 'add'
>>>> requires SM 1.3 or higher, or map_f64_to_f32 directive*
>>>> *ptxas application ptx input, line 109; error : Instruction 'sqrt'
>>>> requires SM 1.3 or higher, or map_f64_to_f32 directive*
>>>> *ptxas application ptx input, line 110; error : Instruction 'cvt'
>>>> requires SM 1.3 or higher, or map_f64_to_f32 directive*
>>>> *ptxas fatal : Ptx assembly aborted due to errors*
>>>> *error : Ptx compilation failed: gpu='sm_11', device
>>>> code='cuModuleLoadDataEx_4'*
>>>> *: Considering profile 'compute_11' for gpu='sm_11' in
>>>> 'cuModuleLoadDataEx_4'*
>>>> *: Retrieving binary for 'cuModuleLoadDataEx_4', for gpu='sm_11', usage
>>>> mode=' '*
>>>> *: Considering profile 'compute_11' for gpu='sm_11' in
>>>> 'cuModuleLoadDataEx_4'*
>>>> *: Control flags for 'cuModuleLoadDataEx_4' disable search path*
>>>> *: Ptx binary found for 'cuModuleLoadDataEx_4',
>>>> architecture='compute_11'*
>>>> *: Ptx compilation for 'cuModuleLoadDataEx_4', for gpu='sm_11', ocg
>>>> options=' '*
>>>> *ptxas application ptx input, line 104; warning : Double is not
>>>> supported. Demoting to float*
>>>> *error: CL_INVALID_BINARY (man page:
>>>> http://www.khronos.org/opencl/sdk/1.1/docs/man/xhtml/errors.html)*
>>>> * at com.jogamp.opencl.CLException.newException(CLException.java:49)*
>>>> * at com.jogamp.opencl.CLProgram.build(CLProgram.java:335)*
>>>> * at com.jogamp.opencl.CLProgram.build(CLProgram.java:174)*
>>>> * at publication.SobelFilterExample.<init>(SobelFilterExample.java:66)*
>>>> * at
>>>> publication.SobelFilterExample.runTest(SobelFilterExample.java:204)*
>>>> * at publication.SobelFilterExample.main(SobelFilterExample.java:155)*
>>>>
>>>> So I suppose it s due to miss installation of JOAL.
>>>>
>>>>
>>>> I'll be so thankful if you can get me out of this deadlock.... I can
>>>> already see the light far away :)
>>>>
>>>> Best regards,
>>>>
>>>>
>>>> Pol
>>>>
>>>>
>>>>
>>>> PS : I'm often connected on google chat so if you prefer answer me
>>>> online please let me know you gmail address.
>>>>
>>>>
>>>>
>>>> 2011/1/12 Rick Lentz <rwlentz at wisc.edu>
>>>>
>>>> Hi Pol,
>>>>>
>>>>> Regarding your first question it is likely that the linker is missing
>>>>> a reference to OpenGL. My instructions were for a prior major release of
>>>>> Ubuntu as well as a prior minor release of NVidia's CUDA SDK. I have
>>>>> updated the setup instructions to reflect the current versions as well as
>>>>> tested against 64bit Mac and Linux OSs.
>>>>>
>>>>> With regard to your specific problem, many of NVidia's examples
>>>>> include use of OpenGL. NVidia's documentation indicates that libgl.so is
>>>>> referenced via static link to appear to be located in /usr/lib/libgl.so (per
>>>>>
>>>>> http://developer.download.nvidia.com/compute/cuda/3_2_prod/drivers/docs/README_Linux.txtunder Chapter 5, Listing of Installed Components, 4th bullet). Chapter 5 of
>>>>> NVidia's documentation goes on to describe the linking process that happens
>>>>> when installing the developer drivers. Perhaps this linking did not happen
>>>>> when you installed the NVidia development drivers for Linux. Did you get an
>>>>> error during installation of the developer drivers reporting something in
>>>>> this regard? Chapter 5 concludes on how to check a Linux dynamic library
>>>>> using the command line tool ldd. In this case, ldd /usr/lib/libGL.so
>>>>>
>>>>> Here is the output I get when running ldd /usr/lib/libGL.so
>>>>>
>>>>> ldd /usr/lib/libGL.so
>>>>> linux-vdso.so.1 => (0x00007ffff33e6000)
>>>>> libnvidia-tls.so.260.19.14 => /usr/lib/tls/libnvidia-tls.so.260.19.14
>>>>> (0x00007fc82b46b000)
>>>>> libnvidia-glcore.so.260.19.14 => /usr/lib/libnvidia-glcore.so.260.19.14
>>>>> (0x00007fc82987f000)
>>>>> libX11.so.6 => /usr/lib/libX11.so.6 (0x00007fc829549000)
>>>>> libXext.so.6 => /usr/lib/libXext.so.6 (0x00007fc829337000)
>>>>> libc.so.6 => /lib/libc.so.6 (0x00007fc828f93000)
>>>>> libdl.so.2 => /lib/libdl.so.2 (0x00007fc828d8f000)
>>>>> libm.so.6 => /lib/libm.so.6 (0x00007fc828b0c000)
>>>>> libxcb.so.1 => /usr/lib/libxcb.so.1 (0x00007fc8288ef000)
>>>>> /lib64/ld-linux-x86-64.so.2 (0x00007fc82b974000)
>>>>> libXau.so.6 => /usr/lib/libXau.so.6 (0x00007fc8286eb000)
>>>>> libXdmcp.so.6 => /usr/lib/libXdmcp.so.6 (0x00007fc8284e5000)
>>>>>
>>>>>
>>>>> If the libGL.so file is present and linked property, you can double
>>>>> check for inclusion of libGL.so path by ensuring its path in your bash
>>>>> profile under the variable LD_LIBRARY_PATH. NVidia asks for the installer
>>>>> to add these additions in the Linux install documents. Here is the line in
>>>>> my .bashrc file that allows the linker to find libGL.so (specifically the
>>>>> ":/usr/lib" portion of this line):
>>>>>
>>>>> export
>>>>> LD_LIBRARY_PATH="/usr/lib:/usr/local/cuda/lib64:/usr/local/cuda/lib:/usr/lib32:/usr/local/lib"
>>>>>
>>>>> I also had issues compiling the most recent version of NVidia's C
>>>>> samples. I reflected a work around in the updated web documents (see the
>>>>> bottom of: http://www.imagejdev.org/setting-host-machine ).
>>>>>
>>>>> Please let me know if you have any more questions or difficulties with
>>>>> the ImageJ OpenCL plugin examples.
>>>>>
>>>>> Sincerely,
>>>>>
>>>>> Rick Lentz
>>>>>
>>>>>
>>>>> On Tue, Jan 11, 2011 at 4:20 AM, Pol kennel <pol.kennel at gmail.com>wrote:
>>>>>
>>>>>> Hello,
>>>>>>
>>>>>> I am a PhD student working on image segmentation by texture analysis
>>>>>> with an application on remote sensing images (Montpellier, France). So i
>>>>>> have to process very large images (e.g. 15000*8000).
>>>>>> All methods I developed are formed as plugin under ImageJ API that I
>>>>>> really like, but not yet applicable on this sort of image.
>>>>>> Recently, I tough to use finally my graphic card (Nvidia GTS240) which
>>>>>> I think will help me a lot in my process. So after several search on the
>>>>>> web, I found this <http://imagejdev.org/plugins/opencl-plugin> (OpenCL
>>>>>> plugin you wrote) great ! it's precisely what i need.
>>>>>>
>>>>>> So I decided to follow your tutorial to setup the environment but met
>>>>>> some problems... :
>>>>>>
>>>>>> 1. When launching makefile in "~/NVIDIA_GPU_Computing_SDK/C/ " and
>>>>>> "~/NVIDIA_GPU_Computing_SDK/OpenCL" I had following error "/usr/bin/ld:
>>>>>> cannot find -lGL". Plus, I dont have the same demo files, maybe because of
>>>>>> version ? (cudatoolkit_3.2.16_linux_64, gpucomputingsdk_3.2.16 and
>>>>>> devdriver_3.2_linux_64_260.19.26 on Ubuntu 10.10). On the other side demos I
>>>>>> have work perfectly (ooclDCT8x8, oclHistogram,...)
>>>>>> 2. *Main problem is that the repository you indicate to get the
>>>>>> example project (http://www.loci.wisc.edu/svn/decon) doesn't
>>>>>> work... *"
>>>>>> http://dev.loci.wisc.edu/svn/software/branches/maven/projects/opencl-decon"
>>>>>> repository works well. Is is the same ?
>>>>>>
>>>>>>
>>>>>> Thanks a lot if you find time to answer me (and sorry for my english).
>>>>>>
>>>>>> And thanks you for work you are doing on ImageJ.
>>>>>>
>>>>>> Best regards,
>>>>>> *
>>>>>> *
>>>>>> --
>>>>>> Pol Kennel
>>>>>>
>>>>>>
>>>>>
>>>>
>>>>
>>>> --
>>>> Pol Kennel
>>>>
>>>>
>>>
>>>
>>> --
>>> Pol Kennel
>>>
>>>
>>
>>
>>
>>
>>
> --
> Pol Kennel
>
>
-------------- next part --------------
An HTML attachment was scrubbed...
URL: <http://imagej.net/pipermail/imagej-devel/attachments/20110114/f5b4105c/attachment.html>
More information about the ImageJ-devel
mailing list