From 5abb164b19a244345672a0b0f37b6e9ca68da7ba Mon Sep 17 00:00:00 2001 From: Wade Walker Date: Fri, 4 Apr 2014 15:55:35 -0500 Subject: Start adding texture interop test. The test here is still not complete, just checking in so I can switch branches. --- test/com/jogamp/opencl/gl/CLGLTest.java | 83 +++++++++++++++++++++++++++++++++ 1 file changed, 83 insertions(+) (limited to 'test/com/jogamp/opencl/gl/CLGLTest.java') diff --git a/test/com/jogamp/opencl/gl/CLGLTest.java b/test/com/jogamp/opencl/gl/CLGLTest.java index b5d85690..e7906cec 100644 --- a/test/com/jogamp/opencl/gl/CLGLTest.java +++ b/test/com/jogamp/opencl/gl/CLGLTest.java @@ -33,6 +33,7 @@ package com.jogamp.opencl.gl; import com.jogamp.common.nio.Buffers; +import com.jogamp.opencl.CLBuffer; import com.jogamp.opencl.CLCommandQueue; import javax.media.opengl.GL2; @@ -51,6 +52,7 @@ import com.jogamp.opencl.util.CLDeviceFilters; import com.jogamp.opencl.util.CLPlatformFilters; import java.io.IOException; +import java.nio.ByteBuffer; import java.nio.IntBuffer; import javax.media.opengl.GLCapabilities; @@ -226,6 +228,87 @@ public class CLGLTest extends UITestCase { } +// @Test(timeout=15000) + @Test + public void textureSharing() { + + out.println(" - - - glcl; textureSharing - - - "); + if(MiscUtils.isOpenCLUnavailable()) + return; + + initGL(); + makeGLCurrent(); + assertTrue(glcontext.isCurrent()); + + @SuppressWarnings("unchecked") + CLPlatform platform = CLPlatform.getDefault(glSharing(glcontext)); + if(platform == null) { + out.println("test aborted"); + return; + } + + @SuppressWarnings("unchecked") + CLDevice device = platform.getMaxFlopsDevice(CLDeviceFilters.glSharing()); + out.println(device); + + CLGLContext context = CLGLContext.create(glcontext, device); + + try { + out.println(context); + + GL2 gl = glcontext.getGL().getGL2(); + + // create and write GL texture + int[] id = new int[1]; + gl.glGenTextures(id.length, id, 0); + gl.glActiveTexture(GL2.GL_TEXTURE0); + gl.glBindTexture (GL2.GL_TEXTURE_2D, id[0]); +// gl.glTexParameteri(GL2.GL_TEXTURE_2D, GL2.GL_TEXTURE_MAG_FILTER, GL2.GL_NEAREST); +// gl.glTexParameteri(GL2.GL_TEXTURE_2D, GL2.GL_TEXTURE_MIN_FILTER, GL2.GL_NEAREST); +// gl.glTexParameteri(GL2.GL_TEXTURE_2D, GL2.GL_TEXTURE_BASE_LEVEL, 0); +// gl.glTexParameteri(GL2.GL_TEXTURE_2D, GL2.GL_TEXTURE_MAX_LEVEL, 0); + + ByteBuffer bufferGL = Buffers.newDirectByteBuffer(new byte [] { + (byte)0, (byte)5, (byte)10, (byte)0xff, + (byte)15, (byte)20, (byte)25, (byte)0xff, + (byte)30, (byte)35, (byte)40, (byte)0xff, + (byte)45, (byte)50, (byte)55, (byte)0xff}); + bufferGL.rewind(); + gl.glTexImage2D(GL2.GL_TEXTURE_2D, 0, GL2.GL_RGBA, 2, 2, 0, GL2.GL_RGBA, GL2.GL_UNSIGNED_BYTE, bufferGL); + gl.glBindTexture(GL2.GL_TEXTURE_2D, 0); + gl.glFinish(); + + // create CLGL buffer + ByteBuffer bufferCL = Buffers.newDirectByteBuffer(2*2*4); + CLGLTexture2d clTexture = context.createFromGLTexture2d(bufferCL, GL2.GL_TEXTURE_2D, id[0], 0, CLBuffer.Mem.READ_ONLY); + +// assertEquals(bufferGL.capacity(), clTexture.getCLCapacity()); +// assertEquals(bufferGL.capacity(), clTexture.getCLSize()); + + CLCommandQueue queue = device.createCommandQueue(); + + // read gl buffer into cl nio buffer + queue.putAcquireGLObject(clTexture) + .putReadImage(clTexture, true) + .putReleaseGLObject(clTexture); + + while(bufferCL.hasRemaining()) { + byte bGL = bufferGL.get(); + byte bCL = bufferCL.get(); + assertEquals(bGL, bCL); + } + + out.println(clTexture); + + clTexture.release(); + gl.glDeleteBuffers(1, id, 0); + } + finally { + context.release(); + deinitGL(); + } + } + private void makeGLCurrent() { // we are patient... while(true) { -- cgit v1.2.3 From 00f4325c5a46bf7c46be8646c1eb6f53b632f30a Mon Sep 17 00:00:00 2001 From: Wade Walker Date: Sun, 6 Apr 2014 14:20:19 -0500 Subject: Finish texture sharing test. Make the test modify a GL texture with a CL kernel, then loop over the texture afterwards to check each texel has the right value. Also make the test loop over all platforms and devices that support sharing. --- src/com/jogamp/opencl/util/CLDeviceFilters.java | 2 +- src/com/jogamp/opencl/util/CLPlatformFilters.java | 18 +++- test/com/jogamp/opencl/gl/CLGLTest.java | 112 ++++++++++++++-------- 3 files changed, 88 insertions(+), 44 deletions(-) (limited to 'test/com/jogamp/opencl/gl/CLGLTest.java') diff --git a/src/com/jogamp/opencl/util/CLDeviceFilters.java b/src/com/jogamp/opencl/util/CLDeviceFilters.java index a5057fb6..6281b844 100644 --- a/src/com/jogamp/opencl/util/CLDeviceFilters.java +++ b/src/com/jogamp/opencl/util/CLDeviceFilters.java @@ -70,7 +70,7 @@ public class CLDeviceFilters { } /** - * Accepts all devices which support OpenGL-OpenCL interoparability. + * Accepts all devices which support OpenGL-OpenCL interoperability. */ public static Filter glSharing() { return new Filter() { diff --git a/src/com/jogamp/opencl/util/CLPlatformFilters.java b/src/com/jogamp/opencl/util/CLPlatformFilters.java index 48d20916..14e1507e 100644 --- a/src/com/jogamp/opencl/util/CLPlatformFilters.java +++ b/src/com/jogamp/opencl/util/CLPlatformFilters.java @@ -67,7 +67,7 @@ public class CLPlatformFilters { } /** - * Accepts all platforms containing at least one devices of which supports OpenGL-OpenCL interoparability. + * Accepts all platforms containing at least one devices of which supports OpenGL-OpenCL interoperability. */ public static Filter glSharing() { return new Filter() { @@ -86,7 +86,7 @@ public class CLPlatformFilters { /** * Accepts all with the given OpenGL context compatible platforms containing at least one - * devices of which supports OpenGL-OpenCL interoparability. + * devices of which supports OpenGL-OpenCL interoperability. */ public static Filter glSharing(final GLContext context) { return new Filter() { @@ -94,10 +94,22 @@ public class CLPlatformFilters { public boolean accept(CLPlatform item) { String glVendor = context.getGL().glGetString(GL.GL_VENDOR); String clVendor = item.getVendor(); - return clVendor.equals(glVendor) && glFilter.accept(item); + return areVendorsCompatible(glVendor,clVendor) && glFilter.accept(item); } }; } + + /** + * We need this test because on at least some AMD cards, the GL vendor is ATI, + * but the CL vendor is AMD. + * @param glVendor OpenGL vendor string. + * @param clVendor OpenCL vendor string. + * @return true if the strings are either the same, or indicate that they're part of the same card. + */ + private static boolean areVendorsCompatible(final String glVendor, final String clVendor) { + return( clVendor.equals(glVendor) + || (glVendor.contains("ATI Technologies") && clVendor.contains("Advanced Micro Devices"))); + } /** * Accepts all platforms supporting the given extensions. diff --git a/test/com/jogamp/opencl/gl/CLGLTest.java b/test/com/jogamp/opencl/gl/CLGLTest.java index e7906cec..8c729c08 100644 --- a/test/com/jogamp/opencl/gl/CLGLTest.java +++ b/test/com/jogamp/opencl/gl/CLGLTest.java @@ -35,6 +35,8 @@ package com.jogamp.opencl.gl; import com.jogamp.common.nio.Buffers; import com.jogamp.opencl.CLBuffer; import com.jogamp.opencl.CLCommandQueue; +import com.jogamp.opencl.CLKernel; +import com.jogamp.opencl.CLProgram; import javax.media.opengl.GL2; import javax.media.opengl.GLException; @@ -228,8 +230,7 @@ public class CLGLTest extends UITestCase { } -// @Test(timeout=15000) - @Test + @Test(timeout=15000) public void textureSharing() { out.println(" - - - glcl; textureSharing - - - "); @@ -241,20 +242,32 @@ public class CLGLTest extends UITestCase { assertTrue(glcontext.isCurrent()); @SuppressWarnings("unchecked") - CLPlatform platform = CLPlatform.getDefault(glSharing(glcontext)); - if(platform == null) { - out.println("test aborted"); + CLPlatform [] clplatforms = CLPlatform.listCLPlatforms(glSharing(glcontext)); + if(clplatforms.length == 0) { + out.println("no platform that supports OpenGL-OpenCL interoperability"); return; } - @SuppressWarnings("unchecked") - CLDevice device = platform.getMaxFlopsDevice(CLDeviceFilters.glSharing()); - out.println(device); + for(CLPlatform clplatform : clplatforms) { + + @SuppressWarnings("unchecked") + CLDevice [] cldevices = clplatform.listCLDevices(CLDeviceFilters.glSharing()); + + for(CLDevice cldevice : cldevices) { + out.println(cldevice); + textureSharingInner(cldevice); + } + } + + deinitGL(); + } - CLGLContext context = CLGLContext.create(glcontext, device); + public void textureSharingInner(CLDevice cldevice) { + + CLGLContext clglcontext = CLGLContext.create(glcontext, cldevice); try { - out.println(context); + out.println(clglcontext); GL2 gl = glcontext.getGL().getGL2(); @@ -263,39 +276,59 @@ public class CLGLTest extends UITestCase { gl.glGenTextures(id.length, id, 0); gl.glActiveTexture(GL2.GL_TEXTURE0); gl.glBindTexture (GL2.GL_TEXTURE_2D, id[0]); -// gl.glTexParameteri(GL2.GL_TEXTURE_2D, GL2.GL_TEXTURE_MAG_FILTER, GL2.GL_NEAREST); -// gl.glTexParameteri(GL2.GL_TEXTURE_2D, GL2.GL_TEXTURE_MIN_FILTER, GL2.GL_NEAREST); -// gl.glTexParameteri(GL2.GL_TEXTURE_2D, GL2.GL_TEXTURE_BASE_LEVEL, 0); -// gl.glTexParameteri(GL2.GL_TEXTURE_2D, GL2.GL_TEXTURE_MAX_LEVEL, 0); - - ByteBuffer bufferGL = Buffers.newDirectByteBuffer(new byte [] { - (byte)0, (byte)5, (byte)10, (byte)0xff, - (byte)15, (byte)20, (byte)25, (byte)0xff, - (byte)30, (byte)35, (byte)40, (byte)0xff, - (byte)45, (byte)50, (byte)55, (byte)0xff}); - bufferGL.rewind(); - gl.glTexImage2D(GL2.GL_TEXTURE_2D, 0, GL2.GL_RGBA, 2, 2, 0, GL2.GL_RGBA, GL2.GL_UNSIGNED_BYTE, bufferGL); + int texWidth = 2; + int texHeight = 2; + gl.glTexImage2D(GL2.GL_TEXTURE_2D, 0, GL2.GL_RGBA, texWidth, texHeight, 0, GL2.GL_RGBA, GL2.GL_UNSIGNED_BYTE, null ); gl.glBindTexture(GL2.GL_TEXTURE_2D, 0); gl.glFinish(); // create CLGL buffer - ByteBuffer bufferCL = Buffers.newDirectByteBuffer(2*2*4); - CLGLTexture2d clTexture = context.createFromGLTexture2d(bufferCL, GL2.GL_TEXTURE_2D, id[0], 0, CLBuffer.Mem.READ_ONLY); - -// assertEquals(bufferGL.capacity(), clTexture.getCLCapacity()); -// assertEquals(bufferGL.capacity(), clTexture.getCLSize()); - - CLCommandQueue queue = device.createCommandQueue(); - - // read gl buffer into cl nio buffer + ByteBuffer bufferCL = Buffers.newDirectByteBuffer(texWidth*texHeight*4); + CLGLTexture2d clTexture = clglcontext.createFromGLTexture2d(bufferCL, GL2.GL_TEXTURE_2D, id[0], 0, CLBuffer.Mem.READ_ONLY); + + // set texel values to a formula that can be read back and verified + String sourceCL = "__kernel void writeTexture (__write_only image2d_t imageTex, unsigned w, unsigned h ) \n" + + "{ \n" + + " for(int y=1; y<=h; ++y) { \n" + + " for(int x=1; x<=w; ++x) { \n" + + " write_imagef(imageTex, (int2)(x-1,y-1), (float4)(((float)x)/((float)(4*w)), ((float)y)/((float)(4*h)), 0.0f, 1.0f)); \n" + + " } \n" + + " } \n" + + "}"; + CLProgram program = clglcontext.createProgram(sourceCL); + program.build(); + System.out.println(program.getBuildStatus()); + System.out.println(program.getBuildLog()); + assertTrue(program.isExecutable()); + + CLKernel clkernel = program.createCLKernel("writeTexture") + .putArg(clTexture) + .putArg(texWidth) + .putArg(texHeight) + .rewind(); + + CLCommandQueue queue = cldevice.createCommandQueue(); + + // write gl texture with cl kernel, then read it to host buffer queue.putAcquireGLObject(clTexture) - .putReadImage(clTexture, true) - .putReleaseGLObject(clTexture); - - while(bufferCL.hasRemaining()) { - byte bGL = bufferGL.get(); - byte bCL = bufferCL.get(); - assertEquals(bGL, bCL); + .put1DRangeKernel(clkernel, 0, 1, 1) + .putReadImage(clTexture, true) + .putReleaseGLObject(clTexture) + .finish(); + + for(int y = 1; y <= texHeight; y++) { + for(int x = 1; x <= texWidth; x++) { + byte bX = bufferCL.get(); + byte bY = bufferCL.get(); + byte bZero = bufferCL.get(); + byte bMinusOne = bufferCL.get(); + byte bXCheck = (byte)(((float)x)/((float)(4*texWidth))*256); + byte bYCheck = (byte)(((float)y)/((float)(4*texHeight))*256); + assertEquals(bXCheck, bX); + assertEquals(bYCheck, bY); + assertEquals(0, bZero); + assertEquals(-1, bMinusOne); + } } out.println(clTexture); @@ -304,8 +337,7 @@ public class CLGLTest extends UITestCase { gl.glDeleteBuffers(1, id, 0); } finally { - context.release(); - deinitGL(); + clglcontext.release(); } } -- cgit v1.2.3 From f98767152049ac141e115dcbb6a6ac66f4831d6a Mon Sep 17 00:00:00 2001 From: Wade Walker Date: Sun, 6 Apr 2014 15:05:46 -0500 Subject: Fix CL-GL interoperability tests on Mac. Fixed detection of compatible interoperability platforms (was silently skipping platform because GL vendor was Nvidia, but CL vendor was Apple). Also fixed CL kernel syntax error about signed-unsigned comparison that ATI's driver on Windows didn't find, and fixed the CL memory object to be write-only instead of read-only (which ATI's Windows driver just ignored). --- src/com/jogamp/opencl/util/CLPlatformFilters.java | 8 +++++--- test/com/jogamp/opencl/gl/CLGLTest.java | 6 +++--- 2 files changed, 8 insertions(+), 6 deletions(-) (limited to 'test/com/jogamp/opencl/gl/CLGLTest.java') diff --git a/src/com/jogamp/opencl/util/CLPlatformFilters.java b/src/com/jogamp/opencl/util/CLPlatformFilters.java index 14e1507e..451c6b1f 100644 --- a/src/com/jogamp/opencl/util/CLPlatformFilters.java +++ b/src/com/jogamp/opencl/util/CLPlatformFilters.java @@ -100,15 +100,17 @@ public class CLPlatformFilters { } /** - * We need this test because on at least some AMD cards, the GL vendor is ATI, - * but the CL vendor is AMD. + * We need this test because: + * - On at least some AMD cards, the GL vendor is ATI, but the CL vendor is AMD. + * - On at least some Macs, the GL vendor is Nvidia, but the CL vendor is Apple. * @param glVendor OpenGL vendor string. * @param clVendor OpenCL vendor string. * @return true if the strings are either the same, or indicate that they're part of the same card. */ private static boolean areVendorsCompatible(final String glVendor, final String clVendor) { return( clVendor.equals(glVendor) - || (glVendor.contains("ATI Technologies") && clVendor.contains("Advanced Micro Devices"))); + || (glVendor.contains("ATI Technologies") && clVendor.contains("Advanced Micro Devices")) + || (glVendor.contains("NVIDIA Corporation") && clVendor.contains("Apple"))); } /** diff --git a/test/com/jogamp/opencl/gl/CLGLTest.java b/test/com/jogamp/opencl/gl/CLGLTest.java index 8c729c08..8c477002 100644 --- a/test/com/jogamp/opencl/gl/CLGLTest.java +++ b/test/com/jogamp/opencl/gl/CLGLTest.java @@ -284,13 +284,13 @@ public class CLGLTest extends UITestCase { // create CLGL buffer ByteBuffer bufferCL = Buffers.newDirectByteBuffer(texWidth*texHeight*4); - CLGLTexture2d clTexture = clglcontext.createFromGLTexture2d(bufferCL, GL2.GL_TEXTURE_2D, id[0], 0, CLBuffer.Mem.READ_ONLY); + CLGLTexture2d clTexture = clglcontext.createFromGLTexture2d(bufferCL, GL2.GL_TEXTURE_2D, id[0], 0, CLBuffer.Mem.WRITE_ONLY); // set texel values to a formula that can be read back and verified String sourceCL = "__kernel void writeTexture (__write_only image2d_t imageTex, unsigned w, unsigned h ) \n" + "{ \n" + - " for(int y=1; y<=h; ++y) { \n" + - " for(int x=1; x<=w; ++x) { \n" + + " for(unsigned y=1; y<=h; ++y) { \n" + + " for(unsigned x=1; x<=w; ++x) { \n" + " write_imagef(imageTex, (int2)(x-1,y-1), (float4)(((float)x)/((float)(4*w)), ((float)y)/((float)(4*h)), 0.0f, 1.0f)); \n" + " } \n" + " } \n" + -- cgit v1.2.3