ceph/src/boost/libs/compute/example/mandelbrot.cpp

   1 //---------------------------------------------------------------------------//
   2 // Copyright (c) 2013-2014 Kyle Lutz <kyle.r.lutz@gmail.com>
   3 //
   4 // Distributed under the Boost Software License, Version 1.0
   5 // See accompanying file LICENSE_1_0.txt or copy at
   6 // http://www.boost.org/LICENSE_1_0.txt
   7 //
   8 // See http://boostorg.github.com/compute for more information.
   9 //---------------------------------------------------------------------------//
  10
  11 #include <iostream>
  12 #include <algorithm>
  13
  14 #include <QtGlobal>
  15 #if QT_VERSION >= 0x050000
  16 #include <QtWidgets>
  17 #else
  18 #include <QtGui>
  19 #endif
  20 #include <QtOpenGL>
  21
  22 #ifndef Q_MOC_RUN
  23 #include <boost/compute/command_queue.hpp>
  24 #include <boost/compute/kernel.hpp>
  25 #include <boost/compute/program.hpp>
  26 #include <boost/compute/system.hpp>
  27 #include <boost/compute/interop/opengl.hpp>
  28 #include <boost/compute/utility/dim.hpp>
  29 #include <boost/compute/utility/source.hpp>
  30 #endif // Q_MOC_RUN
  31
  32 namespace compute = boost::compute;
  33
  34 // opencl source code
  35 const char source[] = BOOST_COMPUTE_STRINGIZE_SOURCE(
  36     // map value to color
  37     float4 color(uint i)
  38     {
  39         uchar c = i;
  40         uchar x = 35;
  41         uchar y = 25;
  42         uchar z = 15;
  43         uchar max = 255;
  44
  45         if(i == 256)
  46             return (float4)(0, 0, 0, 255);
  47         else
  48             return (float4)(max-x*i, max-y*i, max-z*i, max) / 255.f;
  49     }
  50
  51     __kernel void mandelbrot(__write_only image2d_t image)
  52     {
  53         const uint x_coord = get_global_id(0);
  54         const uint y_coord = get_global_id(1);
  55         const uint width = get_global_size(0);
  56         const uint height = get_global_size(1);
  57
  58         float x_origin = ((float) x_coord / width) * 3.25f - 2.0f;
  59         float y_origin = ((float) y_coord / height) * 2.5f - 1.25f;
  60
  61         float x = 0.0f;
  62         float y = 0.0f;
  63
  64         uint i = 0;
  65         while(x*x + y*y <= 4.f && i < 256){
  66             float tmp = x*x - y*y + x_origin;
  67             y = 2*x*y + y_origin;
  68             x = tmp;
  69             i++;
  70         }
  71
  72         int2 coord = { x_coord, y_coord };
  73         write_imagef(image, coord, color(i));
  74     };
  75 );
  76
  77 class MandelbrotWidget : public QGLWidget
  78 {
  79     Q_OBJECT
  80
  81 public:
  82     MandelbrotWidget(QWidget *parent = 0);
  83     ~MandelbrotWidget();
  84
  85     void initializeGL();
  86     void resizeGL(int width, int height);
  87     void paintGL();
  88     void keyPressEvent(QKeyEvent* event);
  89
  90 private:
  91     compute::context context_;
  92     compute::command_queue queue_;
  93     compute::program program_;
  94     GLuint gl_texture_;
  95     compute::opengl_texture cl_texture_;
  96 };
  97
  98 MandelbrotWidget::MandelbrotWidget(QWidget *parent)
  99     : QGLWidget(parent)
 100 {
 101     gl_texture_ = 0;
 102 }
 103
 104 MandelbrotWidget::~MandelbrotWidget()
 105 {
 106 }
 107
 108 void MandelbrotWidget::initializeGL()
 109 {
 110     // setup opengl
 111     glDisable(GL_LIGHTING);
 112
 113     // create the OpenGL/OpenCL shared context
 114     context_ = compute::opengl_create_shared_context();
 115
 116     // get gpu device
 117     compute::device gpu = context_.get_device();
 118     std::cout << "device: " << gpu.name() << std::endl;
 119
 120     // setup command queue
 121     queue_ = compute::command_queue(context_, gpu);
 122
 123     // build mandelbrot program
 124     program_ = compute::program::create_with_source(source, context_);
 125     program_.build();
 126 }
 127
 128 void MandelbrotWidget::resizeGL(int width, int height)
 129 {
 130 #if QT_VERSION >= 0x050000
 131     // scale height/width based on device pixel ratio
 132     width /= windowHandle()->devicePixelRatio();
 133     height /= windowHandle()->devicePixelRatio();
 134 #endif
 135
 136     // resize viewport
 137     glViewport(0, 0, width, height);
 138
 139     // delete old texture
 140     if(gl_texture_){
 141       glDeleteTextures(1, &gl_texture_);
 142       gl_texture_ = 0;
 143     }
 144
 145     // generate new texture
 146     glGenTextures(1, &gl_texture_);
 147     glBindTexture(GL_TEXTURE_2D, gl_texture_);
 148     glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_REPEAT);
 149     glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_REPEAT);
 150     glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR);
 151     glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR);
 152     glTexImage2D(
 153         GL_TEXTURE_2D, 0, GL_RGBA, width, height, 0, GL_RGBA, GL_UNSIGNED_BYTE, 0
 154     );
 155
 156     // create opencl object for the texture
 157     cl_texture_ = compute::opengl_texture(
 158         context_, GL_TEXTURE_2D, 0, gl_texture_, CL_MEM_WRITE_ONLY
 159     );
 160 }
 161
 162 void MandelbrotWidget::paintGL()
 163 {
 164     using compute::dim;
 165
 166     float w = width();
 167     float h = height();
 168
 169     glMatrixMode(GL_PROJECTION);
 170     glLoadIdentity();
 171     glOrtho(0.0, w, 0.0, h, -1.0, 1.0);
 172     glMatrixMode(GL_MODELVIEW);
 173     glLoadIdentity();
 174
 175     // setup the mandelbrot kernel
 176     compute::kernel kernel(program_, "mandelbrot");
 177     kernel.set_arg(0, cl_texture_);
 178
 179     // acquire the opengl texture so it can be used in opencl
 180     compute::opengl_enqueue_acquire_gl_objects(1, &cl_texture_.get(), queue_);
 181
 182     // execute the mandelbrot kernel
 183     queue_.enqueue_nd_range_kernel(
 184         kernel, dim(0, 0), dim(width(), height()), dim(1, 1)
 185     );
 186
 187     // release the opengl texture so it can be used by opengl
 188     compute::opengl_enqueue_release_gl_objects(1, &cl_texture_.get(), queue_);
 189
 190     // ensure opencl is finished before rendering in opengl
 191     queue_.finish();
 192
 193     // draw a single quad with the mandelbrot image texture
 194     glEnable(GL_TEXTURE_2D);
 195     glBindTexture(GL_TEXTURE_2D, gl_texture_);
 196
 197     glBegin(GL_QUADS);
 198     glTexCoord2f(0, 0); glVertex2f(0, 0);
 199     glTexCoord2f(0, 1); glVertex2f(0, h);
 200     glTexCoord2f(1, 1); glVertex2f(w, h);
 201     glTexCoord2f(1, 0); glVertex2f(w, 0);
 202     glEnd();
 203 }
 204
 205 void MandelbrotWidget::keyPressEvent(QKeyEvent* event)
 206 {
 207     if(event->key() == Qt::Key_Escape) {
 208         this->close();
 209     }
 210 }
 211
 212 // the mandelbrot example shows how to create a mandelbrot image in
 213 // OpenCL and render the image as a texture in OpenGL
 214 int main(int argc, char *argv[])
 215 {
 216     QApplication app(argc, argv);
 217
 218     MandelbrotWidget widget;
 219     widget.show();
 220
 221     return app.exec();
 222 }
 223
 224 #include "mandelbrot.moc"