Qt camera integration

miklos/project_integration/camera.cpp

+#include "camera.h"
+
+#include <opencv2/imgproc/imgproc.hpp>
+
+Camera camera;
+
+bool Camera::init(int device)
+{
+    capture_.open(device);
+    if (!capture_.isOpened())
+        return false;
+
+    int camW = 640;
+    int camH = 480;
+  	capture_.set(CV_CAP_PROP_FRAME_WIDTH, camW);
+  	capture_.set(CV_CAP_PROP_FRAME_HEIGHT, camH);
+
+    capture();
+    width_  = frame_.cols;
+    height_ = frame_.rows;
+
+    return true;
+}
+
+void Camera::run()
+{
+    QMutexLocker locker(&mutex);
+    while (true) {  // TODO: never terminates
+        frameCopied.wait(&mutex);
+        capture();
+        emit newFrame();
+    }
+}
+
+void Camera::capture()
+{
+    capture_ >> frame_;
+    cvtColor(frame_, frame_, CV_BGR2GRAY);
+
+    frame_.convertTo(frame_, CV_32F);
+    frame_ /= 255.f;
+}

miklos/project_integration/camera.h

+#ifndef CAMERA_H
+#define CAMERA_H
+
+#include <QMutex>
+#include <QThread>
+#include <QWaitCondition>
+
+#include <opencv2/highgui/highgui.hpp>
+
+class Camera : public QThread {
+    Q_OBJECT
+
+public:
+    bool init(int device);
+    
+    size_t width() const { return width_; }
+    size_t height() const { return height_; }
+
+    float *data() const {
+        return reinterpret_cast<float *>(frame_.data);
+    }
+
+    QMutex mutex;
+    QWaitCondition frameCopied;
+
+signals:
+    void newFrame();
+
+protected:
+    void run();
+
+private:
+    cv::VideoCapture capture_;
+    cv::Mat frame_;
+    size_t width_, height_;
+    
+    void capture();
+};
+
+extern Camera camera;
+
+#endif // CAMERA_H

miklos/project_integration/glwidget.cpp

 #include "glwidget.h"
+#include "camera.h"
 #include "kernel.h"
 
 #include <QGLFunctions>
 
-#include <opencv2/highgui/highgui.hpp>
-#include <opencv2/imgproc/imgproc.hpp>
-
 #include <cuda_runtime_api.h>
 #include <cuda_gl_interop.h>
 
-#include <cstdio>
-
-GLuint positionsVBO;
-struct cudaGraphicsResource* positionsVBO_CUDA;
+static GLuint pixelsVBO;
+static struct cudaGraphicsResource* pixelsVBO_CUDA;
 
 GlWidget::GlWidget(QWidget *parent)
-    : QGLWidget(QGLFormat(), parent), func(context()), start(time(NULL))
+    : QGLWidget(QGLFormat(), parent), gl(context())
 {
 }
 
@@ -25,59 +21,49 @@ GlWidget::~GlWidget()
 
 QSize GlWidget::sizeHint() const
 {
-    return QSize(640, 480);
+    return QSize(camera.width(), camera.height());
 }
 
 void GlWidget::initializeGL()
 {
-    makeCurrent();
-
     // Explicitly set device 0
     cudaGLSetGLDevice(0);
 
     // Create buffer object and register it with CUDA
-    func.glGenBuffers(1, &positionsVBO);
-    func.glBindBuffer(GL_PIXEL_UNPACK_BUFFER, positionsVBO);
-    size_t size = (size_t)640 * 480 * 4 * sizeof(unsigned char); // TODO
-    func.glBufferData(GL_PIXEL_UNPACK_BUFFER, size, 0, GL_DYNAMIC_DRAW);
-    //func.glBindBuffer(GL_PIXEL_UNPACK_BUFFER, 0);  // TODO: what is it?
-    cudaGraphicsGLRegisterBuffer(&positionsVBO_CUDA, positionsVBO, cudaGraphicsMapFlagsWriteDiscard);
+    gl.glGenBuffers(1, &pixelsVBO);
+    gl.glBindBuffer(GL_PIXEL_UNPACK_BUFFER, pixelsVBO);
+    size_t size = camera.width() * camera.height() * 4 * sizeof(unsigned char);
+    gl.glBufferData(GL_PIXEL_UNPACK_BUFFER, size, 0, GL_DYNAMIC_DRAW);
+    cudaGraphicsGLRegisterBuffer(&pixelsVBO_CUDA, pixelsVBO, cudaGraphicsMapFlagsWriteDiscard);
 }
 
 void GlWidget::paintGL()
 {
     // Map buffer object for writing from CUDA
-    unsigned char *positions;
-    cudaGraphicsMapResources(1, &positionsVBO_CUDA, 0);
-    size_t num_bytes; 
-    cudaGraphicsResourceGetMappedPointer((void**)&positions, &num_bytes,  positionsVBO_CUDA);
-
-    float *d_in;
-    cudaMalloc((void **)&d_in, (size_t)640*480*sizeof(float));
-
-    extern cv::VideoCapture camera;
-    cv::Mat mIn;
-    camera >> mIn;
-    printf("width: %d; height: %d;\n", mIn.cols, mIn.rows);
-    cvtColor(mIn, mIn, CV_BGR2GRAY);
-
-    // convert to float representation (opencv loads image values as single bytes by default)
-    mIn.convertTo(mIn, CV_32F);
-    // convert range of each channel to [0,1] (opencv default is [0,255])
-    mIn /= 255.f;
-
-    cudaMemcpy(d_in, mIn.data, (size_t)640*480*sizeof(float), cudaMemcpyHostToDevice);
+    void *d_out;
+    cudaGraphicsMapResources(1, &pixelsVBO_CUDA, 0);
+    size_t size; 
+    cudaGraphicsResourceGetMappedPointer(&d_out, &size,  pixelsVBO_CUDA);
+
+    size_t inBytes = camera.width() * camera.height() * sizeof(float);
+    void *d_in;
+    cudaMalloc(&d_in, inBytes);
+    {
+        QMutexLocker locker(&camera.mutex);
+        cudaMemcpy(d_in, camera.data(), inBytes, cudaMemcpyHostToDevice);
+        cudaDeviceSynchronize();
+        camera.frameCopied.wakeAll();
+    }
 
     // Execute kernel
-    executeKernel(d_in, positions, 640, 480, float(time(0) - start) * 0.1);
+    executeKernel(d_in, d_out, camera.width(), camera.height());
 
     cudaFree(d_in);
 
     // Unmap buffer object
-    cudaGraphicsUnmapResources(1, &positionsVBO_CUDA, 0);
+    cudaGraphicsUnmapResources(1, &pixelsVBO_CUDA, 0);
 
     // Render from buffer object
     glClear(GL_COLOR_BUFFER_BIT | GL_DEPTH_BUFFER_BIT);
-    //func.glBindBuffer(GL_ARRAY_BUFFER, positionsVBO);  // what is it?
-    glDrawPixels( 640, 480, GL_RGBA, GL_UNSIGNED_BYTE, 0 );  // TODO
+    glDrawPixels(camera.width(), camera.height(), GL_RGBA, GL_UNSIGNED_BYTE, 0);
 }

miklos/project_integration/glwidget.h

@@ -18,8 +18,7 @@ protected:
     void paintGL();
 
 private:
-    QGLFunctions func;
-    time_t start;
+    QGLFunctions gl;
 };
 
 #endif // GLWIDGET_H

miklos/project_integration/hello-opengl.pro

@@ -5,8 +5,8 @@ TARGET = hello-opengl
 
 TEMPLATE = app
 
-SOURCES += main.cpp glwidget.cpp
-HEADERS += glwidget.h kernel.h
+SOURCES += main.cpp glwidget.cpp camera.cpp
+HEADERS += glwidget.h kernel.h camera.h
 
 CUDA_SOURCES += kernel.cu
 

miklos/project_integration/kernel.cu

 #include "kernel.h"
 
-__global__ void createVertices(float *in, uchar4* positions, int width, int height)
+__global__ void createVertices(float *in, uchar4* pixel, int w, int h)
 {
     unsigned int x = blockIdx.x * blockDim.x + threadIdx.x;
     unsigned int y = blockIdx.y * blockDim.y + threadIdx.y;
 
-    unsigned char intensity = roundf(255 * in[y * width + x]);
+    unsigned char intensity = roundf(255 * in[y * w + x]);
 
     // Write positions
-    positions[y * width + x].x = intensity;
-    positions[y * width + x].y = intensity;
-    positions[y * width + x].z = intensity;
-    positions[y * width + x].w = 255;
+    size_t i = x + w*(h-1 - y);
+    pixel[i].x = intensity;
+    pixel[i].y = intensity;
+    pixel[i].z = intensity;
+    pixel[i].w = 255;
 }
 
-void executeKernel(float *d_in, void *positions_, int width, int height, float time)
+void executeKernel(void *d_in, void *d_out, size_t w, size_t h)
 {
-    uchar4 *positions = (uchar4 *)positions_;
+    float *in = reinterpret_cast<float *>(d_in);
+    uchar4 *pixel = reinterpret_cast<uchar4 *>(d_out);
     dim3 dimBlock(16, 16, 1);
-    dim3 dimGrid(width / dimBlock.x, height / dimBlock.y, 1);
-    createVertices<<<dimGrid, dimBlock>>>(d_in, positions, width, height);
+    dim3 dimGrid(w / dimBlock.x, h / dimBlock.y, 1);
+    createVertices<<<dimGrid, dimBlock>>>(in, pixel, w, h);
 }

miklos/project_integration/kernel.h

-#ifndef _KERNEL_H
-#define _KERNEL_H
+#ifndef KERNEL_H
+#define KERNEL_H
 
-extern "C" void executeKernel(float *d_in, void *positions_, int width, int height, float time);
+extern "C" void executeKernel(void *d_in, void *d_out, size_t width, size_t height);
 
-#endif // _KERNEL_H
+#endif // KERNEL_H

miklos/project_integration/main.cpp

-#include <iostream>
-
-#include <QApplication>
-
-#include <opencv2/highgui/highgui.hpp>
-
+#include "camera.h"
 #include "glwidget.h"
 
-cv::VideoCapture camera;
+#include <iostream>
+#include <QApplication>
 
 int main(int argc, char *argv[])
 {
     QApplication app(argc, argv);
+    GlWidget w;
 
     // Init camera
-  	camera.open(0);
-  	if(!camera.isOpened()) {
+    if(!camera.init(0)) {
         std::cerr << "ERROR: Could not open camera" << std::endl;
         return 1;
     }
-    int camW = 640;
-    int camH = 480;
-  	camera.set(CV_CAP_PROP_FRAME_WIDTH,camW);
-  	camera.set(CV_CAP_PROP_FRAME_HEIGHT,camH);
 
-    GlWidget w;
+    QObject::connect(&camera, SIGNAL(newFrame()), &w, SLOT(updateGL()));
+
+    camera.start();
     w.show();
 
     return app.exec();