Added an experimental Jocl binding.

Note that Jogamp's Jocl only supports OpenCL1.1, some methods will throw an UnsupportedOperationException.

jme3-examples/src/main/java/jme3test/opencl/HelloOpenCL.java

@@ -58,6 +58,7 @@ public class HelloOpenCL extends SimpleApplication {
         AppSettings settings = new AppSettings(true);
         settings.setOpenCLSupport(true);
         settings.setVSync(true);
+//        settings.setRenderer(AppSettings.JOGL_OPENGL_FORWARD_COMPATIBLE);
         app.setSettings(settings);
         app.start(); // start the game
     }
@@ -208,7 +209,11 @@ public class HelloOpenCL extends SimpleApplication {
             //query supported formats
             for (MemoryAccess ma : MemoryAccess.values()) {
                 for (Image.ImageType type : Image.ImageType.values()) {
-                    System.out.println("Formats for "+ma+" and "+type+": "+Arrays.toString(clContext.querySupportedFormats(ma, type)));
+                    try {
+                        System.out.println("Formats for " + ma + " and " + type + ": " + Arrays.toString(clContext.querySupportedFormats(ma, type)));
+                    } catch (UnsupportedOperationException e) {
+                        LOG.warning(e.getLocalizedMessage());
+                    }
                 }
             }
             

jme3-examples/src/main/java/jme3test/opencl/TestVertexBufferSharing.java

@@ -64,6 +64,7 @@ public class TestVertexBufferSharing extends SimpleApplication {
         AppSettings settings = new AppSettings(true);
         settings.setOpenCLSupport(true);
         settings.setVSync(false);
+//        settings.setRenderer(AppSettings.JOGL_OPENGL_FORWARD_COMPATIBLE);
         app.setSettings(settings);
         app.start(); // start the game
     }

jme3-examples/src/main/java/jme3test/opencl/TestWriteToTexture.java

@@ -63,13 +63,13 @@ public class TestWriteToTexture extends SimpleApplication implements AnalogListe
     private Vector2f C;
     private Image texCL;
     private boolean dragging;
-    private int gcCounter;
 
     public static void main(String[] args){
         TestWriteToTexture app = new TestWriteToTexture();
         AppSettings settings = new AppSettings(true);
         settings.setOpenCLSupport(true);
         settings.setVSync(false);
+//        settings.setRenderer(AppSettings.JOGL_OPENGL_FORWARD_COMPATIBLE);
         app.setSettings(settings);
         app.start(); // start the game
     }
@@ -87,7 +87,6 @@ public class TestWriteToTexture extends SimpleApplication implements AnalogListe
         guiNode.attachChild(pic);
         
         initCounter = 0;
-        gcCounter = 0;
         
         flyCam.setEnabled(false);
         inputManager.setCursorVisible(true);

jme3-jogl/build.gradle

@@ -8,4 +8,5 @@ dependencies {
     compile 'org.jogamp.gluegen:gluegen-rt-main:2.3.2'
     compile 'org.jogamp.jogl:jogl-all-main:2.3.2'
     compile 'org.jogamp.joal:joal-main:2.3.2'
+	compile 'org.jogamp.jocl:jocl-main:2.3.2'
 }

jme3-jogl/src/main/java/com/jme3/opencl/jocl/JoclBuffer.java

@@ -0,0 +1,225 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.opencl.*;
+import java.nio.ByteBuffer;
+import com.jogamp.opencl.*;
+import com.jogamp.opencl.llb.CL;
+import com.jogamp.opencl.llb.gl.CLGL;
+import java.util.EnumSet;
+
+/**
+ *
+ * @author shaman
+ */
+public class JoclBuffer extends Buffer {
+
+    final long id;
+    final CL cl;
+
+    public JoclBuffer(long id) {
+        this.id = id;
+        this.cl = CLPlatform.getLowLevelCLInterface();
+        OpenCLObjectManager.getInstance().registerObject(this);
+    }
+    
+    @Override
+    public long getSize() {
+        Utils.pointers[0].rewind();
+        int ret = cl.clGetMemObjectInfo(id, CL.CL_MEM_SIZE, Utils.pointers[0].elementSize(), Utils.pointers[0].getBuffer(), null);
+        Utils.checkError(ret, "clGetMemObjectInfo");
+        return Utils.pointers[0].get();
+    }
+
+    @Override
+    public MemoryAccess getMemoryAccessFlags() {
+        Utils.pointers[0].rewind();
+        int ret = cl.clGetMemObjectInfo(id, CL.CL_MEM_TYPE, Utils.pointers[0].elementSize(), Utils.pointers[0].getBuffer(), null);
+        Utils.checkError(ret, "clGetMemObjectInfo");
+        long flags = Utils.pointers[0].get();
+        return Utils.getMemoryAccessFromFlag(flags);
+    }
+
+    @Override
+    public void read(CommandQueue queue, ByteBuffer dest, long size, long offset) {
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueReadBuffer(q, id, CL.CL_TRUE, offset, size, dest, 0, null, null);
+        Utils.checkError(ret, "clEnqueueReadBuffer");
+    }
+
+    @Override
+    public Event readAsync(CommandQueue queue, ByteBuffer dest, long size, long offset) {
+        Utils.pointers[0].rewind();
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueReadBuffer(q, id, CL.CL_FALSE, offset, size, dest, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueReadBuffer");
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public void write(CommandQueue queue, ByteBuffer src, long size, long offset) {
+        long q = ((JoclCommandQueue)queue).id;
+        int ret = cl.clEnqueueWriteBuffer(q, id, CL.CL_TRUE, offset, size, src, 0, null, null);
+        Utils.checkError(ret, "clEnqueueWriteBuffer");
+    }
+
+    @Override
+    public Event writeAsync(CommandQueue queue, ByteBuffer src, long size, long offset) {
+        Utils.pointers[0].rewind();
+        long q = ((JoclCommandQueue)queue).id;
+        int ret = cl.clEnqueueWriteBuffer(q, id, CL.CL_FALSE, offset, size, src, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueWriteBuffer");
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public void copyTo(CommandQueue queue, Buffer dest, long size, long srcOffset, long destOffset) {
+        Utils.pointers[0].rewind();
+        long q = ((JoclCommandQueue)queue).id;
+        long did  = ((JoclBuffer) dest).id;
+        int ret = cl.clEnqueueCopyBuffer(q, id, did, srcOffset, destOffset, size, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueCopyBuffer");
+        ret = cl.clWaitForEvents(1, Utils.pointers[0]);
+        Utils.checkError(ret, "clWaitForEvents");
+    }
+
+    @Override
+    public Event copyToAsync(CommandQueue queue, Buffer dest, long size, long srcOffset, long destOffset) {
+        Utils.pointers[0].rewind();
+        long q = ((JoclCommandQueue)queue).id;
+        long did  = ((JoclBuffer) dest).id;
+        int ret = cl.clEnqueueCopyBuffer(q, id, did, srcOffset, destOffset, size, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueCopyBuffer");
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public ByteBuffer map(CommandQueue queue, long size, long offset, MappingAccess access) {
+        long q = ((JoclCommandQueue)queue).id;
+        Utils.errorBuffer.rewind();
+        long flags = Utils.getMappingAccessFlags(access);
+        ByteBuffer b = cl.clEnqueueMapBuffer(q, id, CL.CL_TRUE, flags, offset, size, 0, null, null, Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clEnqueueMapBuffer");
+        return b;
+    }
+
+    @Override
+    public void unmap(CommandQueue queue, ByteBuffer ptr) {
+        long q = ((JoclCommandQueue)queue).id;
+        Utils.pointers[0].rewind();
+        int ret = cl.clEnqueueUnmapMemObject(q, id, ptr, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueUnmapMemObject");
+        ret = cl.clWaitForEvents(1, Utils.pointers[0]);
+        Utils.checkError(ret, "clWaitForEvents");
+    }
+
+    @Override
+    public com.jme3.opencl.Buffer.AsyncMapping mapAsync(CommandQueue queue, long size, long offset, MappingAccess access) {
+        long q = ((JoclCommandQueue)queue).id;
+        Utils.pointers[0].rewind();
+        Utils.errorBuffer.rewind();
+        long flags = Utils.getMappingAccessFlags(access);
+        ByteBuffer b = cl.clEnqueueMapBuffer(q, id, CL.CL_FALSE, flags, offset, size, 0, null, Utils.pointers[0], Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clEnqueueMapBuffer");
+        long event = Utils.pointers[0].get(0);
+        return new com.jme3.opencl.Buffer.AsyncMapping(new JoclEvent(event), b);
+    }
+
+    @Override
+    public Event fillAsync(CommandQueue queue, ByteBuffer pattern, long size, long offset) {
+        throw new UnsupportedOperationException("Not supported by Jocl!");
+    }
+
+    @Override
+    public Event copyToImageAsync(CommandQueue queue, Image dest, long srcOffset, long[] destOrigin, long[] destRegion) {
+        if (destOrigin.length!=3 || destRegion.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[1].put(destOrigin[0]).put(destOrigin[1]).put(destOrigin[2]).position(0);
+        Utils.pointers[2].put(destRegion[0]).put(destRegion[1]).put(destRegion[2]).position(0);
+        long q = ((JoclCommandQueue)queue).id;
+        long i = ((JoclImage) dest).id;
+        int ret = cl.clEnqueueCopyBufferToImage(q, id, i, srcOffset, Utils.pointers[1], Utils.pointers[2], 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueCopyBufferToImage");
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public Event acquireBufferForSharingAsync(CommandQueue queue) {
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[1].put(0, id);
+        long q = ((JoclCommandQueue)queue).id;
+        ((CLGL) cl).clEnqueueAcquireGLObjects(q, 1, Utils.pointers[1], 0, null, Utils.pointers[0]);
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public Event releaseBufferForSharingAsync(CommandQueue queue) {
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[1].put(0, id);
+        long q = ((JoclCommandQueue)queue).id;
+        ((CLGL) cl).clEnqueueReleaseGLObjects(q, 1, Utils.pointers[1], 0, null, Utils.pointers[0]);
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public ObjectReleaser getReleaser() {
+        return new ReleaserImpl(id);
+    }
+    private static class ReleaserImpl implements ObjectReleaser {
+        private long mem;
+        private ReleaserImpl(long mem) {
+            this.mem = mem;
+        }
+        @Override
+        public void release() {
+            if (mem != 0) {
+                int ret = CLPlatform.getLowLevelCLInterface().clReleaseMemObject(mem);
+                mem = 0;
+                Utils.reportError(ret, "clReleaseMemObject");
+            }
+        }
+        
+    }
+}

jme3-jogl/src/main/java/com/jme3/opencl/jocl/JoclCommandQueue.java

@@ -0,0 +1,90 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.opencl.CommandQueue;
+import com.jme3.opencl.OpenCLObjectManager;
+import com.jogamp.opencl.CLCommandQueue;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.llb.CL;
+import com.jogamp.opencl.llb.CLCommandQueueBinding;
+
+/**
+ *
+ * @author shaman
+ */
+public class JoclCommandQueue implements CommandQueue {
+
+    final CL cl;
+    final long id;
+
+    public JoclCommandQueue(long id) {
+        this.id = id;
+        this.cl = CLPlatform.getLowLevelCLInterface();
+        OpenCLObjectManager.getInstance().registerObject(this);
+    }
+    
+    @Override
+    public void flush() {  
+        int ret = cl.clFlush(id);
+        Utils.checkError(ret, "clFlush");
+    }
+
+    @Override
+    public void finish() {
+        int ret = cl.clFinish(id);
+        Utils.checkError(ret, "clFinish");
+    }
+    
+    @Override
+    public ObjectReleaser getReleaser() {
+        return new ReleaserImpl(id, cl);
+    }
+    private static class ReleaserImpl implements ObjectReleaser {
+        private long id;
+        private CLCommandQueueBinding cl;
+
+        private ReleaserImpl(long id, CLCommandQueueBinding cl) {
+            this.id = id;
+            this.cl = cl;
+        }
+        
+        @Override
+        public void release() {
+            if (id != 0) {
+                int ret = cl.clReleaseCommandQueue(id);
+                id = 0;
+                Utils.reportError(ret, "clReleaseCommandQueue");
+            }
+        }
+    }
+}

jme3-jogl/src/main/java/com/jme3/opencl/jocl/JoclContext.java

@@ -0,0 +1,249 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.opencl.*;
+import com.jme3.opencl.Context;
+import com.jme3.opencl.Image.ImageDescriptor;
+import com.jme3.opencl.Image.ImageFormat;
+import com.jme3.scene.VertexBuffer;
+import com.jme3.texture.FrameBuffer;
+import com.jme3.texture.Texture;
+import com.jogamp.opencl.CLContext;
+import com.jogamp.opencl.CLImageFormat;
+import com.jogamp.opencl.CLMemory;
+import com.jogamp.opencl.CLMemory.Mem;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.llb.CL;
+import com.jogamp.opencl.llb.gl.CLGL;
+import com.jogamp.opencl.llb.impl.CLImageFormatImpl;
+import com.jogamp.opengl.GL;
+import java.nio.ByteBuffer;
+import java.nio.IntBuffer;
+import java.util.List;
+import java.util.logging.Level;
+import java.util.logging.Logger;
+
+/**
+ *
+ * @author shaman
+ */
+public class JoclContext extends Context {
+    private static final Logger LOG = Logger.getLogger(JoclContext.class.getName());
+    
+    final CLContext context;
+    final long id;
+    final CL cl; 
+    private final List<JoclDevice> devices;
+
+    public JoclContext(CLContext context, List<JoclDevice> devices) {
+        this.context = context;
+        this.id = context.ID;
+        this.cl = context.getCL();
+        this.devices = devices;
+        OpenCLObjectManager.getInstance().registerObject(this);
+    }
+
+    public CLContext getContext() {
+        return context;
+    }
+
+    @Override
+    public List<JoclDevice> getDevices() {
+        return devices;
+    }
+
+    @Override
+    @SuppressWarnings("element-type-mismatch")
+    public CommandQueue createQueue(Device device) {
+        assert (devices.contains(device)); //this also ensures that device is a JoclDevice
+        long d = ((JoclDevice) device).id;
+        long properties = 0;
+        long q = cl.clCreateCommandQueue(id, d, properties, Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clCreateCommandQueue");
+        return new JoclCommandQueue(q);
+    }
+    
+    @Override
+    public Buffer createBuffer(long size, MemoryAccess access) {
+        long flags = Utils.getMemoryAccessFlags(access);
+        long mem = cl.clCreateBuffer(id, flags, size, null, Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clCreateBuffer");
+        return new JoclBuffer(mem);
+    }
+
+    @Override
+    public Buffer createBufferFromHost(ByteBuffer data, MemoryAccess access) {
+        long flags = Utils.getMemoryAccessFlags(access);
+        flags |= CL.CL_MEM_USE_HOST_PTR;
+        long mem = cl.clCreateBuffer(id, flags, data.capacity(), data, Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clCreateBuffer");
+        return new JoclBuffer(mem);
+    }
+
+    @Override
+    public Image createImage(MemoryAccess access, ImageFormat format, ImageDescriptor descr) {
+        if (descr.type != Image.ImageType.IMAGE_2D && descr.type != Image.ImageType.IMAGE_3D) {
+            throw new UnsupportedOperationException("Jocl only supports 2D and 3D images");
+        }
+        long memFlags = Utils.getMemoryAccessFlags(access);
+        Utils.errorBuffer.rewind();
+        //fill image format
+        CLImageFormatImpl f = CLImageFormatImpl.create();
+        f.setImageChannelOrder(JoclImage.decodeImageChannelOrder(format.channelOrder));
+        f.setImageChannelDataType(JoclImage.decodeImageChannelType(format.channelType));
+        //create image
+        long mem;
+        if (descr.type == Image.ImageType.IMAGE_2D) {
+            mem = cl.clCreateImage2D(id, memFlags, f, descr.width, descr.height, 
+                    descr.hostPtr==null ? 0 : descr.rowPitch, descr.hostPtr, Utils.errorBuffer);
+            Utils.checkError(Utils.errorBuffer, "clCreateImage2D");
+        } else {
+            mem = cl.clCreateImage3D(id, memFlags, f, descr.width, descr.height, descr.depth, 
+                    descr.hostPtr==null ? 0 : descr.rowPitch, descr.hostPtr==null ? 0 : descr.slicePitch, 
+                    descr.hostPtr, Utils.errorBuffer);
+            Utils.checkError(Utils.errorBuffer, "clCreateImage3D");
+        }
+        return new JoclImage(mem);
+    }
+
+    @Override
+    public ImageFormat[] querySupportedFormats(MemoryAccess access, Image.ImageType type) {
+        if (type != Image.ImageType.IMAGE_2D && type != Image.ImageType.IMAGE_3D) {
+            throw new UnsupportedOperationException("Jocl only supports 2D and 3D images");
+        }
+        long memFlags = Utils.getMemoryAccessFlags(access);
+        CLImageFormat[] fx;
+        if (type == Image.ImageType.IMAGE_2D) {
+            fx = context.getSupportedImage2dFormats(Mem.valueOf((int) memFlags));
+        }  else {
+            fx = context.getSupportedImage3dFormats(Mem.valueOf((int) memFlags));
+        }
+        //convert formats
+        ImageFormat[] formats = new ImageFormat[fx.length];
+        for (int i=0; i<fx.length; ++i) {
+            Image.ImageChannelOrder channelOrder = JoclImage.encodeImageChannelOrder(fx[i].getFormatImpl().getImageChannelOrder());
+            Image.ImageChannelType channelType = JoclImage.encodeImageChannelType(fx[i].getFormatImpl().getImageChannelDataType());
+            formats[i] = new ImageFormat(channelOrder, channelType);
+        }
+        return formats;
+    }
+
+    @Override
+    public Buffer bindVertexBuffer(VertexBuffer vb, MemoryAccess access) {
+        int vbId = vb.getId();
+        if (vbId == -1) {
+            throw new IllegalArgumentException("vertex buffer was not yet uploaded to the GPU or is CPU only");
+        }
+        long flags = Utils.getMemoryAccessFlags(access);
+        Utils.errorBuffer.rewind();
+        long mem = ((CLGL) cl).clCreateFromGLBuffer(id, flags, vbId, Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clCreateFromGLBuffer");
+        return new JoclBuffer(mem);
+    }
+
+    @Override
+    public Image bindImage(com.jme3.texture.Image image, Texture.Type textureType, int miplevel, MemoryAccess access) {
+        int imageID = image.getId();
+        if (imageID == -1) {
+            throw new IllegalArgumentException("image was not yet uploaded to the GPU");
+        }
+        long memFlags = Utils.getMemoryAccessFlags(access);
+        int textureTarget = convertTextureType(textureType);
+        Utils.errorBuffer.rewind();
+        long mem;
+        if (textureType == Texture.Type.TwoDimensional) {
+            mem = ((CLGL) cl).clCreateFromGLTexture2D(id, memFlags, textureTarget, miplevel, imageID, Utils.errorBuffer);
+        } else if (textureType == Texture.Type.ThreeDimensional) {
+            mem = ((CLGL) cl).clCreateFromGLTexture3D(id, memFlags, textureTarget, miplevel, imageID, Utils.errorBuffer);
+        } else {
+            throw new UnsupportedOperationException("Jocl only supports 2D and 3D images");
+        }
+        Utils.checkError(Utils.errorBuffer, "clCreateFromGLTexture");
+        return new JoclImage(mem);
+    }
+
+    @Override
+    protected Image bindPureRenderBuffer(FrameBuffer.RenderBuffer buffer, MemoryAccess access) {
+        int renderbuffer = buffer.getId();
+        if (renderbuffer == -1) {
+            throw new IllegalArgumentException("renderbuffer was not yet uploaded to the GPU");
+        }
+        long memFlags = Utils.getMemoryAccessFlags(access);
+        Utils.errorBuffer.rewind();
+        long mem = ((CLGL) cl).clCreateFromGLRenderbuffer(id, memFlags, renderbuffer, Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clCreateFromGLRenderbuffer");
+        return new JoclImage(mem);
+    }
+    
+    private int convertTextureType(Texture.Type textureType) {
+        switch (textureType) {
+            case TwoDimensional: return GL.GL_TEXTURE_2D;
+            case CubeMap: return GL.GL_TEXTURE_CUBE_MAP;
+            default: throw new IllegalArgumentException("unknown or unsupported texture type "+textureType);
+        }
+    }
+
+    @Override
+    public Program createProgramFromSourceCode(String sourceCode) {
+        LOG.log(Level.FINE, "Create program from source:\n{0}", sourceCode);
+        Utils.errorBuffer.rewind();
+        Utils.pointers[0].rewind();
+        Utils.pointers[0].put(0, sourceCode.length());
+        long p = cl.clCreateProgramWithSource(id, 1, new String[]{sourceCode}, Utils.pointers[0], Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clCreateProgramWithSource");
+        return new JoclProgram(p, this);
+    }
+
+    @Override
+    public ObjectReleaser getReleaser() {
+        return new ReleaserImpl(id, devices);
+    }
+    private static class ReleaserImpl implements ObjectReleaser {
+        private long id;
+        private final List<JoclDevice> devices;
+        private ReleaserImpl(long id, List<JoclDevice> devices) {
+            this.id = id;
+            this.devices = devices;
+        }
+        @Override
+        public void release() {
+            if (id != 0) {
+                int ret = CLPlatform.getLowLevelCLInterface().clReleaseContext(id);
+                id = 0;
+                devices.clear();
+                Utils.reportError(ret, "clReleaseContext");
+            }
+        }
+        
+    }
+}

jme3-jogl/src/main/java/com/jme3/opencl/jocl/JoclDevice.java

@@ -0,0 +1,302 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.opencl.Device;
+import com.jme3.opencl.Platform;
+import com.jogamp.opencl.CLDevice;
+import java.util.Arrays;
+import java.util.Collection;
+
+/**
+ *
+ * @author shaman
+ */
+public final class JoclDevice implements Device {
+
+    final long id;
+    final CLDevice device;
+    final JoclPlatform platform;
+
+    public JoclDevice(CLDevice device, JoclPlatform platform) {
+        this.id = device.ID;
+        this.device = device;
+        this.platform = platform;
+    }
+
+    public long getId() {
+        return id;
+    }
+
+    public CLDevice getDevice() {
+        return device;
+    }
+    
+    @Override
+    public JoclPlatform getPlatform() {
+        return platform;
+    }
+
+    @Override
+    public DeviceType getDeviceType() {
+        CLDevice.Type type = device.getType();
+        switch (type) {
+            case ACCELERATOR: return DeviceType.ACCELEARTOR;
+            case CPU: return DeviceType.CPU;
+            case GPU: return DeviceType.GPU;
+            default: return DeviceType.DEFAULT;
+        }
+    }
+
+    @Override
+    public int getVendorId() {
+        return (int) device.getVendorID();
+    }
+
+    @Override
+    public boolean isAvailable() {
+        return device.isAvailable();
+    }
+
+    @Override
+    public boolean hasCompiler() {
+        return device.isCompilerAvailable();
+    }
+
+    @Override
+    public boolean hasDouble() {
+        return hasExtension("cl_khr_fp64");
+    }
+
+    @Override
+    public boolean hasHalfFloat() {
+        return hasExtension("cl_khr_fp16");
+    }
+
+    @Override
+    public boolean hasErrorCorrectingMemory() {
+        return device.isErrorCorrectionSupported();
+    }
+
+    @Override
+    public boolean hasUnifiedMemory() {
+        return device.isMemoryUnified();
+    }
+
+    @Override
+    public boolean hasImageSupport() {
+        return device.isImageSupportAvailable();
+    }
+    
+    @Override
+    public boolean hasWritableImage3D() {
+        return hasExtension("cl_khr_3d_image_writes");
+    }
+
+    @Override
+    public boolean hasOpenGLInterop() {
+        return hasExtension("cl_khr_gl_sharing");
+    }
+    
+    @Override
+    public boolean hasExtension(String extension) {
+        return getExtensions().contains(extension);
+    }
+
+    @Override
+    public Collection<? extends String> getExtensions() {
+        return device.getExtensions();
+    }
+
+    @Override
+    public int getComputeUnits() {
+        return device.getMaxComputeUnits();
+    }
+
+    @Override
+    public int getClockFrequency() {
+        return device.getMaxClockFrequency();
+    }
+
+    @Override
+    public int getAddressBits() {
+        return device.getAddressBits();
+    }
+
+    @Override
+    public boolean isLittleEndian() {
+        return device.isLittleEndian();
+    }
+
+    @Override
+    public long getMaximumWorkItemDimensions() {
+        return device.getMaxWorkItemDimensions();
+    }
+
+    @Override
+    public long[] getMaximumWorkItemSizes() {
+        int[] sizes = device.getMaxWorkItemSizes();
+        long[] s = new long[sizes.length];
+        for (int i=0; i<sizes.length; ++i) {
+            s[i] = sizes[i];
+        }
+        return s;
+    }
+
+    @Override
+    public long getMaxiumWorkItemsPerGroup() {
+        return device.getMaxWorkGroupSize();
+    }
+
+    @Override
+    public int getMaximumSamplers() {
+        return device.getMaxSamplers();
+    }
+
+    @Override
+    public int getMaximumReadImages() {
+        return device.getMaxReadImageArgs();
+    }
+
+    @Override
+    public int getMaximumWriteImages() {
+        return device.getMaxWriteImageArgs();
+    }
+
+    @Override
+    public long[] getMaximumImage2DSize() {
+        return new long[] {
+            device.getMaxImage2dWidth(),
+            device.getMaxImage2dHeight()
+        };
+    }
+
+    @Override
+    public long[] getMaximumImage3DSize() {
+        return new long[] {
+            device.getMaxImage3dWidth(),
+            device.getMaxImage3dHeight(),
+            device.getMaxImage3dDepth()
+        };
+    }
+    
+    @Override
+    public long getMaximumAllocationSize() {
+        return device.getMaxMemAllocSize();
+    }
+    
+    @Override
+    public long getGlobalMemorySize() {
+        return device.getGlobalMemSize();
+    }
+    
+    @Override
+    public long getLocalMemorySize() {
+        return device.getLocalMemSize();
+    }
+    
+    @Override
+    public long getMaximumConstantBufferSize() {
+        return device.getMaxConstantBufferSize();
+    }
+    
+    @Override
+    public int getMaximumConstantArguments() {
+        return (int) device.getMaxConstantArgs();
+    }
+
+    @Override
+    public String getProfile() {
+        return device.getProfile();
+    }
+
+    @Override
+    public String getVersion() {
+        return device.getVersion().toString();
+    }
+
+    @Override
+    public int getVersionMajor() {
+        return Utils.getMajorVersion(getVersion(), "OpenCL ");
+    }
+
+    @Override
+    public int getVersionMinor() {
+        return Utils.getMinorVersion(getVersion(), "OpenCL ");
+    }
+
+    @Override
+    public String getCompilerVersion() {
+        return "OpenCL C 1.1"; //at most OpenCL 1.1 is supported at all
+    }
+
+    @Override
+    public int getCompilerVersionMajor() {
+        return Utils.getMajorVersion(getCompilerVersion(), "OpenCL C ");
+    }
+
+    @Override
+    public int getCompilerVersionMinor() {
+        return Utils.getMinorVersion(getCompilerVersion(), "OpenCL C ");
+    }
+
+    @Override
+    public String getDriverVersion() {
+        return device.getDriverVersion();
+    }
+
+    @Override
+    public int getDriverVersionMajor() {
+        return Utils.getMajorVersion(getDriverVersion(), "");
+    }
+
+    @Override
+    public int getDriverVersionMinor() {
+        return Utils.getMinorVersion(getDriverVersion(), "");
+    }
+
+    @Override
+    public String getName() {
+        return device.getName();
+    }
+
+    @Override
+    public String getVendor() {
+        return device.getVendor();
+    }
+
+    @Override
+    public String toString() {
+        return getName();
+    }
+    
+}

jme3-jogl/src/main/java/com/jme3/opencl/jocl/JoclEvent.java

@@ -0,0 +1,103 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.opencl.Event;
+import com.jme3.opencl.OpenCLObjectManager;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.llb.CL;
+import java.util.logging.Logger;
+
+/**
+ *
+ * @author shaman
+ */
+public class JoclEvent implements Event {
+    private static final Logger LOG = Logger.getLogger(JoclEvent.class.getName());
+    
+    final long id;
+    final CL cl;
+
+    public JoclEvent(long id) {
+        this.id = id;
+        this.cl = CLPlatform.getLowLevelCLInterface();
+        OpenCLObjectManager.getInstance().registerObject(this);
+    }
+
+    @Override
+    public void waitForFinished() {
+        Utils.pointers[0].rewind();
+        Utils.pointers[0].put(0, id);
+        int ret = cl.clWaitForEvents(1, Utils.pointers[0]);
+        Utils.checkError(ret, "clWaitForEvents");
+    }
+
+    @Override
+    public boolean isCompleted() {
+        Utils.tempBuffers[0].b16.rewind();
+        int err = cl.clGetEventInfo(id, CL.CL_EVENT_COMMAND_EXECUTION_STATUS, 4, Utils.tempBuffers[0].b16, null);
+        Utils.checkError(err, "clGetEventInfo");
+        int status = Utils.tempBuffers[0].b16i.get(0);
+        if (status == CL.CL_SUCCESS) {
+            return true;
+        } else if (status < 0) {
+            Utils.checkError(status, "EventStatus");
+            return false;
+        } else {
+            return false;
+        }
+    }
+
+    @Override
+    public ObjectReleaser getReleaser() {
+        return new ReleaserImpl(id);
+    }
+    
+    private static class ReleaserImpl implements ObjectReleaser {
+        private long event;
+
+        private ReleaserImpl(long event) {
+            this.event = event;
+        }
+        
+        @Override
+        public void release() {
+            if (event != 0) {
+                int ret = CLPlatform.getLowLevelCLInterface().clReleaseEvent(event);
+                event = 0;
+                Utils.reportError(ret, "clReleaseEvent");
+                LOG.finer("Event deleted");
+            }
+        }
+        
+    }
+}

jme3-jogl/src/main/java/com/jme3/opencl/jocl/JoclImage.java

@@ -0,0 +1,533 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.math.ColorRGBA;
+import com.jme3.opencl.*;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.llb.CL;
+import com.jogamp.opencl.llb.gl.CLGL;
+import java.nio.ByteBuffer;
+import java.util.logging.Level;
+import java.util.logging.Logger;
+
+/**
+ *
+ * @author shaman
+ */
+public class JoclImage implements Image {
+    private static final Logger LOG = Logger.getLogger(JoclImage.class.getName());
+
+    final long id;
+    final CL cl;
+
+    public JoclImage(long image) {
+        this.id = image;
+        this.cl = CLPlatform.getLowLevelCLInterface();
+        OpenCLObjectManager.getInstance().registerObject(this);
+    }
+
+    public static int decodeImageChannelOrder(ImageChannelOrder order) {
+        switch (order) {
+            case A:
+                return CL.CL_A;
+            case ARGB:
+                return CL.CL_ARGB;
+            case BGRA:
+                return CL.CL_BGRA;
+            case INTENSITY:
+                return CL.CL_INTENSITY;
+            case LUMINANCE:
+                return CL.CL_LUMINANCE;
+            case R:
+                return CL.CL_R;
+            case RA:
+                return CL.CL_RA;
+            case RG:
+                return CL.CL_RG;
+            case RGB:
+                return CL.CL_RGB;
+            case RGBA:
+                return CL.CL_RGBA;
+            case RGBx:
+                return CL.CL_RGBx;
+            case RGx:
+                return CL.CL_RGx;
+            case Rx:
+                return CL.CL_Rx;
+            default:
+                throw new IllegalArgumentException("unknown image channel order: " + order);
+        }
+    }
+
+    public static ImageChannelOrder encodeImageChannelOrder(int order) {
+        switch (order) {
+            case CL.CL_A:
+                return ImageChannelOrder.A;
+            case CL.CL_ARGB:
+                return ImageChannelOrder.ARGB;
+            case CL.CL_BGRA:
+                return ImageChannelOrder.BGRA;
+            case CL.CL_INTENSITY:
+                return ImageChannelOrder.INTENSITY;
+            case CL.CL_LUMINANCE:
+                return ImageChannelOrder.LUMINANCE;
+            case CL.CL_R:
+                return ImageChannelOrder.R;
+            case CL.CL_RA:
+                return ImageChannelOrder.RA;
+            case CL.CL_RG:
+                return ImageChannelOrder.RG;
+            case CL.CL_RGB:
+                return ImageChannelOrder.RGB;
+            case CL.CL_RGBA:
+                return ImageChannelOrder.RGBA;
+            case CL.CL_RGBx:
+                return ImageChannelOrder.RGBx;
+            case CL.CL_RGx:
+                return ImageChannelOrder.RGx;
+            case CL.CL_Rx:
+                return ImageChannelOrder.Rx;
+            default:
+                //throw new com.jme3.opencl.OpenCLException("unknown image channel order id: " + order);
+                LOG.log(Level.WARNING, "Unknown image channel order id: {0}", order);
+                return null;
+        }
+    }
+
+    public static int decodeImageChannelType(ImageChannelType type) {
+        switch (type) {
+            case FLOAT:
+                return CL.CL_FLOAT;
+            case HALF_FLOAT:
+                return CL.CL_HALF_FLOAT;
+            case SIGNED_INT16:
+                return CL.CL_SIGNED_INT16;
+            case SIGNED_INT32:
+                return CL.CL_SIGNED_INT32;
+            case SIGNED_INT8:
+                return CL.CL_SIGNED_INT8;
+            case SNORM_INT16:
+                return CL.CL_SNORM_INT16;
+            case SNORM_INT8:
+                return CL.CL_SNORM_INT8;
+            case UNORM_INT8:
+                return CL.CL_UNORM_INT8;
+            case UNORM_INT_101010:
+                return CL.CL_UNORM_INT_101010;
+            case UNORM_INT16:
+                return CL.CL_UNORM_INT16;
+            case UNORM_SHORT_565:
+                return CL.CL_UNORM_SHORT_565;
+            case UNORM_SHORT_555:
+                return CL.CL_UNORM_SHORT_555;
+            case UNSIGNED_INT16:
+                return CL.CL_UNSIGNED_INT16;
+            case UNSIGNED_INT32:
+                return CL.CL_UNSIGNED_INT32;
+            case UNSIGNED_INT8:
+                return CL.CL_UNSIGNED_INT8;
+            default:
+                throw new IllegalArgumentException("Unknown image channel type: " + type);
+        }
+    }
+
+    public static ImageChannelType encodeImageChannelType(int type) {
+        switch (type) {
+            case CL.CL_FLOAT:
+                return ImageChannelType.FLOAT;
+            case CL.CL_HALF_FLOAT:
+                return ImageChannelType.HALF_FLOAT;
+            case CL.CL_SIGNED_INT16:
+                return ImageChannelType.SIGNED_INT16;
+            case CL.CL_SIGNED_INT32:
+                return ImageChannelType.SIGNED_INT32;
+            case CL.CL_SIGNED_INT8:
+                return ImageChannelType.SIGNED_INT8;
+            case CL.CL_SNORM_INT16:
+                return ImageChannelType.SNORM_INT16;
+            case CL.CL_SNORM_INT8:
+                return ImageChannelType.SNORM_INT8;
+            case CL.CL_UNORM_INT8:
+                return ImageChannelType.UNORM_INT8;
+            case CL.CL_UNORM_INT16:
+                return ImageChannelType.UNORM_INT16;
+            case CL.CL_UNORM_INT_101010:
+                return ImageChannelType.UNORM_INT_101010;
+            case CL.CL_UNORM_SHORT_555:
+                return ImageChannelType.UNORM_SHORT_555;
+            case CL.CL_UNORM_SHORT_565:
+                return ImageChannelType.UNORM_SHORT_565;
+            case CL.CL_UNSIGNED_INT16:
+                return ImageChannelType.UNSIGNED_INT16;
+            case CL.CL_UNSIGNED_INT32:
+                return ImageChannelType.UNSIGNED_INT32;
+            case CL.CL_UNSIGNED_INT8:
+                return ImageChannelType.UNSIGNED_INT8;
+            default:
+                //throw new com.jme3.opencl.OpenCLException("unknown image channel type id: " + type);
+                LOG.log(Level.WARNING, "Unknown image channel type id: {0}", type);
+                return null;
+        }
+    }
+
+    public static int decodeImageType(ImageType type) {
+        switch (type) {
+//            case IMAGE_1D:
+//                return CL.CL_MEM_OBJECT_IMAGE1D;
+//            case IMAGE_1D_ARRAY:
+//                return CL.CL_MEM_OBJECT_IMAGE1D_ARRAY;
+//            case IMAGE_1D_BUFFER:
+//                return CL.CL_MEM_OBJECT_IMAGE1D_BUFFER;
+            case IMAGE_2D:
+                return CL.CL_MEM_OBJECT_IMAGE2D;
+//            case IMAGE_2D_ARRAY:
+//                return CL.CL_MEM_OBJECT_IMAGE2D_ARRAY;
+            case IMAGE_3D:
+                return CL.CL_MEM_OBJECT_IMAGE3D;
+            default:
+                throw new IllegalArgumentException("Unknown or unsupported image type: " + type);
+        }
+    }
+
+    public static ImageType encodeImageType(int type) {
+        switch (type) {
+//            case CL12.CL_MEM_OBJECT_IMAGE1D:
+//                return ImageType.IMAGE_1D;
+//            case CL12.CL_MEM_OBJECT_IMAGE1D_ARRAY:
+//                return ImageType.IMAGE_1D_ARRAY;
+//            case CL12.CL_MEM_OBJECT_IMAGE1D_BUFFER:
+//                return ImageType.IMAGE_1D_BUFFER;
+            case CL.CL_MEM_OBJECT_IMAGE2D:
+                return ImageType.IMAGE_2D;
+//            case CL12.CL_MEM_OBJECT_IMAGE2D_ARRAY:
+//                return ImageType.IMAGE_2D_ARRAY;
+            case CL.CL_MEM_OBJECT_IMAGE3D:
+                return ImageType.IMAGE_3D;
+            default:
+                //throw new com.jme3.opencl.OpenCLException("Unknown image type id: " + type);
+                LOG.log(Level.WARNING, "Unknown or unsupported image type with id: {0}", type);
+                return null;
+        }
+    }
+
+    private long getInfoSize(int param) {
+        Utils.tempBuffers[0].b16l.rewind();
+        int ret = cl.clGetImageInfo(id, param, 8, Utils.tempBuffers[0].b16l, null);
+        Utils.checkError(ret, "clGetImageInfo");
+        return Utils.tempBuffers[0].b16l.get(0);
+    }
+    
+    @Override
+    public long getWidth() {
+        return getInfoSize(CL.CL_IMAGE_WIDTH);
+    }
+
+    @Override
+    public long getHeight() {
+        return getInfoSize(CL.CL_IMAGE_HEIGHT);
+    }
+
+    @Override
+    public long getDepth() {
+        return getInfoSize(CL.CL_IMAGE_DEPTH);
+    }
+
+    @Override
+    public long getRowPitch() {
+        return getInfoSize(CL.CL_IMAGE_ROW_PITCH);
+    }
+
+    @Override
+    public long getSlicePitch() {
+        return getInfoSize(CL.CL_IMAGE_SLICE_PITCH);
+    }
+
+    @Override
+    public long getArraySize() {
+        //return getInfoSize(CL12.CL_IMAGE_ARRAY_SIZE);
+        throw new UnsupportedOperationException("Not supported in Jocl");
+    }
+
+    @Override
+    public ImageFormat getImageFormat() {
+        Utils.tempBuffers[0].b16i.rewind();
+        int ret = cl.clGetImageInfo(id, CL.CL_IMAGE_FORMAT, 8, Utils.tempBuffers[0].b16i, null);
+        Utils.checkError(ret, "clGetImageInfo");
+        int channelOrder = Utils.tempBuffers[0].b16i.get(0);
+        int channelType = Utils.tempBuffers[0].b16i.get(1);
+        return new ImageFormat(encodeImageChannelOrder(channelOrder), encodeImageChannelType(channelType));
+    }
+
+    @Override
+    public ImageType getImageType() {
+        Utils.tempBuffers[0].b16i.rewind();
+        int ret = cl.clGetMemObjectInfo(id, CL.CL_IMAGE_FORMAT, 5, Utils.tempBuffers[0].b16i, null);
+        int type = Utils.tempBuffers[0].b16i.get(0);
+        return encodeImageType(type);
+    }
+
+    @Override
+    public int getElementSize() {
+        return (int) getInfoSize(CL.CL_IMAGE_ELEMENT_SIZE);
+    }
+
+    @Override
+    public void readImage(CommandQueue queue, ByteBuffer dest, long[] origin, long[] region, long rowPitch, long slicePitch) {
+        if (origin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[1].put(origin, 0, 3).position(0);
+        Utils.pointers[2].put(region, 0, 3).position(0);
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueReadImage(q, id, CL.CL_TRUE, Utils.pointers[1], Utils.pointers[2], rowPitch, slicePitch, dest, 0, null, null);
+        Utils.checkError(ret, "clEnqueueReadImage");
+    }
+
+    @Override
+    public Event readImageAsync(CommandQueue queue, ByteBuffer dest, long[] origin, long[] region, long rowPitch, long slicePitch) {
+        if (origin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[1].put(origin, 0, 3).position(0);
+        Utils.pointers[2].put(region, 0, 3).position(0);
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueReadImage(q, id, CL.CL_FALSE, Utils.pointers[1], Utils.pointers[2], rowPitch, slicePitch, dest, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueReadImage");
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public void writeImage(CommandQueue queue, ByteBuffer dest, long[] origin, long[] region, long rowPitch, long slicePitch) {
+        if (origin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[1].put(origin, 0, 3).position(0);
+        Utils.pointers[2].put(region, 0, 3).position(0);
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueWriteImage(q, id, CL.CL_TRUE, Utils.pointers[1], Utils.pointers[2], rowPitch, slicePitch, dest, 0, null, null);
+        Utils.checkError(ret, "clEnqueueWriteImage");
+    }
+
+    @Override
+    public Event writeImageAsync(CommandQueue queue, ByteBuffer dest, long[] origin, long[] region, long rowPitch, long slicePitch) {
+        if (origin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[1].put(origin, 0, 3).position(0);
+        Utils.pointers[2].put(region, 0, 3).position(0);
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueWriteImage(q, id, CL.CL_FALSE, Utils.pointers[1], Utils.pointers[2], rowPitch, slicePitch, dest, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueWriteImage");
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public void copyTo(CommandQueue queue, Image dest, long[] srcOrigin, long[] destOrigin, long[] region) {
+        if (srcOrigin.length!=3 || destOrigin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[3].rewind();
+        Utils.pointers[1].put(srcOrigin, 0, 3).position(0);
+        Utils.pointers[2].put(destOrigin, 0, 3).position(0);
+        Utils.pointers[3].put(region, 0, 3).position(0);
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueCopyImage(q, id, ((JoclImage) dest).id, Utils.pointers[1], Utils.pointers[2], Utils.pointers[3], 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueCopyImage");
+        ret = cl.clWaitForEvents(1, Utils.pointers[0]);
+        Utils.checkError(ret, "clWaitForEvents");
+    }
+
+    @Override
+    public Event copyToAsync(CommandQueue queue, Image dest, long[] srcOrigin, long[] destOrigin, long[] region) {
+        if (srcOrigin.length!=3 || destOrigin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[3].rewind();
+        Utils.pointers[1].put(srcOrigin, 0, 3).position(0);
+        Utils.pointers[2].put(destOrigin, 0, 3).position(0);
+        Utils.pointers[3].put(region, 0, 3).position(0);
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueCopyImage(q, id, ((JoclImage) dest).id, Utils.pointers[1], Utils.pointers[2], Utils.pointers[3], 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueCopyImage");
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public ImageMapping map(CommandQueue queue, long[] origin, long[] region, MappingAccess access) {
+        if (origin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.errorBuffer.rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[3].rewind();
+        Utils.pointers[4].rewind();
+        Utils.pointers[1].put(origin, 0, 3).position(0);
+        Utils.pointers[2].put(region, 0, 3).position(0);
+        long q = ((JoclCommandQueue) queue).id;
+        long flags = Utils.getMappingAccessFlags(access);
+        ByteBuffer buf = cl.clEnqueueMapImage(q, id, CL.CL_TRUE, flags, Utils.pointers[1], Utils.pointers[2], 
+                Utils.pointers[3], Utils.pointers[4], 0, null, null, Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clEnqueueMapBuffer");
+        return new ImageMapping(buf, Utils.pointers[3].get(0), Utils.pointers[4].get(0));
+    }
+
+    @Override
+    public ImageMapping mapAsync(CommandQueue queue, long[] origin, long[] region, MappingAccess access) {
+        if (origin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.errorBuffer.rewind();
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[3].rewind();
+        Utils.pointers[4].rewind();
+        Utils.pointers[1].put(origin, 0, 3).position(0);
+        Utils.pointers[2].put(region, 0, 3).position(0);
+        long q = ((JoclCommandQueue) queue).id;
+        long flags = Utils.getMappingAccessFlags(access);
+        ByteBuffer buf = cl.clEnqueueMapImage(q, id, CL.CL_FALSE, flags, Utils.pointers[1], Utils.pointers[2], 
+                Utils.pointers[3], Utils.pointers[4], 0, null, Utils.pointers[0], Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clEnqueueMapBuffer");
+        long event = Utils.pointers[0].get(0);
+        return new ImageMapping(buf, Utils.pointers[3].get(0), Utils.pointers[4].get(0), 
+                new JoclEvent(event));
+    }
+
+    @Override
+    public void unmap(CommandQueue queue, ImageMapping mapping) {
+        long q = ((JoclCommandQueue)queue).id;
+        Utils.pointers[0].rewind();
+        int ret = cl.clEnqueueUnmapMemObject(q, id, mapping.buffer, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueUnmapMemObject");
+        ret = cl.clWaitForEvents(1, Utils.pointers[0]);
+        Utils.checkError(ret, "clWaitForEvents");
+    }
+
+    @Override
+    public Event fillAsync(CommandQueue queue, long[] origin, long[] region, ColorRGBA color) {
+        if (origin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        throw new UnsupportedOperationException("Not supported by Jocl!");
+    }
+
+    @Override
+    public Event fillAsync(CommandQueue queue, long[] origin, long[] region, int[] color) {
+        if (color.length != 4) {
+            throw new IllegalArgumentException("the passed color array must have length 4");
+        }
+        if (origin.length!=3 || region.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        throw new UnsupportedOperationException("Not supported by Jocl!");
+    }
+
+    @Override
+    public Event copyToBufferAsync(CommandQueue queue, Buffer dest, long[] srcOrigin, long[] srcRegion, long destOffset) {
+        if (srcOrigin.length!=3 || srcRegion.length!=3) {
+            throw new IllegalArgumentException("origin and region must both be arrays of length 3");
+        }
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[2].rewind();
+        Utils.pointers[1].put(srcOrigin, 0, 3).position(0);
+        Utils.pointers[2].put(srcRegion, 0, 3).position(0);
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueCopyImageToBuffer(q, id, ((JoclBuffer) dest).id, 
+                Utils.pointers[1], Utils.pointers[2], destOffset, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueCopyImageToBuffer");
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+
+    @Override
+    public Event acquireImageForSharingAsync(CommandQueue queue) {
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[1].put(0, id);
+        long q = ((JoclCommandQueue)queue).id;
+        ((CLGL) cl).clEnqueueAcquireGLObjects(q, 1, Utils.pointers[1], 0, null, Utils.pointers[0]);
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+    @Override
+    public Event releaseImageForSharingAsync(CommandQueue queue) {
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[1].put(0, id);
+        long q = ((JoclCommandQueue)queue).id;
+        ((CLGL) cl).clEnqueueReleaseGLObjects(q, 1, Utils.pointers[1], 0, null, Utils.pointers[0]);
+        long event = Utils.pointers[0].get(0);
+        return new JoclEvent(event);
+    }
+    
+    @Override
+    public ObjectReleaser getReleaser() {
+        return new ReleaserImpl(id);
+    }
+    private static class ReleaserImpl implements ObjectReleaser {
+        private long mem;
+        private ReleaserImpl(long mem) {
+            this.mem = mem;
+        }
+        @Override
+        public void release() {
+            if (mem != 0) {
+                int ret = CLPlatform.getLowLevelCLInterface().clReleaseMemObject(mem);
+                mem = 0;
+                Utils.reportError(ret, "clReleaseMemObject");
+            }
+        }
+    }
+}

jme3-jogl/src/main/java/com/jme3/opencl/jocl/JoclKernel.java

@@ -0,0 +1,261 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.math.Quaternion;
+import com.jme3.math.Vector2f;
+import com.jme3.math.Vector4f;
+import com.jme3.opencl.*;
+import com.jme3.opencl.Buffer;
+import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.opencl.CLKernel;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.llb.CL;
+import java.nio.*;
+import java.nio.charset.Charset;
+
+import static com.jogamp.common.os.Platform.is32Bit;
+
+/**
+ *
+ * @author shaman
+ */
+public class JoclKernel extends Kernel {
+
+    final long kernel;
+    final CL cl;
+
+    public JoclKernel(long kernel) {
+        this.kernel = kernel;
+        this.cl = CLPlatform.getLowLevelCLInterface();
+        OpenCLObjectManager.getInstance().registerObject(this);
+    }
+    
+    @Override
+    public String getName() {
+        Utils.pointers[0].rewind();
+        int ret = cl.clGetKernelInfo(kernel, CL.CL_KERNEL_FUNCTION_NAME, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clGetKernelInfo");
+        int count = (int) Utils.pointers[0].get(0);
+        ByteBuffer buf = ByteBuffer.allocateDirect(count);
+        ret = cl.clGetKernelInfo(kernel, CL.CL_KERNEL_FUNCTION_NAME, count, buf, null);
+        Utils.checkError(ret, "clGetKernelInfo");
+        byte[] data = new byte[count];
+        buf.get(data);
+        return new String(data, Charset.forName("ASCII"));
+    }
+
+    @Override
+    public int getArgCount() {
+        Utils.tempBuffers[0].b16i.rewind();
+        int ret = cl.clGetKernelInfo(kernel, CL.CL_KERNEL_NUM_ARGS, 4, Utils.tempBuffers[0].b16i, null);
+        Utils.checkError(ret, "clGetKernelInfo");
+        return Utils.tempBuffers[0].b16i.get(0);
+    }
+
+    @Override
+    public long getMaxWorkGroupSize(Device device) {
+        long d = ((JoclDevice) device).id;
+        Utils.tempBuffers[0].b16l.rewind();
+        int ret = cl.clGetKernelWorkGroupInfo(kernel, d, CL.CL_KERNEL_WORK_GROUP_SIZE, 8, Utils.tempBuffers[0].b16l, null);
+        Utils.checkError(ret, "clGetKernelWorkGroupInfo");
+        return Utils.tempBuffers[0].b16l.get(0);
+   }
+    
+    @Override
+    public void setArg(int index, LocalMemPerElement t) {
+        int ret = cl.clSetKernelArg (kernel, index, t.getSize() * workGroupSize.getSizes()[0] * workGroupSize.getSizes()[1] * workGroupSize.getSizes()[2], null);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, LocalMem t) {
+        int ret = cl.clSetKernelArg (kernel, index, t.getSize(), null);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, Buffer t) {
+        Utils.tempBuffers[0].b16l.rewind();
+        Utils.tempBuffers[0].b16l.put(0, ((JoclBuffer) t).id);
+        int ret = cl.clSetKernelArg(kernel, index, is32Bit()?4:8, Utils.tempBuffers[0].b16l);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+    
+    @Override
+    public void setArg(int index, Image i) {
+        Utils.tempBuffers[0].b16l.rewind();
+        Utils.tempBuffers[0].b16l.put(0, ((JoclImage) i).id);
+        int ret = cl.clSetKernelArg(kernel, index, is32Bit()?4:8, Utils.tempBuffers[0].b16l);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, byte b) {
+        ByteBuffer buf = Utils.tempBuffers[0].b16;
+        buf.position(0);
+        buf.put(0, b);
+        int ret = cl.clSetKernelArg(kernel, index, 1, buf);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, short s) {
+        ShortBuffer buf = Utils.tempBuffers[0].b16s;
+        buf.position(0);
+        buf.put(0, s);
+        int ret = cl.clSetKernelArg(kernel, index, 2, buf);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, int i) {
+        IntBuffer buf = Utils.tempBuffers[0].b16i;
+        buf.position(0);
+        buf.limit(1);
+        buf.put(0, i);
+        int ret = cl.clSetKernelArg(kernel, index, 4, buf);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, long l) {
+        LongBuffer buf = Utils.tempBuffers[0].b16l;
+        buf.position(0);
+        buf.limit(1);
+        buf.put(0, l);
+        int ret = cl.clSetKernelArg(kernel, index, 8, buf);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, float f) {
+        FloatBuffer buf = Utils.tempBuffers[0].b16f;
+        buf.position(0);
+        buf.limit(1);
+        buf.put(0, f);
+        int ret = cl.clSetKernelArg(kernel, index, 4, buf);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, double d) {
+        DoubleBuffer buf = Utils.tempBuffers[0].b16d;
+        buf.position(0);
+        buf.limit(1);
+        buf.put(0, d);
+        int ret = cl.clSetKernelArg(kernel, index, 8, buf);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, Vector2f v) {
+        FloatBuffer buf = Utils.tempBuffers[0].b16f;
+        buf.position(0);
+        buf.limit(2);
+        buf.put(0, v.x);
+        buf.put(1, v.y);
+        int ret = cl.clSetKernelArg(kernel, index, 8, buf);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, Vector4f v) {
+        FloatBuffer buf = Utils.tempBuffers[0].b16f;
+        buf.position(0);
+        buf.limit(4);
+        buf.put(0, v.x);
+        buf.put(1, v.y);
+        buf.put(2, v.z);
+        buf.put(3, v.w);
+        int ret = cl.clSetKernelArg(kernel, index, 16, buf);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, Quaternion q) {
+        FloatBuffer buf = Utils.tempBuffers[0].b16f;
+        buf.position(0);
+        buf.limit(4);
+        buf.put(0, q.getX());
+        buf.put(1, q.getY());
+        buf.put(2, q.getZ());
+        buf.put(3, q.getW());
+        int ret = cl.clSetKernelArg(kernel, index, 16, buf);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public void setArg(int index, ByteBuffer buffer, long size) {
+        int ret = cl.clSetKernelArg(kernel, index, size, buffer);
+        Utils.checkError(ret, "clSetKernelArg");
+    }
+
+    @Override
+    public Event Run(CommandQueue queue) {
+        Utils.pointers[0].rewind();
+        Utils.pointers[1].rewind();
+        Utils.pointers[1].put(globalWorkSize.getSizes(), 0, globalWorkSize.getSizes().length);
+        Utils.pointers[1].position(0);
+        PointerBuffer p2 = null;
+        if (workGroupSize.getSizes()[0] > 0) {
+            p2 = Utils.pointers[2].rewind();
+            p2.put(workGroupSize.getSizes(), 0, workGroupSize.getSizes().length);
+            p2.position(0);
+        }
+        long q = ((JoclCommandQueue) queue).id;
+        int ret = cl.clEnqueueNDRangeKernel(q, kernel,
+			globalWorkSize.getDimension(), null, Utils.pointers[1],
+			p2, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clEnqueueNDRangeKernel");
+        return new JoclEvent(Utils.pointers[0].get(0));
+    }
+
+    @Override
+    public ObjectReleaser getReleaser() {
+        return new ReleaserImpl(kernel);
+    }
+    private static class ReleaserImpl implements ObjectReleaser {
+        private long kernel;
+        private ReleaserImpl(long kernel) {
+            this.kernel = kernel;
+        }
+        @Override
+        public void release() {
+            if (kernel != 0) {
+                int ret = CLPlatform.getLowLevelCLInterface().clReleaseKernel(kernel);
+                kernel = 0;
+                Utils.reportError(ret, "clReleaseKernel");
+            }
+        }
+    }
+}

jme3-jogl/src/main/java/com/jme3/opencl/jocl/JoclPlatform.java

@@ -0,0 +1,127 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.opencl.Device;
+import com.jme3.opencl.Platform;
+import com.jogamp.opencl.CLDevice;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.llb.CL;
+import java.util.ArrayList;
+import java.util.Arrays;
+import java.util.Collection;
+import java.util.List;
+
+/**
+ *
+ * @author shaman
+ */
+public final class JoclPlatform implements Platform {
+    
+    final CLPlatform platform;
+    List<JoclDevice> devices;
+    
+    public JoclPlatform(CLPlatform platform) {
+        this.platform = platform;
+    }
+
+    public CLPlatform getPlatform() {
+        return platform;
+    }
+    
+    @Override
+    public List<JoclDevice> getDevices() {
+        if (devices == null) {
+            devices = new ArrayList<>();
+            for (CLDevice d : platform.listCLDevices()) {
+                devices.add(new JoclDevice(d, this));
+            }
+        }
+        return devices;
+    }
+
+    @Override
+    public String getProfile() {
+        return platform.getProfile();
+    }
+
+    @Override
+    public boolean isFullProfile() {
+        return getProfile().contains("FULL_PROFILE");
+    }
+
+    @Override
+    public boolean isEmbeddedProfile() {
+        return getProfile().contains("EMBEDDED_PROFILE");
+    }
+
+    @Override
+    public String getVersion() {
+        return platform.getVendor();
+    }
+
+    @Override
+    public int getVersionMajor() {
+        return Utils.getMajorVersion(getVersion(), "OpenCL ");
+    }
+
+    @Override
+    public int getVersionMinor() {
+        return Utils.getMinorVersion(getVersion(), "OpenCL ");
+    }
+
+    @Override
+    public String getName() {
+        return platform.getName();
+    }
+
+    @Override
+    public String getVendor() {
+        return platform.getVendor();
+    }
+
+    @Override
+    public boolean hasExtension(String extension) {
+        return getExtensions().contains(extension);
+    }
+
+    @Override
+    public boolean hasOpenGLInterop() {
+        return hasExtension("cl_khr_gl_sharing");
+    }
+
+    @Override
+    public Collection<? extends String> getExtensions() {
+        return platform.getExtensions();
+    }
+
+}

jme3-jogl/src/main/java/com/jme3/opencl/jocl/JoclProgram.java

@@ -0,0 +1,152 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.opencl.Kernel;
+import com.jme3.opencl.KernelCompilationException;
+import com.jme3.opencl.OpenCLObjectManager;
+import com.jme3.opencl.Program;
+import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.llb.CL;
+import com.jogamp.opencl.util.CLUtil;
+import java.nio.ByteBuffer;
+import java.util.logging.Level;
+import java.util.logging.Logger;
+
+import static com.jogamp.common.nio.Buffers.newDirectByteBuffer;
+import static com.jogamp.opencl.CLException.newException;
+import static com.jogamp.opencl.llb.CL.CL_SUCCESS;
+/**
+ *
+ * @author shaman
+ */
+public class JoclProgram implements Program {
+    private static final Logger LOG = Logger.getLogger(JoclProgram.class.getName());
+    
+    final long program;
+    final CL cl;
+    private final JoclContext context;
+
+    public JoclProgram(long program, JoclContext context) {
+        this.program = program;
+        this.context = context;
+        this.cl = CLPlatform.getLowLevelCLInterface();
+        OpenCLObjectManager.getInstance().registerObject(this);
+    }
+
+    @Override
+    public void build(String args) throws KernelCompilationException {
+        int ret = cl.clBuildProgram(program, 0, null, args, null);
+        if (ret != CL.CL_SUCCESS) {
+            String log = Log();
+            LOG.log(Level.WARNING, "Unable to compile program:\n{0}", log);
+            if (ret == CL.CL_BUILD_PROGRAM_FAILURE) {
+                throw new KernelCompilationException("Failed to build program", ret, log);
+            } else {
+                Utils.checkError(ret, "clBuildProgram");
+            }
+        } else {
+            LOG.log(Level.INFO, "Program compiled:\n{0}", Log());
+        }
+    }
+
+    @Override
+    public void build() throws KernelCompilationException {
+        build("");
+    }
+    
+    private String Log(long device) {
+        Utils.pointers[0].rewind();
+        int ret = cl.clGetProgramBuildInfo(program, device, CL.CL_PROGRAM_BUILD_LOG, 0, null, Utils.pointers[0]);
+        Utils.checkError(ret, "clGetProgramBuildInfo");
+        int count = (int) Utils.pointers[0].get(0);
+        final ByteBuffer buffer = newDirectByteBuffer(count);
+        ret = cl.clGetProgramBuildInfo(program, device, CL.CL_PROGRAM_BUILD_LOG, buffer.capacity(), buffer, null);
+        Utils.checkError(ret, "clGetProgramBuildInfo");
+        return CLUtil.clString2JavaString(buffer, count);
+    }
+    
+    private String Log() {
+        StringBuilder str = new StringBuilder();
+        for (JoclDevice device : context.getDevices()) {
+            long d = device.id;
+            str.append(device.getName()).append(":\n");
+            str.append(Log(d));
+            str.append('\n');
+        }
+        return str.toString();
+    }
+
+    @Override
+    public Kernel createKernel(String name) {
+        Utils.errorBuffer.rewind();
+        long kernel = cl.clCreateKernel(program, name, Utils.errorBuffer);
+        Utils.checkError(Utils.errorBuffer, "clCreateKernel");
+        return new JoclKernel(kernel);
+    }
+
+    @Override
+    public Kernel[] createAllKernels() {
+        Utils.tempBuffers[0].b16i.rewind();
+        int ret = cl.clCreateKernelsInProgram(program, 0, null, Utils.tempBuffers[0].b16i);
+        Utils.checkError(ret, "clCreateKernelsInProgram");
+        int count = Utils.tempBuffers[0].b16i.get(0);
+        PointerBuffer buf = PointerBuffer.allocateDirect(count);
+        ret = cl.clCreateKernelsInProgram(program, count, buf, null);
+        Utils.checkError(ret, "clCreateKernelsInProgram");
+        Kernel[] kx = new Kernel[count];
+        for (int i=0; i<count; ++i) {
+            kx[i] = new JoclKernel(buf.get());
+        }
+        return kx;
+    }
+
+    @Override
+    public ObjectReleaser getReleaser() {
+        return new ReleaserImpl(program);
+    }
+    private static class ReleaserImpl implements ObjectReleaser {
+        private long program;
+        private ReleaserImpl(long program) {
+            this.program = program;
+        }
+        @Override
+        public void release() {
+            if (program != 0) {
+                int ret = CLPlatform.getLowLevelCLInterface().clReleaseProgram(program);
+                program = 0;
+                Utils.reportError(ret, "clReleaseProgram");
+            }
+        }
+    }
+}

jme3-jogl/src/main/java/com/jme3/opencl/jocl/Utils.java

@@ -0,0 +1,160 @@
+/*
+ * Copyright (c) 2009-2016 jMonkeyEngine
+ * All rights reserved.
+ *
+ * Redistribution and use in source and binary forms, with or without
+ * modification, are permitted provided that the following conditions are
+ * met:
+ *
+ * * Redistributions of source code must retain the above copyright
+ *   notice, this list of conditions and the following disclaimer.
+ *
+ * * Redistributions in binary form must reproduce the above copyright
+ *   notice, this list of conditions and the following disclaimer in the
+ *   documentation and/or other materials provided with the distribution.
+ *
+ * * Neither the name of 'jMonkeyEngine' nor the names of its contributors
+ *   may be used to endorse or promote products derived from this software
+ *   without specific prior written permission.
+ *
+ * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS
+ * "AS IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED
+ * TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
+ * PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
+ * CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
+ * EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
+ * PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
+ * PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
+ * LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
+ * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
+ * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
+ */
+package com.jme3.opencl.jocl;
+
+import com.jme3.opencl.MappingAccess;
+import com.jme3.opencl.MemoryAccess;
+import com.jme3.opencl.OpenCLException;
+import com.jme3.util.BufferUtils;
+import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.opencl.CLEventList;
+import com.jogamp.opencl.CLException;
+import com.jogamp.opencl.CLMemory;
+import com.jogamp.opencl.CLVersion;
+import com.jogamp.opencl.llb.CL;
+import java.lang.reflect.Field;
+import java.nio.*;
+import java.util.EnumSet;
+import java.util.Map;
+import java.util.logging.Level;
+import java.util.logging.Logger;
+
+/**
+ *
+ * @author shaman
+ */
+public class Utils {
+    private static final Logger LOG = Logger.getLogger(Utils.class.getName());
+    private Utils() {}
+    
+    public static int getMajorVersion(String version, String prefix) {
+        String s = version.substring(prefix.length());
+        return Integer.parseInt(s);
+    }
+    
+    public static int getMinorVersion(String version, String prefix) {
+        String s = version.substring(prefix.length());
+        int major = Integer.parseInt(s);
+        s = s.substring((int) (Math.log10(major) + 2));
+        return Integer.parseInt(s);
+    }
+    
+    public static final class TempBuffer {
+        public final ByteBuffer b16;
+        public final ShortBuffer b16s;
+        public final IntBuffer b16i;
+        public final LongBuffer b16l;
+        public final FloatBuffer b16f;
+        public final DoubleBuffer b16d;
+        public TempBuffer() {
+            b16 = BufferUtils.createByteBuffer(16);
+            b16s = b16.asShortBuffer();
+            b16i = b16.asIntBuffer();
+            b16l = b16.asLongBuffer();
+            b16f = b16.asFloatBuffer();
+            b16d = b16.asDoubleBuffer();
+        }
+    }
+    public static final ByteBuffer b80; //needed for ImageDescriptor
+    public static final LongBuffer b80l;
+    public static final TempBuffer[] tempBuffers = new TempBuffer[8];
+    public static final PointerBuffer[] pointers = new PointerBuffer[8];
+    static {
+        for (int i=0; i<8; ++i) {
+            tempBuffers[i] = new TempBuffer();
+            pointers[i] = PointerBuffer.allocateDirect(4);
+        }
+        errorBuffer = BufferUtils.createIntBuffer(1);
+        b80 = BufferUtils.createByteBuffer(80);
+        b80l = b80.asLongBuffer();
+    }
+    
+    public static IntBuffer errorBuffer;
+    public static void checkError(IntBuffer errorBuffer, String callName) {
+        checkError(errorBuffer.get(0), callName);
+    }
+    public static void checkError(int error, String callName) {
+        if (error != CL.CL_SUCCESS) {
+            String errname = getErrorName(error);
+            if (errname == null) {
+                errname = "UNKNOWN";
+            }
+            throw new OpenCLException("OpenCL error in " + callName + ": " + errname + " (0x" + Integer.toHexString(error) + ")", error);
+        }
+    }
+    
+    public static void reportError(int error, String callName) {
+        if (error != CL.CL_SUCCESS) {
+            String errname = getErrorName(error);
+            if (errname == null) {
+                errname = "UNKNOWN";
+            }
+            LOG.log(Level.WARNING, "OpenCL error in {0}: {1} (0x{2})", new Object[]{callName, errname, Integer.toHexString(error)});
+        }
+    }
+    
+    public static String getErrorName(int code) {
+        return CLException.resolveErrorCode(code);
+    }
+    
+     public static long getMemoryAccessFlags(MemoryAccess ma) {
+        switch (ma) {
+            case READ_ONLY: return CL.CL_MEM_READ_ONLY;
+            case WRITE_ONLY: return CL.CL_MEM_WRITE_ONLY;
+            case READ_WRITE: return CL.CL_MEM_READ_WRITE;
+            default: throw new IllegalArgumentException("Unknown memory access: "+ma);
+        }
+    }
+    public static MemoryAccess getMemoryAccessFromFlag(long flag) {
+        if ((flag & CL.CL_MEM_READ_WRITE) > 0) {
+            return MemoryAccess.READ_WRITE;
+        }
+        if ((flag & CL.CL_MEM_READ_ONLY) > 0) {
+            return MemoryAccess.READ_ONLY;
+        }
+        if ((flag & CL.CL_MEM_WRITE_ONLY) > 0) {
+            return MemoryAccess.WRITE_ONLY;
+        }
+        throw new OpenCLException("Unknown memory access flag: "+flag);
+    }
+    
+    public static long getMappingAccessFlags(MappingAccess ma) {
+        switch (ma) {
+            case MAP_READ_ONLY: return CL.CL_MAP_READ;
+            case MAP_READ_WRITE: return CL.CL_MAP_READ | CL.CL_MAP_WRITE;
+            case MAP_WRITE_ONLY: return CL.CL_MAP_WRITE;
+            case MAP_WRITE_INVALIDATE: return CL.CL_MAP_WRITE; //MAP_WRITE_INVALIDATE_REGION not supported
+            default: throw new IllegalArgumentException("Unknown mapping access: "+ma);
+        }
+    }
+
+}

jme3-jogl/src/main/java/com/jme3/system/jogl/JoglContext.java

@@ -36,6 +36,11 @@ import com.jme3.input.JoyInput;
 import com.jme3.input.KeyInput;
 import com.jme3.input.MouseInput;
 import com.jme3.opencl.Context;
+import com.jme3.opencl.DefaultPlatformChooser;
+import com.jme3.opencl.Device;
+import com.jme3.opencl.PlatformChooser;
+import com.jme3.opencl.jocl.JoclDevice;
+import com.jme3.opencl.jocl.JoclPlatform;
 import com.jme3.renderer.Renderer;
 import com.jme3.renderer.RendererException;
 import com.jme3.renderer.jogl.JoglGL;
@@ -56,6 +61,10 @@ import com.jme3.system.JmeContext;
 import com.jme3.system.NanoTimer;
 import com.jme3.system.SystemListener;
 import com.jme3.system.Timer;
+import com.jogamp.opencl.CLDevice;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.gl.CLGLContext;
+import com.jogamp.opencl.llb.CL;
 
 import java.nio.IntBuffer;
 import java.util.concurrent.atomic.AtomicBoolean;
@@ -65,6 +74,8 @@ import java.util.logging.Logger;
 import com.jogamp.opengl.GL;
 import com.jogamp.opengl.GL2GL3;
 import com.jogamp.opengl.GLContext;
+import java.util.ArrayList;
+import java.util.List;
 
 public abstract class JoglContext implements JmeContext {
 
@@ -218,8 +229,95 @@ public abstract class JoglContext implements JmeContext {
         }
     }
     
+    @SuppressWarnings("unchecked")
     protected void initOpenCL() {
         logger.info("Initialize OpenCL with JOGL");
+        
+        //load platforms and devices
+        StringBuilder platformInfos = new StringBuilder();
+        ArrayList<JoclPlatform> platforms = new ArrayList<JoclPlatform>();
+        for (CLPlatform p : CLPlatform.listCLPlatforms()) {
+            platforms.add(new JoclPlatform(p));
+        }
+        platformInfos.append("Available OpenCL platforms:");
+        for (int i=0; i<platforms.size(); ++i) {
+            JoclPlatform platform = platforms.get(i);
+            platformInfos.append("\n * Platform ").append(i+1);
+            platformInfos.append("\n *   Name: ").append(platform.getName());
+            platformInfos.append("\n *   Vendor: ").append(platform.getVendor());
+            platformInfos.append("\n *   Version: ").append(platform.getVersion());
+            platformInfos.append("\n *   Profile: ").append(platform.getProfile());
+            platformInfos.append("\n *   Supports interop: ").append(platform.hasOpenGLInterop());
+            List<JoclDevice> devices = platform.getDevices();
+            platformInfos.append("\n *   Available devices:");
+            for (int j=0; j<devices.size(); ++j) {
+                JoclDevice device = devices.get(j);
+                platformInfos.append("\n *    * Device ").append(j+1);
+                platformInfos.append("\n *    *   Name: ").append(device.getName());
+                platformInfos.append("\n *    *   Vendor: ").append(device.getVendor());
+                platformInfos.append("\n *    *   Version: ").append(device.getVersion());
+                platformInfos.append("\n *    *   Profile: ").append(device.getProfile());
+                platformInfos.append("\n *    *   Compiler version: ").append(device.getCompilerVersion());
+                platformInfos.append("\n *    *   Device type: ").append(device.getDeviceType());
+                platformInfos.append("\n *    *   Compute units: ").append(device.getComputeUnits());
+                platformInfos.append("\n *    *   Work group size: ").append(device.getMaxiumWorkItemsPerGroup());
+                platformInfos.append("\n *    *   Global memory: ").append(device.getGlobalMemorySize()).append("B");
+                platformInfos.append("\n *    *   Local memory: ").append(device.getLocalMemorySize()).append("B");
+                platformInfos.append("\n *    *   Constant memory: ").append(device.getMaximumConstantBufferSize()).append("B");
+                platformInfos.append("\n *    *   Supports double: ").append(device.hasDouble());
+                platformInfos.append("\n *    *   Supports half floats: ").append(device.hasHalfFloat());
+                platformInfos.append("\n *    *   Supports writable 3d images: ").append(device.hasWritableImage3D());
+                platformInfos.append("\n *    *   Supports interop: ").append(device.hasOpenGLInterop());
+            }
+        }
+        logger.info(platformInfos.toString());
+        
+        //choose devices
+        PlatformChooser chooser = null;
+        if (settings.getOpenCLPlatformChooser() != null) {
+            try {
+                chooser = (PlatformChooser) Class.forName(settings.getOpenCLPlatformChooser()).newInstance();
+            } catch (Exception ex) {
+                logger.log(Level.WARNING, "unable to instantiate custom PlatformChooser", ex);
+            }
+        }
+        if (chooser == null) {
+            chooser = new DefaultPlatformChooser();
+        }
+        List<? extends Device> choosenDevices = chooser.chooseDevices(platforms);
+        List<CLDevice> devices = new ArrayList<>(choosenDevices.size());
+        JoclPlatform platform = null;
+        for (Device d : choosenDevices) {
+            if (!(d instanceof JoclDevice)) {
+                logger.log(Level.SEVERE, "attempt to return a custom Device implementation from PlatformChooser: {0}", d);
+                return;
+            }
+            JoclDevice ld = (JoclDevice) d;
+            if (platform == null) {
+                platform = ld.getPlatform();
+            } else if (platform != ld.getPlatform()) {
+                logger.severe("attempt to use devices from different platforms");
+                return;
+            }
+            devices.add(ld.getDevice());
+        }
+        if (devices.isEmpty()) {
+            logger.warning("no devices specified, no OpenCL context created");
+            return;
+        }
+        logger.log(Level.INFO, "chosen platform: {0}", platform.getName());
+        logger.log(Level.INFO, "chosen devices: {0}", choosenDevices);
+        
+        //create context
+        try {
+            CLGLContext c = CLGLContext.create(GLContext.getCurrent(), devices.toArray(new CLDevice[devices.size()]));
+            clContext = new com.jme3.opencl.jocl.JoclContext(c, (List<JoclDevice>) choosenDevices);
+        } catch (Exception ex) {
+            logger.log(Level.SEVERE, "Unable to create OpenCL context", ex);
+            return;
+        }
+        
+        logger.info("OpenCL context created");
     }
 
     public void internalCreate() {

jme3-lwjgl/src/main/java/com/jme3/opencl/lwjgl/LwjglImage.java

@@ -419,9 +419,7 @@ public class LwjglImage implements Image {
                 Utils.pointerBuffers[1], Utils.pointerBuffers[2], 
                 Utils.pointerBuffers[3], Utils.pointerBuffers[4], null, null, Utils.errorBuffer);
         Utils.checkError(Utils.errorBuffer, "clEnqueueMapBuffer");
-        long event = Utils.pointerBuffers[0].get(0);
-        return new ImageMapping(buf, Utils.pointerBuffers[3].get(0), Utils.pointerBuffers[4].get(0), 
-                new LwjglEvent(q.getCLEvent(event)));
+        return new ImageMapping(buf, Utils.pointerBuffers[3].get(0), Utils.pointerBuffers[4].get(0));
     }
 
     @Override
@@ -441,9 +439,10 @@ public class LwjglImage implements Image {
         Utils.errorBuffer.rewind();
         ByteBuffer buf = CL10.clEnqueueMapImage(q, image, CL10.CL_FALSE, flags, 
                 Utils.pointerBuffers[1], Utils.pointerBuffers[2], 
-                Utils.pointerBuffers[3], Utils.pointerBuffers[4], null, null, Utils.errorBuffer);
+                Utils.pointerBuffers[3], Utils.pointerBuffers[4], null, Utils.pointerBuffers[0], Utils.errorBuffer);
         Utils.checkError(Utils.errorBuffer, "clEnqueueMapBuffer");
-        return new ImageMapping(buf, Utils.pointerBuffers[3].get(0), Utils.pointerBuffers[4].get(0));
+        long event = Utils.pointerBuffers[0].get(0);
+        return new ImageMapping(buf, Utils.pointerBuffers[3].get(0), Utils.pointerBuffers[4].get(0), new LwjglEvent(q.getCLEvent(event)));
     }
 
     @Override