1.1 --- a/Makefile	Mon Jul 12 10:38:07 2010 +0300
     1.2 +++ b/Makefile	Tue Jul 13 03:38:29 2010 +0300
     1.3 @@ -4,7 +4,7 @@
     1.4  
     1.5  CXX = g++
     1.6  CXXFLAGS = -pedantic -Wall -g
     1.7 -LDFLAGS = -framework OpenCL
     1.8 +LDFLAGS = -framework OpenGL -framework GLUT -framework OpenCL
     1.9  
    1.10  $(bin): $(obj)
    1.11  	$(CXX) -o $@ $(obj) $(LDFLAGS)

     2.1 --- a/rt.cl	Mon Jul 12 10:38:07 2010 +0300
     2.2 +++ b/rt.cl	Tue Jul 13 03:38:29 2010 +0300
     2.3 @@ -1,6 +1,6 @@
     2.4  struct RendInfo {
     2.5  	int xsz, ysz;
     2.6 -	int num_sph;
     2.7 +	int num_sph, num_lights;
     2.8  	int max_iter;
     2.9  };
    2.10  
    2.11 @@ -10,6 +10,10 @@
    2.12  	float4 color;
    2.13  };
    2.14  
    2.15 +struct Light {
    2.16 +	float4 pos, color;
    2.17 +};
    2.18 +
    2.19  struct Ray {
    2.20  	float4 origin, dir;
    2.21  };
    2.22 @@ -26,6 +30,7 @@
    2.23  __kernel void render(__global float4 *fb,
    2.24  		__global const struct RendInfo *rinf,
    2.25  		__global const struct Sphere *sphlist,
    2.26 +		__global const struct Light *lights,
    2.27  		__global const struct Ray *primrays)
    2.28  {
    2.29  	int idx = get_global_id(0);
    2.30 @@ -38,6 +43,8 @@
    2.31  	} else {
    2.32  		fb[idx] = (float4)(0, 0, 0, 1);
    2.33  	}
    2.34 +
    2.35 +	fb[idx] = primrays[idx].dir * 0.5 + 0.5;
    2.36  }
    2.37  
    2.38  bool intersect(struct Ray ray,

     3.1 --- a/src/clray.cc	Mon Jul 12 10:38:07 2010 +0300
     3.2 +++ b/src/clray.cc	Tue Jul 13 03:38:29 2010 +0300
     3.3 @@ -1,94 +1,126 @@
     3.4  #include <stdio.h>
     3.5 +#include <stdlib.h>
     3.6  #include <string.h>
     3.7 -#include <math.h>
     3.8  #include <errno.h>
     3.9 -#include <assert.h>
    3.10 -#include "ocl.h"
    3.11 +#ifndef __APPLE__
    3.12 +#include <GL/glut.h>
    3.13 +#else
    3.14 +#include <GLUT/glut.h>
    3.15 +#endif
    3.16 +#include "rt.h"
    3.17  
    3.18 -struct RendInfo {
    3.19 -	int xsz, ysz;
    3.20 -	int num_sph;
    3.21 -	int max_iter;
    3.22 -} __attribute__((packed));
    3.23 -
    3.24 -struct Sphere {
    3.25 -	cl_float4 pos;
    3.26 -	cl_float radius;
    3.27 -
    3.28 -	cl_float4 color;
    3.29 -} __attribute__((packed));
    3.30 -
    3.31 -struct Ray {
    3.32 -	cl_float4 origin, dir;
    3.33 -} __attribute__((packed));
    3.34 -
    3.35 -
    3.36 -Ray get_primary_ray(int x, int y, int w, int h, float vfov_deg);
    3.37 +void cleanup();
    3.38 +void disp();
    3.39 +void reshape(int x, int y);
    3.40 +void keyb(unsigned char key, int x, int y);
    3.41 +void mouse(int bn, int status, int x, int y);
    3.42 +void motion(int x, int y);
    3.43  bool write_ppm(const char *fname, float *fb, int xsz, int ysz);
    3.44  
    3.45 -int main()
    3.46 +static float *fb;
    3.47 +static int xsz, ysz;
    3.48 +static bool need_update = true;
    3.49 +
    3.50 +int main(int argc, char **argv)
    3.51  {
    3.52 -	const int xsz = 800;
    3.53 -	const int ysz = 600;
    3.54 +	glutInitWindowSize(800, 600);
    3.55 +	glutInit(&argc, argv);
    3.56 +	glutInitDisplayMode(GLUT_RGB | GLUT_DOUBLE);
    3.57 +	glutCreateWindow("OpenCL Raytracer");
    3.58  
    3.59 -	Sphere sphlist[] = {
    3.60 -		{{0, 0, 10, 1}, 1.0, {1, 0, 0, 1}}
    3.61 -	};
    3.62 -	RendInfo rinf = {xsz, ysz, sizeof sphlist / sizeof *sphlist, 6};
    3.63 -	Ray *prim_rays = new Ray[xsz * ysz];
    3.64 -	float *fb = new float[xsz * ysz * 4];
    3.65 +	xsz = glutGet(GLUT_WINDOW_WIDTH);
    3.66 +	ysz = glutGet(GLUT_WINDOW_HEIGHT);
    3.67  
    3.68 -	/* calculate primary rays */
    3.69 -	for(int i=0; i<ysz; i++) {
    3.70 -		for(int j=0; j<xsz; j++) {
    3.71 -			prim_rays[i * xsz + j] = get_primary_ray(j, i, xsz, ysz, 45.0);
    3.72 -		}
    3.73 -	}
    3.74 +	glutDisplayFunc(disp);
    3.75 +	glutReshapeFunc(reshape);
    3.76 +	glutKeyboardFunc(keyb);
    3.77 +	glutMouseFunc(mouse);
    3.78 +	glutMotionFunc(motion);
    3.79  
    3.80 -	/* setup opencl */
    3.81 -	CLProgram prog("render");
    3.82 -	if(!prog.load("rt.cl")) {
    3.83 +	fb = new float[xsz * ysz * 4];
    3.84 +	if(!init_renderer(xsz, ysz, fb)) {
    3.85  		return 1;
    3.86  	}
    3.87 +	atexit(cleanup);
    3.88  
    3.89 -	prog.set_arg_buffer(0, ARG_WR, xsz * ysz * 4 * sizeof(float), fb);
    3.90 -	prog.set_arg_buffer(1, ARG_RD, sizeof rinf, &rinf);
    3.91 -	prog.set_arg_buffer(2, ARG_RD, sizeof sphlist, sphlist);
    3.92 -	prog.set_arg_buffer(3, ARG_RD, xsz * ysz * sizeof *prim_rays, prim_rays);
    3.93 +	/*glGenTextures(1, &tex);
    3.94 +	glBindTexture(GL_TEXTURE_2D, tex);*/
    3.95 +	glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_CLAMP);
    3.96 +	glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_CLAMP);
    3.97 +	glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR);
    3.98 +	glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR);
    3.99 +	glTexImage2D(GL_TEXTURE_2D, 0, 4, xsz, ysz, 0, GL_RGBA, GL_FLOAT, 0);
   3.100  
   3.101 -	if(!prog.run(1, xsz * ysz)) {
   3.102 -		return 1;
   3.103 -	}
   3.104 -
   3.105 -	CLMemBuffer *mbuf = prog.get_arg_buffer(0);
   3.106 -	map_mem_buffer(mbuf, MAP_RD);
   3.107 -	if(!write_ppm("out.ppm", fb, xsz, ysz)) {
   3.108 -		return 1;
   3.109 -	}
   3.110 -	unmap_mem_buffer(mbuf);
   3.111 -
   3.112 -	delete [] fb;
   3.113 -	delete [] prim_rays;
   3.114 -
   3.115 +	glutMainLoop();
   3.116  	return 0;
   3.117  }
   3.118  
   3.119 -Ray get_primary_ray(int x, int y, int w, int h, float vfov_deg)
   3.120 +void cleanup()
   3.121  {
   3.122 -	float vfov = M_PI * vfov_deg / 180.0;
   3.123 -	float aspect = (float)w / (float)h;
   3.124 +	delete [] fb;
   3.125 +	destroy_renderer();
   3.126 +}
   3.127  
   3.128 -	float ysz = 2.0;
   3.129 -	float xsz = aspect * ysz;
   3.130 +void disp()
   3.131 +{
   3.132 +	if(need_update) {
   3.133 +		render();
   3.134 +		glTexSubImage2D(GL_TEXTURE_2D, 0, 0, 0, xsz, ysz, GL_RGBA, GL_FLOAT, fb);
   3.135 +		need_update = false;
   3.136 +	}
   3.137  
   3.138 -	float px = ((float)x / (float)w) * xsz - xsz / 2.0;
   3.139 -	float py = 1.0 - ((float)y / (float)h) * ysz;
   3.140 -	float pz = 1.0 / tan(0.5 * vfov);
   3.141 +	glEnable(GL_TEXTURE_2D);
   3.142  
   3.143 -	pz *= 1000.0;
   3.144 +	glBegin(GL_QUADS);
   3.145 +	glTexCoord2f(0, 1); glVertex2f(-1, -1);
   3.146 +	glTexCoord2f(1, 1); glVertex2f(1, -1);
   3.147 +	glTexCoord2f(1, 0); glVertex2f(1, 1);
   3.148 +	glTexCoord2f(0, 0); glVertex2f(-1, 1);
   3.149 +	glEnd();
   3.150  
   3.151 -	Ray ray = {{0, 0, 0, 1}, {px, py, pz, 1}};
   3.152 -	return ray;
   3.153 +	glDisable(GL_TEXTURE_2D);
   3.154 +
   3.155 +	glutSwapBuffers();
   3.156 +}
   3.157 +
   3.158 +void reshape(int x, int y)
   3.159 +{
   3.160 +	glViewport(0, 0, x, y);
   3.161 +
   3.162 +	/* reallocate the framebuffer */
   3.163 +	/*delete [] fb;
   3.164 +	fb = new float[x * y * 4];
   3.165 +	set_framebuffer(fb, x, y);*/
   3.166 +}
   3.167 +
   3.168 +void keyb(unsigned char key, int x, int y)
   3.169 +{
   3.170 +	switch(key) {
   3.171 +	case 27:
   3.172 +		exit(0);
   3.173 +
   3.174 +	case 's':
   3.175 +		if(write_ppm("shot.ppm", fb, xsz, ysz)) {
   3.176 +			printf("captured screenshot shot.ppm\n");
   3.177 +		}
   3.178 +		break;
   3.179 +
   3.180 +	case 'r':
   3.181 +		need_update = true;
   3.182 +		glutPostRedisplay();
   3.183 +		break;
   3.184 +
   3.185 +	default:
   3.186 +		break;
   3.187 +	}
   3.188 +}
   3.189 +
   3.190 +void mouse(int bn, int state, int x, int y)
   3.191 +{
   3.192 +}
   3.193 +
   3.194 +void motion(int x, int y)
   3.195 +{
   3.196  }
   3.197  
   3.198  bool write_ppm(const char *fname, float *fb, int xsz, int ysz)

     4.1 --- /dev/null	Thu Jan 01 00:00:00 1970 +0000
     4.2 +++ b/src/rt.cc	Tue Jul 13 03:38:29 2010 +0300
     4.3 @@ -0,0 +1,112 @@
     4.4 +#include <stdio.h>
     4.5 +#include <math.h>
     4.6 +#include <assert.h>
     4.7 +#include "ocl.h"
     4.8 +
     4.9 +struct RendInfo {
    4.10 +	int xsz, ysz;
    4.11 +	int num_sph, num_lights;
    4.12 +	int max_iter;
    4.13 +} __attribute__((packed));
    4.14 +
    4.15 +struct Sphere {
    4.16 +	cl_float4 pos;
    4.17 +	cl_float radius;
    4.18 +
    4.19 +	cl_float4 color;
    4.20 +} __attribute__((packed));
    4.21 +
    4.22 +struct Ray {
    4.23 +	cl_float4 origin, dir;
    4.24 +} __attribute__((packed));
    4.25 +
    4.26 +struct Light {
    4.27 +	cl_float4 pos, color;
    4.28 +} __attribute__((packed));
    4.29 +
    4.30 +
    4.31 +static Ray get_primary_ray(int x, int y, int w, int h, float vfov_deg);
    4.32 +
    4.33 +static Ray *prim_rays;
    4.34 +static CLProgram *prog;
    4.35 +static int global_size;
    4.36 +
    4.37 +bool init_renderer(int xsz, int ysz, float *fb)
    4.38 +{
    4.39 +	Sphere sphlist[] = {
    4.40 +		{{0, 0, 10, 1}, 1.0, {1, 0, 0, 1}}
    4.41 +	};
    4.42 +	Light lightlist[] = {
    4.43 +		{{-10, 10, -20, 1}, {1, 1, 1, 1}}
    4.44 +	};
    4.45 +	RendInfo rinf = {
    4.46 +		xsz, ysz,
    4.47 +		sizeof sphlist / sizeof *sphlist,
    4.48 +		sizeof lightlist / sizeof *lightlist,
    4.49 +		6
    4.50 +	};
    4.51 +	
    4.52 +	/* calculate primary rays */
    4.53 +	prim_rays = new Ray[xsz * ysz];
    4.54 +
    4.55 +	for(int i=0; i<ysz; i++) {
    4.56 +		for(int j=0; j<xsz; j++) {
    4.57 +			prim_rays[i * xsz + j] = get_primary_ray(j, i, xsz, ysz, 45.0);
    4.58 +		}
    4.59 +	}
    4.60 +
    4.61 +	/* setup opencl */
    4.62 +	prog = new CLProgram("render");
    4.63 +	if(!prog->load("rt.cl")) {
    4.64 +		return 1;
    4.65 +	}
    4.66 +
    4.67 +	/* setup argument buffers */
    4.68 +	prog->set_arg_buffer(0, ARG_WR, xsz * ysz * 4 * sizeof(float), fb);
    4.69 +	prog->set_arg_buffer(1, ARG_RD, sizeof rinf, &rinf);
    4.70 +	prog->set_arg_buffer(2, ARG_RD, sizeof sphlist, sphlist);
    4.71 +	prog->set_arg_buffer(3, ARG_RD, sizeof lightlist, lightlist);
    4.72 +	prog->set_arg_buffer(4, ARG_RD, xsz * ysz * sizeof *prim_rays, prim_rays);
    4.73 +
    4.74 +	global_size = xsz * ysz;
    4.75 +	return true;
    4.76 +}
    4.77 +
    4.78 +void destroy_renderer()
    4.79 +{
    4.80 +	delete [] prim_rays;
    4.81 +	delete prog;
    4.82 +}
    4.83 +
    4.84 +bool render()
    4.85 +{
    4.86 +	if(!prog->run(1, global_size)) {
    4.87 +		return false;
    4.88 +	}
    4.89 +
    4.90 +	CLMemBuffer *mbuf = prog->get_arg_buffer(0);
    4.91 +	map_mem_buffer(mbuf, MAP_RD);
    4.92 +	/*if(!write_ppm("out.ppm", fb, xsz, ysz)) {
    4.93 +		return 1;
    4.94 +	}*/
    4.95 +	unmap_mem_buffer(mbuf);
    4.96 +	return true;
    4.97 +}
    4.98 +
    4.99 +static Ray get_primary_ray(int x, int y, int w, int h, float vfov_deg)
   4.100 +{
   4.101 +	float vfov = M_PI * vfov_deg / 180.0;
   4.102 +	float aspect = (float)w / (float)h;
   4.103 +
   4.104 +	float ysz = 2.0;
   4.105 +	float xsz = aspect * ysz;
   4.106 +
   4.107 +	float px = ((float)x / (float)w) * xsz - xsz / 2.0;
   4.108 +	float py = 1.0 - ((float)y / (float)h) * ysz;
   4.109 +	float pz = 1.0 / tan(0.5 * vfov);
   4.110 +
   4.111 +	pz *= 1000.0;
   4.112 +
   4.113 +	Ray ray = {{0, 0, 0, 1}, {px, py, pz, 1}};
   4.114 +	return ray;
   4.115 +}

     5.1 --- /dev/null	Thu Jan 01 00:00:00 1970 +0000
     5.2 +++ b/src/rt.h	Tue Jul 13 03:38:29 2010 +0300
     5.3 @@ -0,0 +1,8 @@
     5.4 +#ifndef RT_H_
     5.5 +#define RT_H_
     5.6 +
     5.7 +bool init_renderer(int xsz, int ysz, float *fb);
     5.8 +void destroy_renderer();
     5.9 +bool render();
    5.10 +
    5.11 +#endif	/* RT_H_ */