added OpenGL support to highgui under WIN32

CMakeLists.txt

@@ -462,6 +462,9 @@ endif()
 
 set(WITH_OPENNI OFF CACHE BOOL "Include OpenNI support")
 set(WITH_XIMEA OFF CACHE BOOL "Include XIMEA cameras support")
+set(WITH_OPENGL OFF CACHE BOOL "Include OpenGL support")
+
+set(HAVE_OPENGL 0)
 
 # ===================================================
 # Macros that checks if module have been installed.
@@ -872,7 +875,6 @@ endif()
 ############################### QT ################################
 
 set(WITH_QT OFF CACHE BOOL "Build with Qt Backend support")
-set(WITH_QT_OPENGL OFF CACHE BOOL "Add OpenGL extension to Qt")
 
 set(HAVE_QT 0)
 set(HAVE_QT_OPENGL 0)
@@ -885,8 +887,9 @@ if (WITH_QT)
         find_package (OpenGL QUIET)
 
         #if (NOT WIN32)
-        if (WITH_QT_OPENGL)
+        if (WITH_OPENGL)
             if (QT_QTOPENGL_FOUND AND OPENGL_FOUND)
+                #set(HAVE_OPENGL 1)
                 set(HAVE_QT_OPENGL 1)
                 add_definitions(-DHAVE_QT_OPENGL)
                 #link_directories("${OPENGL_LIBRARIES}")
@@ -1121,6 +1124,16 @@ if(WIN32)
             set(HIGHGUI_LIBRARIES ${HIGHGUI_LIBRARIES} vfw32 winmm)
         endif()  
     endif()
+
+    if (WITH_OPENGL)
+        find_package(OpenGL QUIET)
+        
+        if (OPENGL_FOUND)
+            set(HAVE_OPENGL 1)
+            set(OPENCV_LINKER_LIBS ${OPENCV_LINKER_LIBS} ${OPENGL_LIBRARIES})
+            include_directories(${OPENGL_INCLUDE_DIR})
+        endif()
+    endif()
 endif()
 
 ############## Android source tree for native camera ###############
@@ -1717,6 +1730,8 @@ else()
     endif()
 endif()
 
+status("    OpenGL support:" HAVE_OPENGL THEN YES ELSE NO)
+
 # media
 status("")
 status("  Media I/O: ")

android/CMakeCache.android.initial.cmake

@@ -55,8 +55,8 @@ set(WITH_PVAPI OFF CACHE BOOL "" )
 #Build with Qt Backend support
 set(WITH_QT OFF CACHE BOOL "" )
 
-#Add OpenGL extension to Qt
-set(WITH_QT_OPENGL OFF CACHE BOOL "" )
+#Add OpenGL support
+set(WITH_OPENGL OFF CACHE BOOL "" )
 
 #Include Intel TBB support
 set(WITH_TBB OFF CACHE BOOL "" )

cvconfig.h.cmake

@@ -195,3 +195,6 @@
 
 /* XIMEA camera support */
 #cmakedefine HAVE_XIMEA
+
+/* OpenGL support*/
+#cmakedefine HAVE_OPENGL

modules/core/include/opencv2/core/devmem2d.hpp

@@ -43,6 +43,8 @@
 #ifndef __OPENCV_CORE_DevMem2D_HPP__
 #define __OPENCV_CORE_DevMem2D_HPP__
 
+#ifdef __cplusplus
+
 #ifdef __CUDACC__ 
     #define __CV_GPU_HOST_DEVICE__ __host__ __device__ __forceinline__ 
 #else
@@ -154,4 +156,6 @@ namespace cv
     }    
 }
 
+#endif // __cplusplus
+
 #endif /* __OPENCV_GPU_DevMem2D_HPP__ */

modules/core/include/opencv2/core/gpumat.hpp

@@ -43,11 +43,16 @@
 #ifndef __OPENCV_GPUMAT_HPP__
 #define __OPENCV_GPUMAT_HPP__
 
+#ifdef __cplusplus
+
 #include "opencv2/core/core.hpp"
 #include "opencv2/core/devmem2d.hpp"
 
 namespace cv { namespace gpu
 {
+    ////////////////////////////////////////////////////////////////////////
+    // GpuMat
+
     //! Smart pointer for GPU memory with reference counting. Its interface is mostly similar with cv::Mat.
     class CV_EXPORTS GpuMat
     {
@@ -212,10 +217,158 @@ namespace cv { namespace gpu
     CV_EXPORTS void ensureSizeIsEnough(int rows, int cols, int type, GpuMat& m);
     CV_EXPORTS void ensureSizeIsEnough(Size size, int type, GpuMat& m);
 
-    //////////////////////////////// Error handling ////////////////////////
+    ////////////////////////////////////////////////////////////////////////
+    // OpenGL
+
+    //! set a CUDA device to use OpenGL interoperability
+    CV_EXPORTS void setGlDevice(int device = 0);
+
+    //! Smart pointer for OpenGL buffer memory with reference counting.
+    class CV_EXPORTS GlBuffer
+    {
+    public:
+        enum Usage
+        {
+            ARRAY_BUFFER = 0x8892,  // buffer will use for OpenGL arrays (vertices, colors, normals, etc)
+            TEXTURE_BUFFER = 0x88EC // buffer will ise for OpenGL textures
+        };
+
+        //! create empty buffer
+        explicit GlBuffer(Usage usage);
+
+        //! create buffer
+        GlBuffer(int rows, int cols, int type, Usage usage);
+        GlBuffer(Size size, int type, Usage usage);
+
+        //! copy from host/device memory
+        GlBuffer(const Mat& mat, Usage usage);
+        GlBuffer(const GpuMat& d_mat, Usage usage);
+
+        ~GlBuffer();
+
+        void create(int rows, int cols, int type, Usage usage);
+        inline void create(Size size, int type, Usage usage) { create(size.height, size.width, type, usage); }
+        inline void create(int rows, int cols, int type) { create(rows, cols, type, usage()); }
+        inline void create(Size size, int type) { create(size.height, size.width, type, usage()); }
+
+        void release();
+
+        //! copy from host/device memory
+        void copyFrom(const Mat& mat);
+        void copyFrom(const GpuMat& d_mat);
+
+        void bind() const;
+        void unbind() const;
+
+        //! map to host memory
+        Mat mapHost();
+        void unmapHost();
+
+        //! map to device memory
+        GpuMat mapDevice();
+        void unmapDevice();
+
+        int rows() const;
+        int cols() const;
+        Size size() const;
+        bool empty() const;
+
+        int type() const;
+        int depth() const;
+        int channels() const;
+        int elemSize() const;
+        int elemSize1() const;
+
+        Usage usage() const;
+
+    private:
+        class Impl;
+        Ptr<Impl> impl_;
+    };
+
+    //! Smart pointer for OpenGL 2d texture memory with reference counting.
+    class CV_EXPORTS GlTexture
+    {
+    public:
+        //! create empty texture
+        GlTexture();
+
+        //! create texture
+        GlTexture(int rows, int cols, int type);
+        GlTexture(Size size, int type);
 
-    CV_EXPORTS void error(const char *error_string, const char *file, const int line, const char *func);
+        //! copy from host/device memory
+        explicit GlTexture(const Mat& mat, bool bgra = true);
+        explicit GlTexture(const GlBuffer& buf, bool bgra = true);
 
+        ~GlTexture();
+
+        void create(int rows, int cols, int type);
+        inline void create(Size size, int type) { create(size.height, size.width, type); }
+        void release();
+
+        //! copy from host/device memory
+        void copyFrom(const Mat& mat, bool bgra = true);
+        void copyFrom(const GlBuffer& buf, bool bgra = true);
+
+        void bind() const;
+        void unbind() const;
+
+        int rows() const;
+        int cols() const;
+        Size size() const;
+        bool empty() const;
+
+        int type() const;
+        int depth() const;
+        int channels() const;
+        int elemSize() const;
+        int elemSize1() const;
+
+    private:
+        class Impl;
+        Ptr<Impl> impl_;
+    };
+
+    //! render functions
+    CV_EXPORTS void render(const GlTexture& tex);
+
+    //! OpenGL extension table
+    class CV_EXPORTS GlFuncTab
+    {
+    public:
+        virtual ~GlFuncTab() {}
+
+        virtual void genBuffers(int n, unsigned int* buffers) const = 0;        
+        virtual void deleteBuffers(int n, const unsigned int* buffers) const = 0;
+
+        virtual void bufferData(unsigned int target, ptrdiff_t size, const void* data, unsigned int usage) const = 0;
+        virtual void bufferSubData(unsigned int target, ptrdiff_t offset, ptrdiff_t size, const void* data) const = 0;
+
+        virtual void bindBuffer(unsigned int target, unsigned int buffer) const = 0;
+
+        virtual void* mapBuffer(unsigned int target, unsigned int access) const = 0;
+        virtual void unmapBuffer(unsigned int target) const = 0;
+
+        virtual bool isGlContextInitialized() const = 0;
+    };
+
+    CV_EXPORTS void setGlFuncTab(const GlFuncTab* tab);
+
+    ////////////////////////////////////////////////////////////////////////
+    // Error handling
+
+    CV_EXPORTS void error(const char* error_string, const char* file, const int line, const char* func = "");
+    CV_EXPORTS bool checkGlError(const char* file, const int line, const char* func = "");
+
+    #if defined(__GNUC__)
+        #define CV_CheckGlError() CV_DbgAssert( (cv::gpu::checkGlError(__FILE__, __LINE__, __func__)) )
+    #else
+        #define CV_CheckGlError() CV_DbgAssert( (cv::gpu::checkGlError(__FILE__, __LINE__)) )
+    #endif
+
+    ////////////////////////////////////////////////////////////////////////
+    ////////////////////////////////////////////////////////////////////////
     ////////////////////////////////////////////////////////////////////////
 
     inline GpuMat::GpuMat() 
@@ -456,4 +609,6 @@ namespace cv { namespace gpu
     }
 }}
 
+#endif // __cplusplus
+
 #endif // __OPENCV_GPUMAT_HPP__

modules/core/include/opencv2/core/types_c.h

@@ -250,7 +250,9 @@ enum {
  CV_StsBadMemBlock=            -214, /* an allocated block has been corrupted */
  CV_StsAssert=                 -215, /* assertion failed */    
  CV_GpuNotSupported=           -216,  
- CV_GpuApiCallError=           -217
+ CV_GpuApiCallError=           -217,
+ CV_OpenGlNotSupported=        -218,
+ CV_OpenGlApiCallError=        -219
 };
 
 /****************************************************************************************\

modules/core/src/cuda/matrix_operations.cu

@@ -106,7 +106,7 @@ namespace cv { namespace gpu { namespace device
         CopyToFunc func = tab[depth];
 
         if (func == 0) 
-            cv::gpu::error("Unsupported copyTo operation", __FILE__, __LINE__);
+            cv::gpu::error("Unsupported copyTo operation", __FILE__, __LINE__, "copy_to_with_mask");
 
         func(mat_src, mat_dst, mask, channels, stream);
     }
@@ -246,7 +246,7 @@ namespace cv { namespace gpu { namespace device
             return saturate_cast<D>(alpha * src + beta);
         }
 
-        const double alpha, beta;
+        double alpha, beta;
     };
 
     namespace detail
@@ -338,7 +338,7 @@ namespace cv { namespace gpu { namespace device
 
         caller_t func = tab[sdepth][ddepth];
         if (!func)
-            cv::gpu::error("Unsupported convert operation", __FILE__, __LINE__);
+            cv::gpu::error("Unsupported convert operation", __FILE__, __LINE__, "convert_gpu");
 
         func(src, dst, alpha, beta, stream);
     }

modules/core/src/gpumat.cpp

@@ -44,17 +44,27 @@
 #include "opencv2/core/gpumat.hpp"
 
 #include <iostream>
-#include <sstream>
 
 #ifdef HAVE_CUDA
     #include <cuda_runtime.h>
     #include <npp.h>
 #endif
 
+#ifdef HAVE_OPENGL
+    #include <GL/gl.h>
+
+    #ifdef HAVE_CUDA
+        #include <cuda_gl_interop.h>
+    #endif
+#endif
+
 using namespace std;
 using namespace cv;
 using namespace cv::gpu;
 
+////////////////////////////////////////////////////////////////////////
+// GpuMat
+
 cv::gpu::GpuMat::GpuMat(const GpuMat& m) 
     : flags(m.flags), rows(m.rows), cols(m.cols), step(m.step), data(m.data), refcount(m.refcount), datastart(m.datastart), dataend(m.dataend)
 {
@@ -932,6 +942,1074 @@ void cv::gpu::GpuMat::release()
     refcount = 0;
 }
 
+////////////////////////////////////////////////////////////////////////
+// OpenGL
+
+namespace
+{
+    void throw_nogl() 
+    {
+    #ifndef HAVE_OPENGL
+        CV_Error(CV_OpenGlNotSupported, "The library is compiled without OpenGL support"); 
+    #else
+        CV_Error(CV_OpenGlNotSupported, "OpenGL context doesn't exist"); 
+    #endif
+    }
+
+    class EmptyGlFuncTab : public GlFuncTab
+    {
+    public:
+        void genBuffers(int, unsigned int*) const { throw_nogl(); }
+        void deleteBuffers(int, const unsigned int*) const { throw_nogl(); }
+
+        void bufferData(unsigned int, ptrdiff_t, const void*, unsigned int) const { throw_nogl(); }
+        void bufferSubData(unsigned int, ptrdiff_t, ptrdiff_t, const void*) const { throw_nogl(); }
+
+        void bindBuffer(unsigned int, unsigned int) const { throw_nogl(); }
+
+        void* mapBuffer(unsigned int, unsigned int) const { throw_nogl(); return 0; }
+        void unmapBuffer(unsigned int) const { throw_nogl(); }
+
+        bool isGlContextInitialized() const { return false; }
+    };
+
+    const GlFuncTab* g_glFuncTab = 0;
+
+    const GlFuncTab* glFuncTab()
+    {
+        static EmptyGlFuncTab empty;
+        return g_glFuncTab ? g_glFuncTab : &empty;
+    }
+}
+
+void cv::gpu::setGlFuncTab(const GlFuncTab* tab)
+{
+    g_glFuncTab = tab;
+}
+
+#ifdef HAVE_OPENGL
+    #ifndef GL_DYNAMIC_DRAW
+        #define GL_DYNAMIC_DRAW 0x88E8
+    #endif
+
+    #ifndef GL_READ_WRITE
+        #define GL_READ_WRITE 0x88BA
+    #endif
+
+    #ifndef GL_BGR
+        #define GL_BGR 0x80E0
+    #endif
+
+    #ifndef GL_BGRA
+        #define GL_BGRA 0x80E1
+    #endif
+
+    namespace
+    {    
+        const GLenum gl_types[] = {GL_UNSIGNED_BYTE, GL_BYTE, GL_UNSIGNED_SHORT, GL_SHORT, GL_INT, GL_FLOAT, GL_DOUBLE};
+
+    #ifdef HAVE_CUDA
+        bool g_isCudaGlDeviceInitialized = false;
+    #endif
+    }
+#endif // HAVE_OPENGL
+
+
+
+void cv::gpu::setGlDevice(int device)
+{
+#ifndef HAVE_CUDA
+    throw_nogpu();
+#else
+    #ifndef HAVE_OPENGL
+        throw_nogl();
+    #else
+        if (!glFuncTab()->isGlContextInitialized())
+            throw_nogl();
+
+        cudaSafeCall( cudaGLSetGLDevice(device) ); 
+
+        g_isCudaGlDeviceInitialized = true;
+    #endif
+#endif
+}
+
+////////////////////////////////////////////////////////////////////////
+// CudaGlInterop
+
+#if defined HAVE_CUDA && defined HAVE_OPENGL
+namespace
+{
+    class CudaGlInterop
+    {
+    public:
+        CudaGlInterop() : resource_(0) 
+        {
+        }
+
+        ~CudaGlInterop() 
+        { 
+            if (resource_)
+            {
+                cudaGraphicsUnregisterResource(resource_);
+                resource_ = 0;
+            } 
+        }
+
+        void registerBuffer(unsigned int buffer)
+        {
+            if (!g_isCudaGlDeviceInitialized)
+                cvError(CV_GpuApiCallError, "registerBuffer", "cuda GL device wasn't initialized, call setGlDevice", __FILE__, __LINE__);
+
+            cudaGraphicsResource_t resource;
+            cudaSafeCall( cudaGraphicsGLRegisterBuffer(&resource, buffer, cudaGraphicsMapFlagsNone) );
+
+            resource_ = resource;
+        }
+
+        void copyFrom(const GpuMat& mat, cudaStream_t stream = 0)
+        {
+            CV_Assert(resource_ != 0);
+
+            cudaSafeCall( cudaGraphicsMapResources(1, &resource_, stream) );
+
+            void* dst_ptr;
+            size_t num_bytes;
+            cudaSafeCall( cudaGraphicsResourceGetMappedPointer(&dst_ptr, &num_bytes, resource_) );
+            
+            const void* src_ptr = mat.ptr();
+            size_t widthBytes = mat.cols * mat.elemSize();
+
+            CV_Assert(widthBytes * mat.rows <= num_bytes);
+
+            if (stream == 0)
+                cudaSafeCall( cudaMemcpy2D(dst_ptr, widthBytes, src_ptr, mat.step, widthBytes, mat.rows, cudaMemcpyDeviceToDevice) );
+            else
+                cudaSafeCall( cudaMemcpy2DAsync(dst_ptr, widthBytes, src_ptr, mat.step, widthBytes, mat.rows, cudaMemcpyDeviceToDevice, stream) );
+
+            cudaGraphicsUnmapResources(1, &resource_, stream);
+        }
+
+        GpuMat map(int rows, int cols, int type, cudaStream_t stream = 0)
+        {
+            CV_Assert(resource_ != 0);
+
+            cudaSafeCall( cudaGraphicsMapResources(1, &resource_, stream) );
+
+            void* ptr;
+            size_t num_bytes;
+            cudaSafeCall( cudaGraphicsResourceGetMappedPointer(&ptr, &num_bytes, resource_) );
+
+            CV_Assert( static_cast<size_t>(cols) * CV_ELEM_SIZE(type) * rows <= num_bytes );
+
+            return GpuMat(rows, cols, type, ptr);
+        }
+
+        void unmap(cudaStream_t stream = 0)
+        {
+            cudaGraphicsUnmapResources(1, &resource_, stream);
+        }
+
+    private:
+        cudaGraphicsResource_t resource_;
+    };
+}
+#endif // HAVE_CUDA && HAVE_OPENGL
+
+////////////////////////////////////////////////////////////////////////
+// GlBuffer
+
+#ifndef HAVE_OPENGL
+
+class cv::gpu::GlBuffer::Impl
+{
+};
+
+#else
+
+class cv::gpu::GlBuffer::Impl
+{
+public:
+    explicit Impl(unsigned int target) : rows_(0), cols_(0), type_(0), target_(target), buffer_(0) 
+    {
+    }
+
+    Impl(int rows, int cols, int type, unsigned int target);
+
+    Impl(const Mat& m, unsigned int target);
+
+    ~Impl();
+
+    void copyFrom(const Mat& m);
+
+#ifdef HAVE_CUDA
+    void copyFrom(const GpuMat& mat, cudaStream_t stream = 0);
+#endif
+
+    void bind() const;
+    void unbind() const;
+
+    Mat mapHost();
+    void unmapHost();
+
+#ifdef HAVE_CUDA
+    GpuMat mapDevice(cudaStream_t stream = 0);
+    void unmapDevice(cudaStream_t stream = 0);
+#endif
+
+    int rows() const { return rows_; }
+    int cols() const { return cols_; }
+    int type() const { return type_; }    
+    int target() const { return target_; }
+
+private:
+    int rows_;
+    int cols_;
+    int type_;
+
+    unsigned int target_;
+
+    unsigned int buffer_;
+
+#ifdef HAVE_CUDA
+    CudaGlInterop cudaGlInterop_;
+#endif
+};
+
+cv::gpu::GlBuffer::Impl::Impl(int rows, int cols, int type, unsigned int target) : rows_(0), cols_(0), type_(0), target_(target), buffer_(0) 
+{ 
+    if (!glFuncTab()->isGlContextInitialized())
+        throw_nogl();
+
+    CV_DbgAssert(rows > 0 && cols > 0);
+    CV_DbgAssert(CV_MAT_DEPTH(type) >= 0 && CV_MAT_DEPTH(type) <= CV_64F);
+
+    unsigned int buffer;    
+    glFuncTab()->genBuffers(1, &buffer);
+    CV_CheckGlError();
+
+    size_t size = rows * cols * CV_ELEM_SIZE(type);
+
+    glFuncTab()->bindBuffer(target_, buffer);
+    CV_CheckGlError();
+
+    glFuncTab()->bufferData(target_, size, 0, GL_DYNAMIC_DRAW);
+    CV_CheckGlError();
+
+    glFuncTab()->bindBuffer(target_, 0);
+    
+#ifdef HAVE_CUDA
+    if (g_isCudaGlDeviceInitialized)
+        cudaGlInterop_.registerBuffer(buffer);
+#endif
+
+    rows_ = rows;
+    cols_ = cols;
+    type_ = type;
+    buffer_ = buffer;
+}
+
+cv::gpu::GlBuffer::Impl::Impl(const Mat& m, unsigned int target) : rows_(0), cols_(0), type_(0), target_(target), buffer_(0) 
+{ 
+    if (!glFuncTab()->isGlContextInitialized())
+        throw_nogl();
+
+    CV_DbgAssert(m.rows > 0 && m.cols > 0);
+    CV_DbgAssert(m.depth() >= 0 && m.depth() <= CV_64F);
+    CV_Assert(m.isContinuous());
+
+    unsigned int buffer;  
+    glFuncTab()->genBuffers(1, &buffer);
+    CV_CheckGlError();
+
+    size_t size = m.rows * m.cols * m.elemSize();
+
+    glFuncTab()->bindBuffer(target_, buffer);
+    CV_CheckGlError();
+
+    glFuncTab()->bufferData(target_, size, m.data, GL_DYNAMIC_DRAW);
+    CV_CheckGlError();
+
+    glFuncTab()->bindBuffer(target_, 0);
+    
+#ifdef HAVE_CUDA
+    if (g_isCudaGlDeviceInitialized)
+        cudaGlInterop_.registerBuffer(buffer);
+#endif
+
+    rows_ = m.rows;
+    cols_ = m.cols;
+    type_ = m.type();
+    buffer_ = buffer;
+}
+
+cv::gpu::GlBuffer::Impl::~Impl() 
+{ 
+    try
+    {
+        if (buffer_)
+            glFuncTab()->deleteBuffers(1, &buffer_);
+    }
+#ifdef _DEBUG
+    catch(const exception& e)
+    {
+        cerr << e.what() << endl;
+    }
+#endif
+    catch(...)
+    {
+    }
+}
+
+void cv::gpu::GlBuffer::Impl::copyFrom(const Mat& m)
+{
+    CV_Assert(buffer_ != 0);
+
+    CV_DbgAssert(rows_ == m.rows && cols_ == m.cols && type_ == m.type());
+    CV_Assert(m.isContinuous());
+
+    bind();
+
+    size_t size = m.rows * m.cols * m.elemSize();
+
+    glFuncTab()->bufferSubData(target_, 0, size, m.data);
+    CV_CheckGlError();
+
+    unbind();
+}
+
+#ifdef HAVE_CUDA
+
+void cv::gpu::GlBuffer::Impl::copyFrom(const GpuMat& mat, cudaStream_t stream) 
+{
+    if (!g_isCudaGlDeviceInitialized)
+        cvError(CV_GpuApiCallError, "copyFrom", "cuda GL device wasn't initialized, call setGlDevice", __FILE__, __LINE__);
+
+    CV_Assert(buffer_ != 0);
+
+    CV_DbgAssert(rows_ == mat.rows && cols_ == mat.cols && type_ == mat.type());
+
+    cudaGlInterop_.copyFrom(mat, stream);
+}
+
+#endif // HAVE_CUDA
+
+void cv::gpu::GlBuffer::Impl::bind() const 
+{
+    CV_Assert(buffer_ != 0);
+
+    glFuncTab()->bindBuffer(target_, buffer_); 
+    CV_CheckGlError();
+}
+
+void cv::gpu::GlBuffer::Impl::unbind() const
+{ 
+    glFuncTab()->bindBuffer(target_, 0);
+}
+
+Mat cv::gpu::GlBuffer::Impl::mapHost()
+{
+    void* ptr = glFuncTab()->mapBuffer(target_, GL_READ_WRITE);
+    CV_CheckGlError();
+
+    return Mat(rows_, cols_, type_, ptr);
+}
+
+void cv::gpu::GlBuffer::Impl::unmapHost()
+{
+    glFuncTab()->unmapBuffer(target_);
+}
+
+#ifdef HAVE_CUDA
+
+GpuMat cv::gpu::GlBuffer::Impl::mapDevice(cudaStream_t stream)
+{
+    if (!g_isCudaGlDeviceInitialized)
+        cvError(CV_GpuApiCallError, "copyFrom", "cuda GL device wasn't initialized, call setGlDevice", __FILE__, __LINE__);
+
+    CV_Assert(buffer_ != 0);
+
+    return cudaGlInterop_.map(rows_, cols_, type_, stream);
+}
+
+void cv::gpu::GlBuffer::Impl::unmapDevice(cudaStream_t stream)
+{
+    if (!g_isCudaGlDeviceInitialized)
+        cvError(CV_GpuApiCallError, "copyFrom", "cuda GL device wasn't initialized, call setGlDevice", __FILE__, __LINE__);
+
+    cudaGlInterop_.unmap(stream);
+}
+
+#endif // HAVE_CUDA
+
+#endif // HAVE_OPENGL
+
+cv::gpu::GlBuffer::GlBuffer(Usage usage)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl(usage);
+#endif
+}
+
+cv::gpu::GlBuffer::GlBuffer(int rows, int cols, int type, Usage usage)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl(rows, cols, type, usage);
+#endif
+}
+
+cv::gpu::GlBuffer::GlBuffer(Size size, int type, Usage usage)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl(size.height, size.width, type, usage);
+#endif
+}
+
+cv::gpu::GlBuffer::GlBuffer(const Mat& mat, Usage usage)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl(mat, usage);
+#endif
+}
+
+cv::gpu::GlBuffer::GlBuffer(const GpuMat& d_mat, Usage usage)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    #ifndef HAVE_CUDA
+        throw_nogpu();
+    #else
+        impl_ = new Impl(d_mat.rows, d_mat.cols, d_mat.type(), usage);
+        impl_->copyFrom(d_mat);
+    #endif
+#endif
+}
+
+cv::gpu::GlBuffer::~GlBuffer()
+{
+}
+
+void cv::gpu::GlBuffer::create(int rows_, int cols_, int type_, Usage usage_)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    if (rows_ != rows() || cols_ != cols() || type_ != type() || usage_ != usage())
+        impl_ = new Impl(rows_, cols_, type_, usage_);
+#endif
+}
+
+void cv::gpu::GlBuffer::release()
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl(usage());
+#endif
+}
+
+void cv::gpu::GlBuffer::copyFrom(const Mat& mat)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    create(mat.rows, mat.cols, mat.type());
+    impl_->copyFrom(mat);
+#endif
+}
+
+void cv::gpu::GlBuffer::copyFrom(const GpuMat& d_mat)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    #ifndef HAVE_CUDA
+        throw_nogpu();
+    #else
+        create(d_mat.rows, d_mat.cols, d_mat.type());
+        impl_->copyFrom(d_mat);
+    #endif
+#endif
+}
+
+void cv::gpu::GlBuffer::bind() const
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_->bind();
+#endif
+}
+
+void cv::gpu::GlBuffer::unbind() const
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_->unbind();
+#endif
+}
+
+Mat cv::gpu::GlBuffer::mapHost()
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+    return Mat();
+#else
+    return impl_->mapHost();
+#endif
+}
+
+void cv::gpu::GlBuffer::unmapHost()
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_->unmapHost();
+#endif
+}
+
+GpuMat cv::gpu::GlBuffer::mapDevice()
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+    return GpuMat();
+#else
+    #ifndef HAVE_CUDA
+        throw_nogpu();
+        return GpuMat();
+    #else
+        return impl_->mapDevice();
+    #endif
+#endif
+}
+
+void cv::gpu::GlBuffer::unmapDevice()
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    #ifndef HAVE_CUDA
+        throw_nogpu();
+    #else
+        impl_->unmapDevice();
+    #endif
+#endif
+}
+
+int cv::gpu::GlBuffer::rows() const
+{
+#ifndef HAVE_OPENGL
+    return 0;
+#else
+    return impl_->rows();
+#endif
+}
+
+int cv::gpu::GlBuffer::cols() const
+{
+#ifndef HAVE_OPENGL
+    return 0;
+#else
+    return impl_->cols();
+#endif
+}
+
+Size cv::gpu::GlBuffer::size() const
+{
+    return Size(cols(), rows());
+}
+
+bool cv::gpu::GlBuffer::empty() const
+{
+    return rows() == 0 || cols() == 0;
+}
+
+int cv::gpu::GlBuffer::type() const
+{
+#ifndef HAVE_OPENGL
+    return 0;
+#else
+    return impl_->type();
+#endif
+}
+
+int cv::gpu::GlBuffer::depth() const 
+{ 
+    return CV_MAT_DEPTH(type()); 
+}
+
+int cv::gpu::GlBuffer::channels() const 
+{ 
+    return CV_MAT_CN(type()); 
+}
+
+int cv::gpu::GlBuffer::elemSize() const 
+{ 
+    return CV_ELEM_SIZE(type()); 
+}
+
+int cv::gpu::GlBuffer::elemSize1() const 
+{ 
+    return CV_ELEM_SIZE1(type()); 
+}
+
+GlBuffer::Usage cv::gpu::GlBuffer::usage() const
+{
+#ifndef HAVE_OPENGL
+    return ARRAY_BUFFER;
+#else
+    return static_cast<Usage>(impl_->target());
+#endif
+}
+
+//////////////////////////////////////////////////////////////////////////////////////////
+// GlTexture
+
+#ifndef HAVE_OPENGL
+
+class cv::gpu::GlTexture::Impl
+{
+};
+
+#else
+
+class cv::gpu::GlTexture::Impl
+{
+public:
+    Impl();
+
+    Impl(int rows, int cols, int type);
+
+    Impl(const Mat& mat, bool bgra);
+    Impl(const GlBuffer& buf, bool bgra);
+
+    ~Impl();
+
+    void copyFrom(const Mat& mat, bool bgra);
+    void copyFrom(const GlBuffer& buf, bool bgra);
+
+    void bind() const;
+    void unbind() const;
+
+    int rows() const { return rows_; }
+    int cols() const { return cols_; }
+    int type() const { return type_; }
+
+private:    
+    int rows_;
+    int cols_;
+    int type_;
+    unsigned int tex_;
+};
+
+cv::gpu::GlTexture::Impl::Impl() : rows_(0), cols_(0), type_(0), tex_(0)
+{
+}
+
+cv::gpu::GlTexture::Impl::Impl(int rows, int cols, int type) : rows_(0), cols_(0), type_(0), tex_(0)
+{
+    if (!glFuncTab()->isGlContextInitialized())
+        throw_nogl();
+
+    int depth = CV_MAT_DEPTH(type);
+    int cn = CV_MAT_CN(type);
+
+    CV_DbgAssert(rows > 0 && cols > 0);
+    CV_Assert(cn == 1 || cn == 3 || cn == 4);
+    CV_Assert(depth >= 0 && depth <= CV_32F);
+
+    GLuint tex;
+    glGenTextures(1, &tex);
+    CV_CheckGlError();
+
+    glBindTexture(GL_TEXTURE_2D, tex);
+    CV_CheckGlError();
+
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_REPEAT);
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_REPEAT);
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR);
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR); 
+    CV_CheckGlError();
+
+    GLenum format = cn == 1 ? GL_LUMINANCE : cn == 3 ? GL_BGR : GL_BGRA;
+
+    glPixelStorei(GL_UNPACK_ALIGNMENT, 1);
+    CV_CheckGlError();
+
+    glTexImage2D(GL_TEXTURE_2D, 0, cn, cols, rows, 0, format, gl_types[depth], 0);
+    CV_CheckGlError();
+
+    rows_ = rows;
+    cols_ = cols;
+    type_ = type;
+    tex_ = tex;
+}
+
+cv::gpu::GlTexture::Impl::Impl(const Mat& mat, bool bgra) : rows_(0), cols_(0), type_(0), tex_(0)
+{
+    if (!glFuncTab()->isGlContextInitialized())
+        throw_nogl();
+
+    int depth = mat.depth();
+    int cn = mat.channels();
+
+    CV_DbgAssert(mat.rows > 0 && mat.cols > 0);
+    CV_Assert(cn == 1 || cn == 3 || cn == 4);
+    CV_Assert(depth >= 0 && depth <= CV_32F);
+    CV_Assert(mat.isContinuous());
+
+    GLuint tex;
+    glGenTextures(1, &tex);
+    CV_CheckGlError();
+
+    glBindTexture(GL_TEXTURE_2D, tex);
+    CV_CheckGlError();
+
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_REPEAT);
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_REPEAT);
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR);
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR); 
+    CV_CheckGlError();
+
+    GLenum format = cn == 1 ? GL_LUMINANCE : (cn == 3 ? (bgra ? GL_BGR : GL_RGB) : (bgra ? GL_BGRA : GL_RGBA));
+
+    glPixelStorei(GL_UNPACK_ALIGNMENT, 1);
+    CV_CheckGlError();
+
+    glTexImage2D(GL_TEXTURE_2D, 0, cn, mat.cols, mat.rows, 0, format, gl_types[depth], mat.data);
+    CV_CheckGlError();
+
+    rows_ = mat.rows;
+    cols_ = mat.cols;
+    type_ = mat.type();
+    tex_ = tex;
+}
+
+cv::gpu::GlTexture::Impl::Impl(const GlBuffer& buf, bool bgra) : rows_(0), cols_(0), type_(0), tex_(0)
+{
+    if (!glFuncTab()->isGlContextInitialized())
+        throw_nogl();
+
+    int depth = buf.depth();
+    int cn = buf.channels();
+
+    CV_DbgAssert(buf.rows() > 0 && buf.cols() > 0);
+    CV_Assert(cn == 1 || cn == 3 || cn == 4);
+    CV_Assert(depth >= 0 && depth <= CV_32F);
+    CV_Assert(buf.usage() == GlBuffer::TEXTURE_BUFFER);
+
+    GLuint tex;
+    glGenTextures(1, &tex);
+    CV_CheckGlError();
+
+    glBindTexture(GL_TEXTURE_2D, tex);
+    CV_CheckGlError();
+
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_REPEAT);
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_REPEAT);
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR);
+    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR); 
+    CV_CheckGlError();
+
+    GLenum format = cn == 1 ? GL_LUMINANCE : (cn == 3 ? (bgra ? GL_BGR : GL_RGB) : (bgra ? GL_BGRA : GL_RGBA));
+
+    buf.bind();
+
+    glPixelStorei(GL_UNPACK_ALIGNMENT, 1);
+    CV_CheckGlError();
+
+    glTexImage2D(GL_TEXTURE_2D, 0, cn, buf.cols(), buf.rows(), 0, format, gl_types[depth], 0);
+    CV_CheckGlError();
+
+    buf.unbind();
+
+    rows_ = buf.rows();
+    cols_ = buf.cols();
+    type_ = buf.type();
+    tex_ = tex;
+}
+
+cv::gpu::GlTexture::Impl::~Impl()
+{
+    if (tex_)
+        glDeleteTextures(1, &tex_);
+}
+
+void cv::gpu::GlTexture::Impl::copyFrom(const Mat& mat, bool bgra)
+{
+    CV_Assert(tex_ != 0);
+    CV_DbgAssert(mat.cols == cols_ && mat.rows == rows_ && mat.type() == type_);
+
+    glPixelStorei(GL_UNPACK_ALIGNMENT, 1);
+    CV_CheckGlError();
+
+    int cn = mat.channels();
+    GLenum format = cn == 1 ? GL_LUMINANCE : (cn == 3 ? (bgra ? GL_BGR : GL_RGB) : (bgra ? GL_BGRA : GL_RGBA));
+
+    glTexSubImage2D(GL_TEXTURE_2D, 0, 0, 0, cols_, rows_, format, gl_types[mat.depth()], mat.data);
+    CV_CheckGlError();
+}
+
+void cv::gpu::GlTexture::Impl::copyFrom(const GlBuffer& buf, bool bgra)
+{
+    CV_Assert(tex_ != 0);
+    CV_DbgAssert(buf.cols() == cols_ && buf.rows() == rows_ && buf.type() == type_);
+    CV_Assert(buf.usage() == GlBuffer::TEXTURE_BUFFER);
+
+    buf.bind();
+
+    glPixelStorei(GL_UNPACK_ALIGNMENT, 1);
+    CV_CheckGlError();
+
+    int cn = buf.channels();
+    GLenum format = cn == 1 ? GL_LUMINANCE : (cn == 3 ? (bgra ? GL_BGR : GL_RGB) : (bgra ? GL_BGRA : GL_RGBA));
+
+    glTexSubImage2D(GL_TEXTURE_2D, 0, 0, 0, cols_, rows_, format, gl_types[buf.depth()], 0);
+    CV_CheckGlError();
+
+    buf.unbind();
+}
+
+void cv::gpu::GlTexture::Impl::bind() const
+{
+    CV_Assert(tex_ != 0);
+
+    glEnable(GL_TEXTURE_2D);
+    CV_CheckGlError();
+
+    glBindTexture(GL_TEXTURE_2D, tex_);
+    CV_CheckGlError();
+}
+
+void cv::gpu::GlTexture::Impl::unbind() const
+{
+    glBindTexture(GL_TEXTURE_2D, 0);
+
+    glDisable(GL_TEXTURE_2D);
+}
+
+#endif // HAVE_OPENGL
+
+cv::gpu::GlTexture::GlTexture()
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl;
+#endif
+}
+
+cv::gpu::GlTexture::GlTexture(int rows, int cols, int type)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl(rows, cols, type);
+#endif
+}
+
+cv::gpu::GlTexture::GlTexture(Size size, int type)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl(size.height, size.width, type);
+#endif
+}
+
+cv::gpu::GlTexture::GlTexture(const Mat& mat, bool bgra)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl(mat, bgra);
+#endif
+}
+
+cv::gpu::GlTexture::GlTexture(const GlBuffer& buf, bool bgra)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl(buf, bgra);
+#endif
+}
+
+cv::gpu::GlTexture::~GlTexture()
+{
+}
+
+void cv::gpu::GlTexture::create(int rows_, int cols_, int type_)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    if (rows_ != rows() || cols_ != cols() || type_ != type())
+        impl_ = new Impl(rows_, cols_, type_);
+#endif
+}
+
+void cv::gpu::GlTexture::release()
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_ = new Impl;
+#endif
+}
+
+void cv::gpu::GlTexture::copyFrom(const Mat& mat, bool bgra)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    create(mat.rows, mat.cols, mat.type());
+    impl_->copyFrom(mat, bgra);
+#endif
+}
+
+void cv::gpu::GlTexture::copyFrom(const GlBuffer& buf, bool bgra)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    create(buf.rows(), buf.cols(), buf.type());
+    impl_->copyFrom(buf, bgra);
+#endif
+}
+
+void cv::gpu::GlTexture::bind() const
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_->bind();
+#endif
+}
+
+void cv::gpu::GlTexture::unbind() const
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    impl_->unbind();
+#endif
+}
+
+int cv::gpu::GlTexture::rows() const
+{
+#ifndef HAVE_OPENGL
+    return 0;
+#else
+    return impl_->rows();
+#endif
+}
+
+int cv::gpu::GlTexture::cols() const
+{
+#ifndef HAVE_OPENGL
+    return 0;
+#else
+    return impl_->cols();
+#endif
+}
+
+Size cv::gpu::GlTexture::size() const
+{
+    return Size(cols(), rows());
+}
+
+bool cv::gpu::GlTexture::empty() const
+{
+    return rows() == 0 || cols() == 0;
+}
+
+int cv::gpu::GlTexture::type() const
+{
+#ifndef HAVE_OPENGL
+    return 0;
+#else
+    return impl_->type();
+#endif
+}
+
+int cv::gpu::GlTexture::depth() const 
+{ 
+    return CV_MAT_DEPTH(type()); 
+}
+
+int cv::gpu::GlTexture::channels() const 
+{ 
+    return CV_MAT_CN(type()); 
+}
+
+int cv::gpu::GlTexture::elemSize() const 
+{ 
+    return CV_ELEM_SIZE(type()); 
+}
+
+int cv::gpu::GlTexture::elemSize1() const 
+{ 
+    return CV_ELEM_SIZE1(type()); 
+}
+
+////////////////////////////////////////////////////////////////////////
+// Rendering
+
+void cv::gpu::render(const GlTexture& tex)
+{
+#ifndef HAVE_OPENGL
+    throw_nogl();
+#else
+    if (!tex.empty())
+    {
+        tex.bind();
+
+        glDisable(GL_DEPTH_TEST);
+
+        glMatrixMode(GL_PROJECTION);
+        glLoadIdentity();
+        glOrtho(0, 1, 1, 0, -1, 1);
+
+        glMatrixMode(GL_MODELVIEW);
+        glLoadIdentity();
+
+        glTexEnvi(GL_TEXTURE_ENV, GL_TEXTURE_ENV_MODE, GL_REPLACE);
+
+        glBegin(GL_QUADS);
+            glVertex2d(0.0, 0.0);
+            glTexCoord2d(1.0, 0.0);	 
+
+            glVertex2d(1.0, 0.0);
+            glTexCoord2d(1.0, 1.0);
+
+            glVertex2d(1.0, 1.0);
+            glTexCoord2d(0.0, 1.0);	 
+
+            glVertex2d(0.0, 1.0);
+            glTexCoord2d(0.0, 0.0);	
+        glEnd();
+
+        CV_CheckGlError();
+
+        tex.unbind();
+    }
+#endif
+}
+
+////////////////////////////////////////////////////////////////////////
+// Error handling
+
 void cv::gpu::error(const char *error_string, const char *file, const int line, const char *func)
 {
     int code = CV_GpuApiCallError;
@@ -947,3 +2025,47 @@ void cv::gpu::error(const char *error_string, const char *file, const int line,
     else    
         cv::error( cv::Exception(code, error_string, func, file, line) );
 }
+
+bool cv::gpu::checkGlError(const char* file, const int line, const char* func)
+{
+#ifndef HAVE_OPENGL
+    return true;
+#else
+    GLenum err = glGetError();
+
+    if (err != GL_NO_ERROR)
+    {
+        const char* msg;
+
+        switch (err)
+        {
+        case GL_INVALID_ENUM:
+            msg = "An unacceptable value is specified for an enumerated argument";
+            break;
+        case GL_INVALID_VALUE:
+            msg = "A numeric argument is out of range";
+            break;
+        case GL_INVALID_OPERATION:
+            msg = "The specified operation is not allowed in the current state";
+            break;
+        case GL_STACK_OVERFLOW:
+            msg = "This command would cause a stack overflow";
+            break;
+        case GL_STACK_UNDERFLOW:
+            msg = "This command would cause a stack underflow";
+            break;
+        case GL_OUT_OF_MEMORY:
+            msg = "There is not enough memory left to execute the command";
+            break;
+        default:
+            msg = "Unknown error";
+        };
+
+        cvError(CV_OpenGlApiCallError, func, msg, file, line);
+
+        return false;
+    }
+
+    return true;
+#endif
+}

modules/core/src/system.cpp

@@ -630,7 +630,9 @@ CV_IMPL const char* cvErrorStr( int status )
     case CV_StsBadMemBlock :         return "Memory block has been corrupted";
     case CV_StsAssert :              return "Assertion failed";
     case CV_GpuNotSupported :        return "No GPU support";
-    case CV_GpuApiCallError :        return "Gpu Api call";
+    case CV_GpuApiCallError :        return "Gpu API call";
+    case CV_OpenGlNotSupported :     return "No OpenGL support";
+    case CV_OpenGlApiCallError :     return "OpenGL API call";
     };
 
     sprintf(buf, "Unknown %s code %d", status >= 0 ? "status":"error", status);

modules/gpu/src/cuda/bilateral_filter.cu

@@ -206,7 +206,7 @@ namespace cv { namespace gpu { namespace device
                 }
                 break;
             default:
-                cv::gpu::error("Unsupported channels count", __FILE__, __LINE__);
+                cv::gpu::error("Unsupported channels count", __FILE__, __LINE__, "bilateral_filter_caller");
             }
 
             if (stream != 0)

modules/gpu/src/cuda/hog.cu

@@ -314,7 +314,7 @@ namespace cv { namespace gpu { namespace device
             else if (nthreads == 512)
                 normalize_hists_kernel_many_blocks<512, nblocks><<<grid, threads>>>(block_hist_size, img_block_width, block_hists, threshold);
             else
-                cv::gpu::error("normalize_hists: histogram's size is too big, try to decrease number of bins", __FILE__, __LINE__);
+                cv::gpu::error("normalize_hists: histogram's size is too big, try to decrease number of bins", __FILE__, __LINE__, "normalize_hists");
 
             cudaSafeCall( cudaGetLastError() );
 

modules/gpu/src/cuda/split_merge.cu

@@ -291,7 +291,7 @@ namespace cv { namespace gpu { namespace device
             MergeFunction merge_func = merge_func_tbl[merge_func_id];
 
             if (merge_func == 0)
-                cv::gpu::error("Unsupported channel count or data type", __FILE__, __LINE__);
+                cv::gpu::error("Unsupported channel count or data type", __FILE__, __LINE__, "merge_caller");
 
             merge_func(src, dst, stream);
         }
@@ -498,7 +498,7 @@ namespace cv { namespace gpu { namespace device
             SplitFunction split_func = split_func_tbl[split_func_id];
 
             if (split_func == 0)
-                cv::gpu::error("Unsupported channel count or data type", __FILE__, __LINE__);
+                cv::gpu::error("Unsupported channel count or data type", __FILE__, __LINE__, "split_caller");
 
             split_func(src, dst, stream);
         }

modules/gpu/src/cuda/stereobm.cu

@@ -346,7 +346,7 @@ namespace cv { namespace gpu { namespace device
             int winsz2 = winsz >> 1;
 
             if (winsz2 == 0 || winsz2 >= calles_num)
-                cv::gpu::error("Unsupported window size", __FILE__, __LINE__);
+                cv::gpu::error("Unsupported window size", __FILE__, __LINE__, "stereoBM_GPU");
 
             //cudaSafeCall( cudaFuncSetCacheConfig(&stereoKernel, cudaFuncCachePreferL1) );
             //cudaSafeCall( cudaFuncSetCacheConfig(&stereoKernel, cudaFuncCachePreferShared) );

modules/gpu/src/cuda/stereocsbp.cu

@@ -336,7 +336,7 @@ namespace cv { namespace gpu { namespace device
             case 1: init_data_cost<T, 1><<<grid, threads, 0, stream>>>(h, w, level); break;
             case 3: init_data_cost<T, 3><<<grid, threads, 0, stream>>>(h, w, level); break;
             case 4: init_data_cost<T, 4><<<grid, threads, 0, stream>>>(h, w, level); break;
-            default: cv::gpu::error("Unsupported channels count", __FILE__, __LINE__);
+            default: cv::gpu::error("Unsupported channels count", __FILE__, __LINE__, "init_data_cost_caller_");
             }
         }
 
@@ -355,7 +355,7 @@ namespace cv { namespace gpu { namespace device
             case 1: init_data_cost_reduce<T, winsz, 1><<<grid, threads, smem_size, stream>>>(level, rows, cols, h); break;
             case 3: init_data_cost_reduce<T, winsz, 3><<<grid, threads, smem_size, stream>>>(level, rows, cols, h); break;
             case 4: init_data_cost_reduce<T, winsz, 4><<<grid, threads, smem_size, stream>>>(level, rows, cols, h); break;
-            default: cv::gpu::error("Unsupported channels count", __FILE__, __LINE__);
+            default: cv::gpu::error("Unsupported channels count", __FILE__, __LINE__, "init_data_cost_reduce_caller_");
             }
         }
 
@@ -533,7 +533,7 @@ namespace cv { namespace gpu { namespace device
             case 1: compute_data_cost<T, 1><<<grid, threads, 0, stream>>>(disp_selected_pyr, data_cost, h, w, level, nr_plane); break;
             case 3: compute_data_cost<T, 3><<<grid, threads, 0, stream>>>(disp_selected_pyr, data_cost, h, w, level, nr_plane); break;
             case 4: compute_data_cost<T, 4><<<grid, threads, 0, stream>>>(disp_selected_pyr, data_cost, h, w, level, nr_plane); break;
-            default: cv::gpu::error("Unsupported channels count", __FILE__, __LINE__);
+            default: cv::gpu::error("Unsupported channels count", __FILE__, __LINE__, "compute_data_cost_caller_");
             }
         }
 
@@ -553,7 +553,7 @@ namespace cv { namespace gpu { namespace device
             case 1: compute_data_cost_reduce<T, winsz, 1><<<grid, threads, smem_size, stream>>>(disp_selected_pyr, data_cost, level, rows, cols, h, nr_plane); break;
             case 3: compute_data_cost_reduce<T, winsz, 3><<<grid, threads, smem_size, stream>>>(disp_selected_pyr, data_cost, level, rows, cols, h, nr_plane); break;
             case 4: compute_data_cost_reduce<T, winsz, 4><<<grid, threads, smem_size, stream>>>(disp_selected_pyr, data_cost, level, rows, cols, h, nr_plane); break;
-            default: cv::gpu::error("Unsupported channels count", __FILE__, __LINE__);
+            default: cv::gpu::error("Unsupported channels count", __FILE__, __LINE__, "compute_data_cost_reduce_caller_");
             }
         }
 

modules/gpu/src/opencv2/gpu/device/common.hpp

@@ -66,7 +66,7 @@
 
 namespace cv { namespace gpu 
 {
-    void error(const char *error_string, const char *file, const int line, const char *func = "");
+    void error(const char *error_string, const char *file, const int line, const char *func);
 }}
 
 static inline void ___cudaSafeCall(cudaError_t err, const char *file, const int line, const char *func = "")

modules/gpu/src/opencv2/gpu/device/utility.hpp

@@ -76,7 +76,7 @@ namespace cv { namespace gpu { namespace device
             return mask.ptr(y)[x] != 0;
         }
 
-        const PtrStepb mask;
+        PtrStepb mask;
     };
 
     struct MaskCollection

modules/highgui/include/opencv2/highgui/highgui.hpp

@@ -44,6 +44,7 @@
 #define __OPENCV_HIGHGUI_HPP__
 
 #include "opencv2/core/core.hpp"
+#include "opencv2/core/gpumat.hpp"
 #include "opencv2/highgui/highgui_c.h"
 
 #ifdef __cplusplus
@@ -54,54 +55,34 @@ struct CvVideoWriter;
 namespace cv
 {
 
-enum { WINDOW_AUTOSIZE=1 };
+enum { 
+    // Flags for namedWindow
+    WINDOW_NORMAL   = CV_WINDOW_NORMAL,   // the user can resize the window (no constraint) / also use to switch a fullscreen window to a normal size
+    WINDOW_AUTOSIZE = CV_WINDOW_AUTOSIZE, // the user cannot resize the window, the size is constrainted by the image displayed
+    WINDOW_OPENGL   = CV_WINDOW_OPENGL,   // window with opengl support
 
-CV_EXPORTS_W void namedWindow( const string& winname, int flags=WINDOW_AUTOSIZE );
-CV_EXPORTS_W void destroyWindow( const string& winname );
-CV_EXPORTS_W void destroyAllWindows();
-CV_EXPORTS_W int startWindowThread();
-CV_EXPORTS_W void resizeWindow( const string& name, int width, int height );
-CV_EXPORTS_W void moveWindow( const string& name, int x, int y );
-
-CV_EXPORTS_W void setWindowProperty(const string& winname, int prop_id, double prop_value);//YV
-CV_EXPORTS_W double getWindowProperty(const string& winname, int prop_id);//YV
-
-
-//Only for Qt
-//------------------------
-CV_EXPORTS CvFont fontQt(const string& nameFont, int pointSize=-1,
-                         Scalar color=Scalar::all(0), int weight=CV_FONT_NORMAL,
-                         int style=CV_STYLE_NORMAL, int spacing=0);
-CV_EXPORTS void addText( const Mat& img, const string& text, Point org, CvFont font);
-
-CV_EXPORTS void displayOverlay(const string& winname, const string& text, int delayms);
-CV_EXPORTS void displayStatusBar(const string& winname, const string& text, int delayms);
-
-typedef void (CV_CDECL *OpenGLCallback)(void* userdata);
-CV_EXPORTS void createOpenGLCallback(const string& winname, CvOpenGLCallback callbackOpenGL, void* userdata=0);
+    // Flags for set / getWindowProperty
+    WND_PROP_FULLSCREEN   = CV_WND_PROP_FULLSCREEN,  // fullscreen property
+    WND_PROP_AUTOSIZE     = CV_WND_PROP_AUTOSIZE,    // autosize property
+    WND_PROP_ASPECT_RATIO = CV_WND_PROP_ASPECTRATIO, // window's aspect ration
+    WND_PROP_OPENGL       = CV_WND_PROP_OPENGL       // opengl support
+};
 
-CV_EXPORTS void saveWindowParameters(const string& windowName);
-CV_EXPORTS void loadWindowParameters(const string& windowName);
-CV_EXPORTS  int startLoop(int (*pt2Func)(int argc, char *argv[]), int argc, char* argv[]);
-CV_EXPORTS  void stopLoop();
+CV_EXPORTS_W void namedWindow(const string& winname, int flags = WINDOW_AUTOSIZE);
+CV_EXPORTS_W void destroyWindow(const string& winname);
+CV_EXPORTS_W void destroyAllWindows();
 
-typedef void (CV_CDECL *ButtonCallback)(int state, void* userdata);
-CV_EXPORTS int createButton( const string& bar_name, ButtonCallback on_change,
-                             void* userdata=NULL, int type=CV_PUSH_BUTTON,
-                             bool initial_button_state=0);
-//-------------------------
+CV_EXPORTS_W int startWindowThread();
 
-CV_EXPORTS_W void imshow( const string& winname, InputArray mat );
+CV_EXPORTS_W int waitKey(int delay = 0);
 
-typedef void (CV_CDECL *TrackbarCallback)(int pos, void* userdata);
+CV_EXPORTS_W void imshow(const string& winname, InputArray mat);
 
-CV_EXPORTS int createTrackbar( const string& trackbarname, const string& winname,
-                               int* value, int count,
-                               TrackbarCallback onChange=0,
-                               void* userdata=0);
+CV_EXPORTS_W void resizeWindow(const string& winname, int width, int height);
+CV_EXPORTS_W void moveWindow(const string& winname, int x, int y);
 
-CV_EXPORTS_W int getTrackbarPos( const string& trackbarname, const string& winname );
-CV_EXPORTS_W void setTrackbarPos( const string& trackbarname, const string& winname, int pos );
+CV_EXPORTS_W void setWindowProperty(const string& winname, int prop_id, double prop_value);//YV
+CV_EXPORTS_W double getWindowProperty(const string& winname, int prop_id);//YV
 
 enum
 {
@@ -127,10 +108,65 @@ enum
     EVENT_FLAG_ALTKEY    =32
 };
         
-typedef void (*MouseCallback )(int event, int x, int y, int flags, void* param);
+typedef void (*MouseCallback)(int event, int x, int y, int flags, void* userdata);
 
 //! assigns callback for mouse events
-CV_EXPORTS void setMouseCallback( const string& windowName, MouseCallback onMouse, void* param=0);
+CV_EXPORTS void setMouseCallback(const string& winname, MouseCallback onMouse, void* userdata = 0);
+
+
+typedef void (CV_CDECL *TrackbarCallback)(int pos, void* userdata);
+
+CV_EXPORTS int createTrackbar(const string& trackbarname, const string& winname,
+                              int* value, int count,
+                              TrackbarCallback onChange = 0,
+                              void* userdata = 0);
+
+CV_EXPORTS_W int getTrackbarPos(const string& trackbarname, const string& winname);
+CV_EXPORTS_W void setTrackbarPos(const string& trackbarname, const string& winname, int pos);
+
+// OpenGL support
+
+typedef void (CV_CDECL *OpenGLCallback)(void* userdata);
+CV_EXPORTS void createOpenGLCallback(const string& winname, OpenGLCallback onOpenGlDraw, void* userdata = 0);
+
+typedef void (CV_CDECL *OpenGlDrawCallback)(void* userdata);
+static inline void setOpenGlDrawCallback(const string& winname, OpenGlDrawCallback onOpenGlDraw, void* userdata = 0)
+{
+    createOpenGLCallback(winname, onOpenGlDraw, userdata);
+}
+
+CV_EXPORTS void setOpenGlContext(const string& winname);
+
+CV_EXPORTS void updateWindow(const string& winname);
+
+CV_EXPORTS void imshow(const string& winname, const gpu::GpuMat& d_mat);
+
+CV_EXPORTS void imshow(const string& winname, const gpu::GlBuffer& buf);
+
+CV_EXPORTS void imshow(const string& winname, const gpu::GlTexture& tex);
+
+
+//Only for Qt
+
+CV_EXPORTS CvFont fontQt(const string& nameFont, int pointSize=-1,
+                         Scalar color=Scalar::all(0), int weight=CV_FONT_NORMAL,
+                         int style=CV_STYLE_NORMAL, int spacing=0);
+CV_EXPORTS void addText( const Mat& img, const string& text, Point org, CvFont font);
+
+CV_EXPORTS void displayOverlay(const string& winname, const string& text, int delayms);
+CV_EXPORTS void displayStatusBar(const string& winname, const string& text, int delayms);
+
+CV_EXPORTS void saveWindowParameters(const string& windowName);
+CV_EXPORTS void loadWindowParameters(const string& windowName);
+CV_EXPORTS  int startLoop(int (*pt2Func)(int argc, char *argv[]), int argc, char* argv[]);
+CV_EXPORTS  void stopLoop();
+
+typedef void (CV_CDECL *ButtonCallback)(int state, void* userdata);
+CV_EXPORTS int createButton( const string& bar_name, ButtonCallback on_change,
+                             void* userdata=NULL, int type=CV_PUSH_BUTTON,
+                             bool initial_button_state=0);
+
+//-------------------------
 
 enum
 {
@@ -167,8 +203,6 @@ CV_EXPORTS_W bool imencode( const string& ext, InputArray img,
                             CV_OUT vector<uchar>& buf,
                             const vector<int>& params=vector<int>());
 
-CV_EXPORTS_W int waitKey(int delay=0);
-
 #ifndef CV_NO_VIDEO_CAPTURE_CPP_API
 
 template<> void CV_EXPORTS Ptr<CvCapture>::delete_obj();

modules/highgui/include/opencv2/highgui/highgui_c.h

@@ -76,9 +76,6 @@ CVAPI(void) cvAddText(const CvArr* img, const char* text, CvPoint org, CvFont *a
 CVAPI(void) cvDisplayOverlay(const char* name, const char* text, int delayms);
 CVAPI(void) cvDisplayStatusBar(const char* name, const char* text, int delayms);
 
-typedef void (CV_CDECL *CvOpenGLCallback)(void* userdata);
-CVAPI(void) cvCreateOpenGLCallback( const char* window_name, CvOpenGLCallback callbackOpenGL, void* userdata CV_DEFAULT(NULL), double angle CV_DEFAULT(-1), double zmin CV_DEFAULT(-1), double zmax CV_DEFAULT(-1));
-
 CVAPI(void) cvSaveWindowParameters(const char* name);
 CVAPI(void) cvLoadWindowParameters(const char* name);
 CVAPI(int) cvStartLoop(int (*pt2Func)(int argc, char *argv[]), int argc, char* argv[]);
@@ -99,18 +96,20 @@ CVAPI(int) cvStartWindowThread();
 enum
 {
     //These 3 flags are used by cvSet/GetWindowProperty
-    CV_WND_PROP_FULLSCREEN = 0,//to change/get window's fullscreen property
-    CV_WND_PROP_AUTOSIZE   = 1,//to change/get window's autosize property
-    CV_WND_PROP_ASPECTRATIO= 2,//to change/get window's aspectratio property
-    //
+    CV_WND_PROP_FULLSCREEN = 0, //to change/get window's fullscreen property
+    CV_WND_PROP_AUTOSIZE   = 1, //to change/get window's autosize property
+    CV_WND_PROP_ASPECTRATIO= 2, //to change/get window's aspectratio property
+    CV_WND_PROP_OPENGL     = 3, //to change/get window's opengl support
+    
     //These 2 flags are used by cvNamedWindow and cvSet/GetWindowProperty
-    CV_WINDOW_NORMAL       = 0x00000000,//the user can resize the window (no constraint)  / also use to switch a fullscreen window to a normal size
-    CV_WINDOW_AUTOSIZE     = 0x00000001,//the user cannot resize the window, the size is constrainted by the image displayed
-    //
+    CV_WINDOW_NORMAL       = 0x00000000, //the user can resize the window (no constraint)  / also use to switch a fullscreen window to a normal size
+    CV_WINDOW_AUTOSIZE     = 0x00000001, //the user cannot resize the window, the size is constrainted by the image displayed
+    CV_WINDOW_OPENGL       = 0x00001000, //window with opengl support
+    
     //Those flags are only for Qt
-    CV_GUI_EXPANDED         = 0x00000000,//status bar and tool bar
-    CV_GUI_NORMAL           = 0x00000010,//old fashious way
-    //
+    CV_GUI_EXPANDED         = 0x00000000, //status bar and tool bar
+    CV_GUI_NORMAL           = 0x00000010, //old fashious way
+    
     //These 3 flags are used by cvNamedWindow and cvSet/GetWindowProperty
     CV_WINDOW_FULLSCREEN   = 1,//change the window to fullscreen
     CV_WINDOW_FREERATIO    = 0x00000100,//the image expends as much as it can (no ratio constraint)
@@ -250,6 +249,14 @@ CVAPI(void) cvConvertImage( const CvArr* src, CvArr* dst, int flags CV_DEFAULT(0
 /* wait for key event infinitely (delay<=0) or for "delay" milliseconds */
 CVAPI(int) cvWaitKey(int delay CV_DEFAULT(0));
 
+// OpenGL support
+
+typedef void (CV_CDECL *CvOpenGLCallback)(void* userdata);
+CVAPI(void) cvCreateOpenGLCallback( const char* window_name, CvOpenGLCallback callbackOpenGL, void* userdata CV_DEFAULT(NULL), double angle CV_DEFAULT(-1), double zmin CV_DEFAULT(-1), double zmax CV_DEFAULT(-1));
+
+CVAPI(void) cvSetOpenGlContext(const char* window_name);
+CVAPI(void) cvUpdateWindow(const char* window_name);
+
 /****************************************************************************************\
 *                         Working with Video Files and Cameras                           *
 \****************************************************************************************/

modules/highgui/src/precomp.hpp

@@ -176,14 +176,18 @@ CvCapture * cvCreateCameraCapture_PvAPI  (const int     index);
 CvVideoWriter* cvCreateVideoWriter_GStreamer( const char* filename, int fourcc,
                                             double fps, CvSize frameSize, int is_color );
 
-//Yannick Verdie 2010                                 
+//Yannick Verdie 2010  
+void cvSetModeWindow_W32(const char* name, double prop_value);
+void cvSetModeWindow_GTK(const char* name, double prop_value);
+void cvSetModeWindow_CARBON(const char* name, double prop_value);
+
 double cvGetModeWindow_W32(const char* name);
 double cvGetModeWindow_GTK(const char* name);
 double cvGetModeWindow_CARBON(const char* name);
 
-void cvSetModeWindow_W32(const char* name, double prop_value);
-void cvSetModeWindow_GTK(const char* name, double prop_value);
-void cvSetModeWindow_CARBON(const char* name, double prop_value);
+double cvGetPropWindowAutoSize_W32(const char* name);
+double cvGetRatioWindow_W32(const char* name);
+double cvGetOpenGlProp_W32(const char* name);
 
 //for QT
 #if defined (HAVE_QT)
@@ -195,6 +199,11 @@ double cvGetRatioWindow_QT(const char* name);
 void cvSetRatioWindow_QT(const char* name,double prop_value);
 #endif
 
+// OpenGL
+typedef void (CV_CDECL *CvOpenGlCleanCallback)(void* userdata);
+void icvSetOpenGlCleanCallback(const char* window_name, CvOpenGlCleanCallback callback, void* userdata);
+
+
 /*namespace cv
 {
 

modules/highgui/src/window.cpp

@@ -54,13 +54,13 @@ CV_IMPL void cvSetWindowProperty(const char* name, int prop_id, double prop_valu
             break;
     
         #if defined (HAVE_QT)
-        cvSetModeWindow_QT(name,prop_value);
+            cvSetModeWindow_QT(name,prop_value);
         #elif defined WIN32 || defined _WIN32 
-        cvSetModeWindow_W32(name,prop_value);
+            cvSetModeWindow_W32(name,prop_value);
         #elif defined (HAVE_GTK)
-        cvSetModeWindow_GTK(name,prop_value);
+            cvSetModeWindow_GTK(name,prop_value);
         #elif defined (HAVE_CARBON)
-        cvSetModeWindow_CARBON(name,prop_value);
+            cvSetModeWindow_CARBON(name,prop_value);
         #endif
     break;
     
@@ -83,12 +83,12 @@ CV_IMPL void cvSetWindowProperty(const char* name, int prop_id, double prop_valu
 /* return -1 if error */
 CV_IMPL double cvGetWindowProperty(const char* name, int prop_id)
 {
+    if (!name) 
+        return -1;
+
     switch(prop_id)
     {
-    case CV_WND_PROP_FULLSCREEN:
-    
-        if (!name)//bad argument
-            return -1;
+    case CV_WND_PROP_FULLSCREEN:    
         
         #if defined (HAVE_QT)
             return cvGetModeWindow_QT(name);
@@ -101,25 +101,38 @@ CV_IMPL double cvGetWindowProperty(const char* name, int prop_id)
         #else
             return -1;
         #endif
+    break;
     
     case CV_WND_PROP_AUTOSIZE:
-    
-        if (!name)//bad argument
-            return -1;
-            
+                
         #if defined (HAVE_QT)
             return cvGetPropWindow_QT(name);
+        #elif defined WIN32 || defined _WIN32 
+            return cvGetPropWindowAutoSize_W32(name);
         #else
             return -1;
         #endif  
+    break;
         
     case CV_WND_PROP_ASPECTRATIO:
+
         #if defined (HAVE_QT)
             return cvGetRatioWindow_QT(name);
+        #elif defined WIN32 || defined _WIN32 
+            return cvGetRatioWindow_W32(name);
         #else
             return -1;
         #endif  
     break;
+
+    case CV_WND_PROP_OPENGL:
+
+        #if defined WIN32 || defined _WIN32 
+            return cvGetOpenGlProp_W32(name);
+        #else
+            return -1;
+        #endif 
+    break;
     
     default:
         return -1;
@@ -153,19 +166,12 @@ void cv::moveWindow( const string& winname, int x, int y )
 
 void cv::setWindowProperty(const string& winname, int prop_id, double prop_value)
 {
-    cvSetWindowProperty( winname.c_str(),prop_id,prop_value);
+    cvSetWindowProperty( winname.c_str(), prop_id, prop_value);
 }
 
 double cv::getWindowProperty(const string& winname, int prop_id)
 {
-    return  cvGetWindowProperty(winname.c_str(),prop_id);
-}
-
-void cv::imshow( const string& winname, InputArray _img )
-{
-    Mat img = _img.getMat();
-    CvMat c_img = img;
-    cvShowImage( winname.c_str(), &c_img );
+    return cvGetWindowProperty(winname.c_str(), prop_id);
 }
 
 int cv::waitKey(int delay)
@@ -201,6 +207,284 @@ int cv::startWindowThread()
     return cvStartWindowThread();
 }
 
+// OpenGL support
+
+void cv::createOpenGLCallback(const string& name, OpenGLCallback callback, void* userdata)
+{
+    cvCreateOpenGLCallback(name.c_str(), callback, userdata);
+}
+
+void cv::setOpenGlContext(const string& windowName)
+{
+    cvSetOpenGlContext(windowName.c_str());
+}
+
+void cv::updateWindow(const string& windowName)
+{
+    cvUpdateWindow(windowName.c_str());
+}
+
+#ifdef HAVE_OPENGL
+namespace
+{
+    const int CV_TEXTURE_MAGIC_VAL        = 0x00287653;
+
+    struct GlObjBase
+    {
+        int flag;
+        GlObjBase* next;
+        GlObjBase* prev;
+        std::string winname;
+
+        virtual ~GlObjBase() {}
+    };
+
+    GlObjBase* g_glObjs = 0;
+
+    GlObjBase* findGlObjByName(const std::string& winname)
+    {
+        GlObjBase* obj = g_glObjs;
+
+        while(obj && obj->winname != winname)
+            obj = obj->next;
+
+        return obj;
+    }
+
+    void addGlObj(GlObjBase* glObj)
+    {
+        glObj->next = g_glObjs;
+        glObj->prev = 0;
+        if (g_glObjs)
+            g_glObjs->prev = glObj;
+        g_glObjs = glObj;
+    }
+
+    void removeGlObj(GlObjBase* glObj)
+    {
+        if (glObj->prev)
+            glObj->prev->next = glObj->next;
+        else
+            g_glObjs = glObj->next;
+
+        if (glObj->next)
+            glObj->next->prev = glObj->prev;
+
+        delete glObj;
+    }
+
+    template <typename T>
+    struct GlObj : GlObjBase
+    {
+        T obj;
+    };
+
+    void CV_CDECL glDrawTextureCallback(void* userdata)
+    {
+        GlObj<cv::gpu::GlTexture>* texObj = static_cast<GlObj<cv::gpu::GlTexture>*>(userdata);
+
+        CV_DbgAssert(texObj->flag == CV_TEXTURE_MAGIC_VAL);
+
+        cv::gpu::render(texObj->obj);
+    }
+
+    void CV_CDECL glCleanCallback(void* userdata)
+    {
+        GlObjBase* glObj = static_cast<GlObjBase*>(userdata);
+
+        removeGlObj(glObj);
+    }
+}
+#endif // HAVE_OPENGL
+
+void cv::imshow( const string& winname, InputArray _img )
+{
+    Mat img = _img.getMat();
+
+#ifndef HAVE_OPENGL
+    CvMat c_img = img;
+    cvShowImage(winname.c_str(), &c_img);
+#else
+    double useGl = getWindowProperty(winname, WND_PROP_OPENGL);
+    if (useGl <= 0)
+    {
+        CvMat c_img = img;
+        cvShowImage(winname.c_str(), &c_img);
+    }
+    else
+    {
+        double autoSize = getWindowProperty(winname, WND_PROP_AUTOSIZE);
+
+        if (autoSize > 0)
+            resizeWindow(winname, img.cols, img.rows);
+
+        setOpenGlContext(winname);
+
+        GlObjBase* glObj = findGlObjByName(winname);
+
+        if (glObj && glObj->flag != CV_TEXTURE_MAGIC_VAL)
+        {
+            icvSetOpenGlCleanCallback(winname.c_str(), 0, 0);
+            glObj = 0;
+        }
+
+        if (glObj)
+        {
+            GlObj<cv::gpu::GlTexture>* texObj = static_cast<GlObj<cv::gpu::GlTexture>*>(glObj);
+            texObj->obj.copyFrom(img);
+        }
+        else
+        {
+            GlObj<cv::gpu::GlTexture>* texObj = new GlObj<cv::gpu::GlTexture>;
+            texObj->obj.copyFrom(img);
+
+            glObj = texObj;
+            glObj->flag = CV_TEXTURE_MAGIC_VAL;
+            glObj->winname = winname;
+
+            addGlObj(glObj);
+
+            icvSetOpenGlCleanCallback(winname.c_str(), glCleanCallback, glObj);
+        }        
+
+        setOpenGlDrawCallback(winname, glDrawTextureCallback, glObj);
+
+        updateWindow(winname);
+    }
+#endif
+}
+
+void cv::imshow(const string& winname, const gpu::GlBuffer& buf)
+{
+#ifndef HAVE_OPENGL
+    CV_Error(CV_OpenGlNotSupported, "The library is compiled without OpenGL support"); 
+#else
+    CV_Assert(buf.usage() == gpu::GlBuffer::TEXTURE_BUFFER);
+
+    namedWindow(winname, WINDOW_OPENGL | WINDOW_AUTOSIZE);
+
+    double autoSize = getWindowProperty(winname, WND_PROP_AUTOSIZE);
+
+    if (autoSize > 0)
+        resizeWindow(winname, buf.cols(), buf.rows());
+
+    setOpenGlContext(winname);
+
+    GlObjBase* glObj = findGlObjByName(winname);
+
+    if (glObj && glObj->flag != CV_TEXTURE_MAGIC_VAL)
+    {
+        icvSetOpenGlCleanCallback(winname.c_str(), 0, 0);
+        glObj = 0;
+    }
+
+    if (glObj)
+    {
+        GlObj<cv::gpu::GlTexture>* texObj = static_cast<GlObj<cv::gpu::GlTexture>*>(glObj);
+        texObj->obj.copyFrom(buf);
+    }
+    else
+    {
+        GlObj<cv::gpu::GlTexture>* texObj = new GlObj<cv::gpu::GlTexture>;
+        texObj->obj.copyFrom(buf);
+
+        glObj = texObj;
+        glObj->flag = CV_TEXTURE_MAGIC_VAL;
+        glObj->winname = winname;
+
+        addGlObj(glObj);
+
+        icvSetOpenGlCleanCallback(winname.c_str(), glCleanCallback, glObj);
+    }        
+
+    setOpenGlDrawCallback(winname, glDrawTextureCallback, glObj);
+
+    updateWindow(winname);
+#endif
+}
+
+void cv::imshow(const string& winname, const gpu::GpuMat& d_mat)
+{
+#ifndef HAVE_OPENGL
+    CV_Error(CV_OpenGlNotSupported, "The library is compiled without OpenGL support"); 
+#else
+    setOpenGlContext(winname);
+    gpu::GlBuffer buf(d_mat, gpu::GlBuffer::TEXTURE_BUFFER);
+    imshow(winname, buf);
+#endif
+}
+
+void cv::imshow(const string& winname, const gpu::GlTexture& tex)
+{
+#ifndef HAVE_OPENGL
+    CV_Error(CV_OpenGlNotSupported, "The library is compiled without OpenGL support"); 
+#else
+    namedWindow(winname, WINDOW_OPENGL | WINDOW_AUTOSIZE);
+
+    double autoSize = getWindowProperty(winname, WND_PROP_AUTOSIZE);
+
+    if (autoSize > 0)
+        resizeWindow(winname, tex.cols(), tex.rows());
+
+    setOpenGlContext(winname);
+
+    GlObjBase* glObj = findGlObjByName(winname);
+
+    if (glObj && glObj->flag != CV_TEXTURE_MAGIC_VAL)
+    {
+        icvSetOpenGlCleanCallback(winname.c_str(), 0, 0);
+        glObj = 0;
+    }
+
+    if (glObj)
+    {
+        GlObj<cv::gpu::GlTexture>* texObj = static_cast<GlObj<cv::gpu::GlTexture>*>(glObj);
+        texObj->obj = tex;
+    }
+    else
+    {
+        GlObj<cv::gpu::GlTexture>* texObj = new GlObj<cv::gpu::GlTexture>;
+        texObj->obj = tex;
+
+        glObj = texObj;
+        glObj->flag = CV_TEXTURE_MAGIC_VAL;
+        glObj->winname = winname;
+
+        addGlObj(glObj);
+
+        icvSetOpenGlCleanCallback(winname.c_str(), glCleanCallback, glObj);
+    }
+
+    setOpenGlDrawCallback(winname, glDrawTextureCallback, glObj);
+
+    updateWindow(winname);
+#endif
+}
+
+#ifndef HAVE_OPENGL
+
+CV_IMPL void cvCreateOpenGLCallback(const char*, CvOpenGLCallback, void*, double, double, double)
+{
+    CV_Error(CV_OpenGlNotSupported, "The library is compiled without OpenGL support"); 
+}
+
+CV_IMPL void cvSetOpenGlContext(const char*)
+{
+    CV_Error(CV_OpenGlNotSupported, "The library is compiled without OpenGL support"); 
+}
+
+CV_IMPL void cvUpdateWindow(const char*)
+{
+    CV_Error(CV_OpenGlNotSupported, "The library is compiled without OpenGL support"); 
+}
+
+void icvSetOpenGlCleanCallback(const char*, CvOpenGlCleanCallback, void*)
+{
+    CV_Error(CV_OpenGlNotSupported, "The library is compiled without OpenGL support"); 
+}
+
+#endif // !HAVE_OPENGL
+
 #if defined (HAVE_QT)
 
 CvFont cv::fontQt(const string& nameFont, int pointSize, Scalar color, int weight,  int style, int spacing)
@@ -219,11 +503,6 @@ void cv::displayStatusBar(const string& name,  const string& text, int delayms)
     cvDisplayStatusBar(name.c_str(),text.c_str(), delayms);
 }
 
-void cv::createOpenGLCallback(const string& name,  OpenGLCallback callback, void* param)
-{
-    cvCreateOpenGLCallback(name.c_str(),callback, param);
-}
-
 void cv::displayOverlay(const string& name,  const string& text, int delayms)
 {
     cvDisplayOverlay(name.c_str(),text.c_str(), delayms);
@@ -379,7 +658,6 @@ CV_IMPL int cvStartWindowThread()
     return -1;
 }
 
-
 //-------- Qt ---------
 CV_IMPL void cvAddText( const CvArr*, const char*, CvPoint org, CvFont* font)
 {

modules/highgui/src/window_w32.cpp

@@ -54,6 +54,13 @@
 #include <stdio.h>
 #include <assert.h>
 
+#ifdef HAVE_OPENGL
+#include <memory>
+#include "opencv2/highgui/highgui.hpp"
+#include "opencv2/core/gpumat.hpp"
+#include <GL\gl.h>
+#endif
+
 static const char* trackbar_text =
 "                                                                                             ";
 
@@ -81,7 +88,7 @@ static const char* trackbar_text =
 
 #endif
 
-void  FillBitmapInfo( BITMAPINFO* bmi, int width, int height, int bpp, int origin )
+void FillBitmapInfo( BITMAPINFO* bmi, int width, int height, int bpp, int origin )
 {
     assert( bmi && width >= 0 && height >= 0 && (bpp == 8 || bpp == 24 || bpp == 32));
 
@@ -127,7 +134,6 @@ typedef struct CvTrackbar
 }
 CvTrackbar;
 
-
 typedef struct CvWindow
 {
     int signature;
@@ -155,6 +161,24 @@ typedef struct CvWindow
         CvTrackbar* first;
     }
     toolbar;
+
+    int width;
+    int height;
+
+    // OpenGL support
+
+#ifdef HAVE_OPENGL
+    bool useGl;
+    HGLRC hGLRC;
+
+    CvOpenGLCallback glDrawCallback;
+    void* glDrawData;
+
+    CvOpenGlCleanCallback glCleanCallback;
+    void* glCleanData;
+
+    cv::gpu::GlFuncTab* glFuncTab;
+#endif
 }
 CvWindow;
 
@@ -464,6 +488,420 @@ void cvSetModeWindow_W32( const char*, double)
 }
 #endif
 
+double cvGetPropWindowAutoSize_W32(const char* name)
+{
+    double result = -1;
+
+    CV_FUNCNAME( "cvSetCloseCallback" );
+
+    __BEGIN__;
+
+    CvWindow* window;
+
+    if (!name)
+        CV_ERROR( CV_StsNullPtr, "NULL name string" );
+
+    window = icvFindWindowByName( name );
+    if (!window)
+        EXIT;
+
+    result = window->flags & CV_WINDOW_AUTOSIZE;
+
+    __END__;
+
+    return result;
+}
+
+double cvGetRatioWindow_W32(const char* name)
+{
+	double result = -1;
+	
+	CV_FUNCNAME( "cvGetRatioWindow_W32" );
+
+    __BEGIN__;
+
+    CvWindow* window;
+
+    if (!name)
+        CV_ERROR( CV_StsNullPtr, "NULL name string" );
+
+    window = icvFindWindowByName( name );
+    if (!window)
+        CV_ERROR( CV_StsNullPtr, "NULL window" );
+        
+    result = static_cast<double>(window->width) / window->height;
+        
+    __END__;
+
+    return result;   
+}
+
+double cvGetOpenGlProp_W32(const char* name)
+{
+	double result = -1;
+
+#ifdef HAVE_OPENGL	
+	CV_FUNCNAME( "cvGetOpenGlProp_W32" );
+
+    __BEGIN__;
+
+    CvWindow* window;
+
+    if (!name)
+        CV_ERROR( CV_StsNullPtr, "NULL name string" );
+
+    window = icvFindWindowByName( name );
+    if (!window)
+        __CV_EXIT__;
+        
+    result = window->useGl;
+        
+    __END__;
+#endif
+
+    return result;   
+}
+
+
+// OpenGL support
+
+#ifdef HAVE_OPENGL
+
+#ifndef APIENTRY
+    #define APIENTRY
+#endif
+
+#ifndef APIENTRYP
+    #define APIENTRYP APIENTRY *
+#endif
+
+#ifndef GL_VERSION_1_5
+    /* GL types for handling large vertex buffer objects */
+    typedef ptrdiff_t GLintptr;
+    typedef ptrdiff_t GLsizeiptr;
+#endif
+
+#ifndef GL_BGR
+    #define GL_BGR 0x80E0
+#endif
+
+#ifndef GL_BGRA
+    #define GL_BGRA 0x80E1
+#endif
+
+namespace
+{
+    typedef void (APIENTRYP PFNGLGENBUFFERSPROC   ) (GLsizei n, GLuint *buffers);
+    typedef void (APIENTRYP PFNGLDELETEBUFFERSPROC) (GLsizei n, const GLuint *buffers);
+
+    typedef void (APIENTRYP PFNGLBUFFERDATAPROC   ) (GLenum target, GLsizeiptr size, const GLvoid *data, GLenum usage);
+    typedef void (APIENTRYP PFNGLBUFFERSUBDATAPROC) (GLenum target, GLintptr offset, GLsizeiptr size, const GLvoid* data);
+
+    typedef void (APIENTRYP PFNGLBINDBUFFERPROC   ) (GLenum target, GLuint buffer);
+
+    typedef GLvoid* (APIENTRYP PFNGLMAPBUFFERPROC) (GLenum target, GLenum access);
+    typedef GLboolean (APIENTRYP PFNGLUNMAPBUFFERPROC) (GLenum target);
+
+    class GlFuncTab_W32 : public cv::gpu::GlFuncTab
+    {
+    public:
+        PFNGLGENBUFFERSPROC    glGenBuffersExt;
+        PFNGLDELETEBUFFERSPROC glDeleteBuffersExt;
+
+        PFNGLBUFFERDATAPROC    glBufferDataExt;
+        PFNGLBUFFERSUBDATAPROC glBufferSubDataExt;
+
+        PFNGLBINDBUFFERPROC    glBindBufferExt;
+
+        PFNGLMAPBUFFERPROC     glMapBufferExt;
+        PFNGLUNMAPBUFFERPROC   glUnmapBufferExt;
+
+        bool initialized;
+
+        GlFuncTab_W32()
+        {
+            glGenBuffersExt    = 0;
+            glDeleteBuffersExt = 0;
+
+            glBufferDataExt    = 0;
+            glBufferSubDataExt = 0;
+
+            glBindBufferExt    = 0;
+
+            glMapBufferExt     = 0;
+            glUnmapBufferExt   = 0;
+
+            initialized = false;
+        }
+
+        void genBuffers(int n, unsigned int* buffers) const
+        {
+            CV_FUNCNAME( "genBuffers" );
+
+            __BEGIN__;
+
+            if (!glGenBuffersExt)
+                CV_ERROR(CV_OpenGlApiCallError, "Current OpenGL implementation doesn't support required extension");
+
+            glGenBuffersExt(n, buffers);
+            CV_CheckGlError();
+
+            __END__;
+        }
+
+        void deleteBuffers(int n, const unsigned int* buffers) const
+        {
+            CV_FUNCNAME( "deleteBuffers" );
+
+            __BEGIN__;
+
+            if (!glDeleteBuffersExt)
+                CV_ERROR(CV_OpenGlApiCallError, "Current OpenGL implementation doesn't support required extension");
+
+            glDeleteBuffersExt(n, buffers);
+            CV_CheckGlError();
+
+            __END__;
+        }
+
+        void bufferData(unsigned int target, ptrdiff_t size, const void* data, unsigned int usage) const
+        {
+            CV_FUNCNAME( "bufferData" );
+
+            __BEGIN__;
+
+            if (!glBufferDataExt)
+                CV_ERROR(CV_OpenGlApiCallError, "Current OpenGL implementation doesn't support required extension");
+
+            glBufferDataExt(target, size, data, usage);
+            CV_CheckGlError();
+
+            __END__;
+        }
+
+        void bufferSubData(unsigned int target, ptrdiff_t offset, ptrdiff_t size, const void* data) const
+        {
+            CV_FUNCNAME( "bufferSubData" );
+
+            __BEGIN__;
+
+            if (!glBufferSubDataExt)
+                CV_ERROR(CV_OpenGlApiCallError, "Current OpenGL implementation doesn't support required extension");
+
+            glBufferSubDataExt(target, offset, size, data);
+            CV_CheckGlError();
+
+            __END__;
+        }
+
+        void bindBuffer(unsigned int target, unsigned int buffer) const
+        {
+            CV_FUNCNAME( "bindBuffer" );
+
+            __BEGIN__;
+
+            if (!glBindBufferExt)
+                CV_ERROR(CV_OpenGlApiCallError, "Current OpenGL implementation doesn't support required extension");
+
+            glBindBufferExt(target, buffer);
+            CV_CheckGlError();
+
+            __END__;
+        }
+
+        void* mapBuffer(unsigned int target, unsigned int access) const
+        {
+            CV_FUNCNAME( "mapBuffer" );
+
+            void* res = 0;
+
+            __BEGIN__;
+
+            if (!glMapBufferExt)
+                CV_ERROR(CV_OpenGlApiCallError, "Current OpenGL implementation doesn't support required extension");
+
+            res = glMapBufferExt(target, access);
+            CV_CheckGlError();
+
+            __END__;
+
+            return res;
+        }
+
+        void unmapBuffer(unsigned int target) const
+        {
+            CV_FUNCNAME( "unmapBuffer" );
+
+            __BEGIN__;
+
+            if (!glUnmapBufferExt)
+                CV_ERROR(CV_OpenGlApiCallError, "Current OpenGL implementation doesn't support required extension");
+
+            glUnmapBufferExt(target);
+            CV_CheckGlError();
+
+            __END__;
+        }
+
+        bool isGlContextInitialized() const
+        {
+            return initialized;
+        }
+    };
+
+    void initGl(CvWindow* window)
+    {
+        std::auto_ptr<GlFuncTab_W32> glFuncTab(new GlFuncTab_W32);
+
+        // Load extensions
+        PROC func;
+
+        func = wglGetProcAddress("glGenBuffers");
+        glFuncTab->glGenBuffersExt = (PFNGLGENBUFFERSPROC)func;
+
+        func = wglGetProcAddress("glDeleteBuffers");
+        glFuncTab->glDeleteBuffersExt = (PFNGLDELETEBUFFERSPROC)func;
+
+        func = wglGetProcAddress("glBufferData");
+        glFuncTab->glBufferDataExt = (PFNGLBUFFERDATAPROC)func;
+
+        func = wglGetProcAddress("glBufferSubData");
+        glFuncTab->glBufferSubDataExt = (PFNGLBUFFERSUBDATAPROC)func;
+
+        func = wglGetProcAddress("glBindBuffer");
+        glFuncTab->glBindBufferExt = (PFNGLBINDBUFFERPROC)func;
+
+        func = wglGetProcAddress("glMapBuffer");
+        glFuncTab->glMapBufferExt = (PFNGLMAPBUFFERPROC)func;
+
+        func = wglGetProcAddress("glUnmapBuffer");
+        glFuncTab->glUnmapBufferExt = (PFNGLUNMAPBUFFERPROC)func;
+
+        glFuncTab->initialized = true;
+
+        window->glFuncTab = glFuncTab.release();
+
+        cv::gpu::setGlFuncTab(window->glFuncTab);
+    }
+
+    void createGlContext(HWND hWnd, HDC& hGLDC, HGLRC& hGLRC, bool& useGl)
+    {
+        CV_FUNCNAME( "createGlContext" );
+
+        __BEGIN__;
+
+        useGl = false;
+
+        int PixelFormat;
+
+        static PIXELFORMATDESCRIPTOR pfd =
+        {
+            sizeof(PIXELFORMATDESCRIPTOR), // Size Of This Pixel Format Descriptor
+            1,                             // Version Number
+            PFD_DRAW_TO_WINDOW |           // Format Must Support Window
+            PFD_SUPPORT_OPENGL |           // Format Must Support OpenGL
+            PFD_DOUBLEBUFFER,              // Must Support Double Buffering
+            PFD_TYPE_RGBA,                 // Request An RGBA Format
+            32,                            // Select Our Color Depth
+            0, 0, 0, 0, 0, 0,              // Color Bits Ignored
+            0,                             // No Alpha Buffer
+            0,                             // Shift Bit Ignored
+            0,                             // No Accumulation Buffer
+            0, 0, 0, 0,                    // Accumulation Bits Ignored
+            16,                            // 16Bit Z-Buffer (Depth Buffer)  
+            0,                             // No Stencil Buffer
+            0,                             // No Auxiliary Buffer
+            PFD_MAIN_PLANE,                // Main Drawing Layer
+            0,                             // Reserved
+            0, 0, 0	                       // Layer Masks Ignored
+        };
+
+        hGLDC = GetDC(hWnd);
+        if (!hGLDC)
+            CV_ERROR( CV_OpenGlApiCallError, "Can't Create A GL Device Context" );
+
+        PixelFormat = ChoosePixelFormat(hGLDC, &pfd);
+        if (!PixelFormat)
+            CV_ERROR( CV_OpenGlApiCallError, "Can't Find A Suitable PixelFormat" );
+
+        if (!SetPixelFormat(hGLDC, PixelFormat, &pfd))
+            CV_ERROR( CV_OpenGlApiCallError, "Can't Set The PixelFormat" );
+
+        hGLRC = wglCreateContext(hGLDC);
+        if (!hGLRC)
+            CV_ERROR( CV_OpenGlApiCallError, "Can't Create A GL Rendering Context" );
+
+        if (!wglMakeCurrent(hGLDC, hGLRC))
+            CV_ERROR( CV_OpenGlApiCallError, "Can't Activate The GL Rendering Context" );
+
+        useGl = true;
+
+        __END__;
+    }
+
+    void releaseGlContext(CvWindow* window)
+    {
+        CV_FUNCNAME( "releaseGlContext" );
+
+        __BEGIN__;
+
+        if (window->hGLRC)
+        {
+            wglDeleteContext(window->hGLRC);
+            window->hGLRC = NULL;
+        }
+
+        if (window->dc)	
+        {
+            ReleaseDC(window->hwnd, window->dc);
+            window->dc = NULL;
+        }
+
+        window->useGl = false;
+
+        __END__;
+    }
+
+    void drawGl(CvWindow* window)
+    {
+        CV_FUNCNAME( "drawGl" );
+
+        __BEGIN__;
+
+        if (!wglMakeCurrent(window->dc, window->hGLRC))
+            CV_ERROR( CV_OpenGlApiCallError, "Can't Activate The GL Rendering Context" );
+
+	    glClear(GL_COLOR_BUFFER_BIT | GL_DEPTH_BUFFER_BIT);
+
+        if (window->glDrawCallback)
+            window->glDrawCallback(window->glDrawData);
+
+        CV_CheckGlError();
+
+        if (!SwapBuffers(window->dc))
+            CV_ERROR( CV_OpenGlApiCallError, "Can't swap OpenGL buffers" );
+
+        __END__;
+    }
+
+    void resizeGl(CvWindow* window)
+    {
+        CV_FUNCNAME( "resizeGl" );
+
+        __BEGIN__;
+
+        if (!wglMakeCurrent(window->dc, window->hGLRC))
+            CV_ERROR( CV_OpenGlApiCallError, "Can't Activate The GL Rendering Context" );
+
+        glViewport(0, 0, window->width, window->height);
+
+        __END__;
+    }
+}
+
+#endif // HAVE_OPENGL
+
+
 CV_IMPL int cvNamedWindow( const char* name, int flags )
 {
     int result = 0;
@@ -483,9 +921,61 @@ CV_IMPL int cvNamedWindow( const char* name, int flags )
         CV_ERROR( CV_StsNullPtr, "NULL name string" );
 
     // Check the name in the storage
-    if( icvFindWindowByName( name ) != 0 )
+    window = icvFindWindowByName( name );
+    if (window != 0)
     {
         result = 1;
+
+        #ifdef HAVE_OPENGL
+            if (window->useGl && !(flags & CV_WINDOW_OPENGL))
+            {
+                wglMakeCurrent(window->dc, window->hGLRC);
+
+                if (window->glCleanCallback)
+                {
+                    window->glCleanCallback(window->glCleanData);
+                    window->glCleanCallback = 0;
+                    window->glCleanData = 0;
+                }
+
+                releaseGlContext(window);
+
+                window->dc = CreateCompatibleDC(0);
+                window->hGLRC = 0;
+                window->useGl = false;
+            }
+            else if (!window->useGl && (flags & CV_WINDOW_OPENGL))
+            {
+                if (window->dc && window->image)
+                    DeleteObject(SelectObject(window->dc, window->image));
+
+                if (window->dc)
+                    DeleteDC(window->dc);
+
+                bool useGl = false;
+                HDC hGLDC = 0;
+                HGLRC hGLRC = 0;
+
+                createGlContext(window->hwnd, hGLDC, hGLRC, useGl);
+
+                if (!useGl)
+                {
+                    window->dc = CreateCompatibleDC(0);
+                    window->hGLRC = 0;
+                    window->useGl = false;
+
+                    result = 0;
+                }
+                else
+                {
+                    window->dc = hGLDC;
+                    window->hGLRC = hGLRC;
+                    window->useGl = true;
+                    initGl(window);
+                }
+            }
+        #endif // HAVE_OPENGL
+
         EXIT;
     }
 
@@ -506,6 +996,18 @@ CV_IMPL int cvNamedWindow( const char* name, int flags )
     if( !hWnd )
         CV_ERROR( CV_StsError, "Frame window can not be created" );
 
+#ifndef HAVE_OPENGL
+    if (flags & CV_WINDOW_OPENGL)
+        CV_ERROR( CV_OpenGlNotSupported, "Library was built without OpenGL support" );
+#else
+    bool useGl = false;
+    HDC hGLDC = 0;
+    HGLRC hGLRC = 0;
+
+    if (flags & CV_WINDOW_OPENGL)
+        createGlContext(hWnd, hGLDC, hGLRC, useGl);
+#endif
+
     ShowWindow(hWnd, SW_SHOW);
 
     len = (int)strlen(name);
@@ -518,7 +1020,32 @@ CV_IMPL int cvNamedWindow( const char* name, int flags )
     memcpy( window->name, name, len + 1 );
     window->flags = flags;
     window->image = 0;
+
+#ifndef HAVE_OPENGL
     window->dc = CreateCompatibleDC(0);
+#else
+    window->glFuncTab = 0;
+    if (!useGl)
+    {
+        window->dc = CreateCompatibleDC(0);
+        window->hGLRC = 0;
+        window->useGl = false;
+    }
+    else
+    {
+        window->dc = hGLDC;
+        window->hGLRC = hGLRC;
+        window->useGl = true;
+        initGl(window);
+    }
+
+    window->glDrawCallback = 0;
+    window->glDrawData = 0;
+
+    window->glCleanCallback = 0;
+    window->glCleanData = 0;
+#endif
+
     window->last_key = 0;
     window->status = CV_WINDOW_NORMAL;//YV
 
@@ -544,12 +1071,128 @@ CV_IMPL int cvNamedWindow( const char* name, int flags )
     return result;
 }
 
+#ifdef HAVE_OPENGL
+
+CV_IMPL void cvSetOpenGlContext(const char* name)
+{
+    CV_FUNCNAME( "cvSetOpenGlContext" );
+
+    __BEGIN__;
+
+    CvWindow* window;
+
+    if(!name)
+        CV_ERROR( CV_StsNullPtr, "NULL name string" );
+
+    window = icvFindWindowByName( name );
+    if (!window)
+        CV_ERROR( CV_StsNullPtr, "NULL window" );
+
+    if (!window->useGl)
+        CV_ERROR( CV_OpenGlNotSupported, "Window doesn't support OpenGL" );
+
+    if (!wglMakeCurrent(window->dc, window->hGLRC))
+        CV_ERROR( CV_OpenGlApiCallError, "Can't Activate The GL Rendering Context" );
+
+    cv::gpu::setGlFuncTab(window->glFuncTab);
+
+    __END__;
+}
+
+CV_IMPL void cvUpdateWindow(const char* name)
+{
+    CV_FUNCNAME( "cvUpdateWindow" );
+
+    __BEGIN__;
+
+    CvWindow* window;
+
+    if (!name)
+        CV_ERROR( CV_StsNullPtr, "NULL name string" );
+
+    window = icvFindWindowByName( name );
+    if (!window)
+        EXIT;
+
+    InvalidateRect(window->hwnd, 0, 0);
+
+    __END__;
+}
+
+CV_IMPL void cvCreateOpenGLCallback(const char* name, CvOpenGLCallback callback, void* userdata, double, double, double)
+{
+    CV_FUNCNAME( "cvCreateOpenGLCallback" );
+
+    __BEGIN__;
+
+    CvWindow* window;
+
+    if(!name)
+        CV_ERROR( CV_StsNullPtr, "NULL name string" );
+
+    window = icvFindWindowByName( name );
+    if( !window )
+        EXIT;
+
+    if (!window->useGl)
+        CV_ERROR( CV_OpenGlNotSupported, "Window was created without OpenGL context" );
+
+    window->glDrawCallback = callback;
+    window->glDrawData = userdata;
+
+    __END__;
+}
+
+void icvSetOpenGlCleanCallback(const char* name, CvOpenGlCleanCallback callback, void* userdata)
+{
+    CV_FUNCNAME( "icvSetOpenGlCleanCallback" );
+
+    __BEGIN__;
+
+    CvWindow* window;
+
+    if (!name)
+        CV_ERROR(CV_StsNullPtr, "NULL name string");
+
+    window = icvFindWindowByName(name);
+    if (!window)
+        EXIT;
+
+    if (window->glCleanCallback)
+        window->glCleanCallback(window->glCleanData);
+
+    window->glCleanCallback = callback;
+    window->glCleanData = userdata;
+
+    __END__;
+}
+
+#endif // HAVE_OPENGL
 
 static void icvRemoveWindow( CvWindow* window )
 {
     CvTrackbar* trackbar = NULL;
     RECT wrect={0,0,0,0};
 
+#ifdef HAVE_OPENGL
+    if (window->useGl)
+    {
+        delete window->glFuncTab;
+         cv::gpu::setGlFuncTab(0);
+
+        wglMakeCurrent(window->dc, window->hGLRC);
+
+        if (window->glCleanCallback)
+        {
+            window->glCleanCallback(window->glCleanData);
+            window->glCleanCallback = 0;
+            window->glCleanData = 0;
+        }
+
+        releaseGlContext(window);
+    }
+#endif
+
     if( window->frame )
         GetWindowRect( window->frame, &wrect );
     if( window->name )
@@ -559,7 +1202,7 @@ static void icvRemoveWindow( CvWindow* window )
     if( window->hwnd )
         icvSetWindowLongPtr( window->hwnd, CV_USERDATA, 0 );
     if( window->frame )
-    icvSetWindowLongPtr( window->frame, CV_USERDATA, 0 );
+        icvSetWindowLongPtr( window->frame, CV_USERDATA, 0 );
 
     if( window->toolbar.toolbar )
         icvSetWindowLongPtr(window->toolbar.toolbar, CV_USERDATA, 0);
@@ -722,7 +1365,6 @@ static void icvUpdateWindowPos( CvWindow* window )
                rect.bottom - rect.top + 1, TRUE );
 }
 
-
 CV_IMPL void
 cvShowImage( const char* name, const CvArr* arr )
 {
@@ -745,7 +1387,12 @@ cvShowImage( const char* name, const CvArr* arr )
     window = icvFindWindowByName(name);
 	if(!window)
 	{
-		cvNamedWindow(name, 1);
+        #ifndef HAVE_OPENGL
+		    cvNamedWindow(name, CV_WINDOW_AUTOSIZE);
+        #else
+		    cvNamedWindow(name, CV_WINDOW_AUTOSIZE | CV_WINDOW_OPENGL);
+        #endif
+
 		window = icvFindWindowByName(name);
 	}
 
@@ -757,6 +1404,15 @@ cvShowImage( const char* name, const CvArr* arr )
 
     CV_CALL( image = cvGetMat( arr, &stub ));
 
+#ifdef HAVE_OPENGL
+    if (window->useGl)
+    {
+        cv::Mat im(image);
+        cv::imshow(name, im);
+        return;
+    }
+#endif
+
     if (window->image)
         // if there is something wrong with these system calls, we cannot display image...
         if (icvGetBitmapData( window, &size, &channels, &dst_ptr ))
@@ -1069,6 +1725,13 @@ static LRESULT CALLBACK HighGUIProc( HWND hwnd, UINT uMsg, WPARAM wParam, LPARAM
             //DeleteDC(hdc);
             EndPaint(hwnd, &paint);
         }
+#ifdef HAVE_OPENGL
+        else if(window->useGl) 
+        {
+            drawGl(window);            
+            return DefWindowProc(hwnd, uMsg, wParam, lParam);
+        }
+#endif
         else
         {
             return DefWindowProc(hwnd, uMsg, wParam, lParam);
@@ -1094,6 +1757,15 @@ static LRESULT CALLBACK HighGUIProc( HWND hwnd, UINT uMsg, WPARAM wParam, LPARAM
     case WM_KEYDOWN:
         window->last_key = (int)wParam;
         return 0;
+
+    case WM_SIZE:
+        window->width = LOWORD(lParam);
+        window->height = HIWORD(lParam);
+
+#ifdef HAVE_OPENGL
+        if (window->useGl)
+            resizeGl(window);
+#endif
     }
 
     return DefWindowProc(hwnd, uMsg, wParam, lParam);

samples/gpu/highgui_gpu.cpp

+#include <iostream>
+#include <string>
+
+#include "opencv2/core/core.hpp"
+#include "opencv2/core/gpumat.hpp"
+#include "opencv2/gpu/gpu.hpp"
+#include "opencv2/highgui/highgui.hpp"
+#include "opencv2/contrib/contrib.hpp"
+
+using namespace std;
+using namespace cv;
+using namespace cv::gpu;
+
+struct Timer
+{
+    Timer(const string& msg_)
+    {
+        msg = msg_;
+
+        tm.reset();
+        tm.start();
+    }
+
+    ~Timer()
+    {
+        tm.stop();
+        cout << msg << " " << tm.getTimeMilli() << " ms\n";
+    }
+
+    string msg;
+    TickMeter tm;
+};
+
+int main(int argc, char* argv[])
+{
+    if (argc < 2)
+    {
+        cout << "Usage: " << argv[0] << " image" << endl;
+        return -1;
+    }
+
+    try
+    {
+        bool haveCuda = getCudaEnabledDeviceCount() > 0;
+
+        namedWindow("OpenGL Mat", WINDOW_OPENGL | WINDOW_AUTOSIZE);
+        namedWindow("OpenGL GlBuffer", WINDOW_OPENGL | WINDOW_AUTOSIZE);
+        namedWindow("OpenGL GlTexture", WINDOW_OPENGL | WINDOW_AUTOSIZE);
+        if (haveCuda)
+            namedWindow("OpenGL GpuMat", WINDOW_OPENGL | WINDOW_AUTOSIZE);
+        namedWindow("Mat", WINDOW_AUTOSIZE);
+
+        Mat img = imread(argv[1]);
+        
+        if (haveCuda)
+            setGlDevice();
+
+        setOpenGlContext("OpenGL GlBuffer");
+        GlBuffer buf(img, GlBuffer::TEXTURE_BUFFER);
+
+        setOpenGlContext("OpenGL GlTexture");
+        GlTexture tex(img);
+        
+        GpuMat d_img;
+        if (haveCuda)
+            d_img.upload(img);
+
+        cout << "=== First call\n\n";
+
+        {
+            Timer t("OpenGL Mat      ");
+            imshow("OpenGL Mat", img);
+        }
+        {
+            Timer t("OpenGL GlBuffer ");
+            imshow("OpenGL GlBuffer", buf);
+        }
+        {
+            Timer t("OpenGL GlTexture");
+            imshow("OpenGL GlTexture", tex);
+        }
+        if (haveCuda)
+        {
+            Timer t("OpenGL GpuMat   ");
+            imshow("OpenGL GpuMat", d_img);
+        }
+        {
+            Timer t("Mat             ");
+            imshow("Mat", img);
+        }
+
+        cout << "\n=== Second call\n\n";   
+
+        {
+            Timer t("OpenGL Mat      ");
+            imshow("OpenGL Mat", img);
+        }
+        {
+            Timer t("OpenGL GlBuffer ");
+            imshow("OpenGL GlBuffer", buf);
+        }
+        {
+            Timer t("OpenGL GlTexture");
+            imshow("OpenGL GlTexture", tex);
+        }
+        if (haveCuda)
+        {
+            Timer t("OpenGL GpuMat   ");
+            imshow("OpenGL GpuMat", d_img);
+        }
+        {
+            Timer t("Mat             ");
+            imshow("Mat", img);
+        }
+
+        cout << "\n";
+
+        waitKey();
+    }
+    catch(const exception& e)
+    {
+        cout << e.what() << endl;
+    }
+
+    return 0;
+}
\ No newline at end of file