From ac3d99f95b8953cdb5829c7536abc48d5fd65daf Mon Sep 17 00:00:00 2001
From: Andreas Kloeckner <inform@tiker.net>
Date: Wed, 23 Jun 2010 10:41:19 -0400
Subject: [PATCH] Add GL interop fixes and example. (Paolo Simone Gasparello)
---
doc/source/runtime.rst | 11 ++++++
examples/gl_interop_demo.py | 75 +++++++++++++++++++++++++++++++++++++
src/wrapper/wrap_cl.hpp | 6 ++-
3 files changed, 90 insertions(+), 2 deletions(-)
create mode 100644 examples/gl_interop_demo.py
diff --git a/doc/source/runtime.rst b/doc/source/runtime.rst
index 276ca34b..11ddf72c 100644
--- a/doc/source/runtime.rst
+++ b/doc/source/runtime.rst
@@ -119,6 +119,17 @@ Platforms, Devices and Contexts
:func:`create_some_context`. See, e.g. this
`explanation by AMD <http://developer.amd.com/support/KnowledgeBase/Lists/KnowledgeBase/DispForm.aspx?ID=71>`_.
+ .. note::
+
+ For
+ :attr:`context_properties.CL_GL_CONTEXT_KHR`,
+ :attr:`context_properties.CL_EGL_DISPLAY_KHR`,
+ :attr:`context_properties.CL_GLX_DISPLAY_KHR`,
+ :attr:`context_properties.CL_WGL_HDC_KHR`, and
+ :attr:`context_properties.CL_CGL_SHAREGROUP_KHR`
+ the value in the key-value pair is a PyOpenGL context or display
+ instance.
+
.. versionchanged:: 0.91.2
Constructor arguments *dev_type* added.
diff --git a/examples/gl_interop_demo.py b/examples/gl_interop_demo.py
new file mode 100644
index 00000000..e773db83
--- /dev/null
+++ b/examples/gl_interop_demo.py
@@ -0,0 +1,75 @@
+from OpenGL.GL import *
+from OpenGL.GLUT import *
+from OpenGL.raw.GL.VERSION.GL_1_5 import glBufferData as rawGlBufferData
+try:
+ from OpenGL.WGL import wglGetCurrentDisplay as GetCurrentDisplay, wglGetCurrentContext as GetCurrentContext
+except:
+ pass
+try:
+ from OpenGL.GLX import glXGetCurrentDisplay as GetCurrentDisplay, glXGetCurrentContext as GetCurrentContext
+except:
+ pass
+import pyopencl as cl
+
+
+n_vertices = 10000
+
+src = """
+
+__kernel void generate_sin(__global float2* a)
+{
+ int id = get_global_id(0);
+ int n = get_global_size(0);
+ float r = (float)id / (float)n;
+ float x = r * 16.0f * 3.1415f;
+ a[id].x = r * 2.0f - 1.0f;
+ a[id].y = native_sin(x);
+}
+
+"""
+
+def initialize():
+ plats = cl.get_platforms()
+ ctx_props = cl.context_properties
+ props = [(ctx_props.PLATFORM, plats[0]), (ctx_props.GL_CONTEXT_KHR,
+ GetCurrentContext()), (ctx_props.GLX_DISPLAY_KHR, GetCurrentDisplay())]
+ ctx = cl.Context(properties=props)
+ glClearColor(1, 1, 1, 1)
+ glColor(0, 0, 1)
+ vbo = glGenBuffers(1)
+ glBindBuffer(GL_ARRAY_BUFFER, vbo)
+ rawGlBufferData(GL_ARRAY_BUFFER, n_vertices * 2 * 4, None, GL_STATIC_DRAW)
+ glEnableClientState(GL_VERTEX_ARRAY)
+ glVertexPointer(2, GL_FLOAT, 0, None)
+ coords_dev = cl.GLBuffer(ctx, cl.mem_flags.READ_WRITE, int(vbo))
+ prog = cl.Program(ctx, src).build()
+ queue = cl.CommandQueue(ctx)
+ cl.enqueue_acquire_gl_objects(queue, [coords_dev])
+ prog.generate_sin(queue, (n_vertices,), coords_dev)
+ cl.enqueue_release_gl_objects(queue, [coords_dev])
+ queue.finish()
+ glFlush()
+
+def display():
+ glClear(GL_COLOR_BUFFER_BIT)
+ glDrawArrays(GL_LINE_STRIP, 0, n_vertices)
+ glFlush()
+
+def reshape(w, h):
+ glViewport(0, 0, w, h)
+ glMatrixMode(GL_PROJECTION)
+ glLoadIdentity()
+ glMatrixMode(GL_MODELVIEW)
+
+if __name__ == '__main__':
+ import sys
+ glutInit(sys.argv)
+ if len(sys.argv) > 1:
+ n_vertices = int(sys.argv[1])
+ glutInitWindowSize(800, 160)
+ glutInitWindowPosition(0, 0)
+ glutCreateWindow('OpenCL/OpenGL Interop Tutorial: Sin Generator')
+ glutDisplayFunc(display)
+ glutReshapeFunc(reshape)
+ initialize()
+ glutMainLoop()
diff --git a/src/wrapper/wrap_cl.hpp b/src/wrapper/wrap_cl.hpp
index ea0cbdaf..3d9a1914 100644
--- a/src/wrapper/wrap_cl.hpp
+++ b/src/wrapper/wrap_cl.hpp
@@ -693,7 +693,10 @@ namespace pyopencl
|| prop == CL_CGL_SHAREGROUP_KHR
)
{
- py::extract<cl_context_properties> value(prop_tuple[1]);
+ py::object ctypes = py::import("ctypes");
+ py::object prop = prop_tuple[1], c_void_p = ctypes.attr("c_void_p");
+ py::object ptr = ctypes.attr("cast")(prop, c_void_p);
+ py::extract<cl_context_properties> value(ptr.attr("value"));
props.push_back(value);
}
#endif
@@ -2849,7 +2852,6 @@ namespace pyopencl
std::vector<cl_context_properties> props
= parse_context_properties(py_properties);
-
typedef CL_API_ENTRY cl_int CL_API_CALL
(*func_ptr_type)(const cl_context_properties * /* properties */,
cl_gl_context_info /* param_name */,
--
GitLab