initial commit

.gitignore

@@ -0,0 +1 @@
+build

CMakeLists.txt

@@ -0,0 +1,8 @@
+cmake_minimum_required(VERSION 3.12)
+project(raytracing-in-a-weekend)
+
+set(CMAKE_CXX_STANDARD 14)
+enable_language(CUDA)
+
+add_executable(main src/main.cu)
+set_target_properties(main PROPERTIES CUDA_ARCHITECTURES all-major)

run.sh

@@ -0,0 +1,6 @@
+#!/bin/sh
+mkdir -p build
+cd build
+cmake ..
+make
+./main

src/main.cu

@@ -0,0 +1,82 @@
+#include <cassert>
+#include <ctime>
+#include <iostream>
+#include <time.h>
+
+#define checkCudaErrors(val) check_cuda((val), #val, __FILE__, __LINE__)
+
+void check_cuda(cudaError_t result, const char *func, const char *file, int line) {
+	if (result) {
+                std::cout << "CUDA error: " << cudaGetErrorString(result) << " (error code " << static_cast<unsigned int>(result) << ")";
+		std::cout << " at " << file << ":" << line << " '" << func << "' \n";
+		cudaDeviceReset();
+		exit(-1);
+	}
+}
+
+__global__ void render(float *fb, int max_x, int max_y) {
+    int x = threadIdx.x + blockIdx.x * blockDim.x;
+    int y = threadIdx.y + blockIdx.y * blockDim.y;
+    if ((x >= max_x) || (y >= max_y)) return;
+    int pixel_idx = 3*(y*max_x + x);
+
+    fb[pixel_idx + 0] = float(x) / max_x;
+    fb[pixel_idx + 1] = float(y) / max_y;
+    fb[pixel_idx + 2] = 0.2;
+}
+
+int main() {
+	const char *image_filename = "image.ppm";
+	int nx = 1200;
+	int ny = 600;
+	int tx = 8;
+	int ty = 8;
+
+	int num_pixels = nx*ny;
+	size_t fb_size = 3*num_pixels*sizeof(float);
+
+	float *fb;
+	checkCudaErrors(cudaMallocManaged(&fb, fb_size));
+
+	clock_t start = clock();
+	{
+		dim3 blocks(nx/tx+1, ny/ty+1);
+		dim3 threads(tx, ty);
+
+		render<<<blocks, threads>>>(fb, nx, ny);
+                checkCudaErrors(cudaGetLastError());
+                checkCudaErrors(cudaDeviceSynchronize());
+	}
+	clock_t stop = clock();
+	double timer_seconds = ((double)(stop - start)) / CLOCKS_PER_SEC;
+	std::cout << "took " << timer_seconds << " seconds.\n";
+
+	FILE *f = fopen(image_filename, "w");
+	assert(f);
+
+	char header[128] = { 0 };
+	int header_size = snprintf(header, sizeof(header), "P3\n%d %d\n255\n", nx, ny);
+	fwrite(header, header_size, 1, f);
+
+	for (int y=ny-1; y >= 0; y--) {
+		for (int x=0; x < nx; x++) {
+			size_t pixel_idx = 3*(y*nx + x);
+			float r = fb[pixel_idx + 0];
+			float g = fb[pixel_idx + 1];
+			float b = fb[pixel_idx + 2];
+			int ir = int(255.99*r);
+			int ig = int(255.99*g);
+			int ib = int(255.99*b);
+
+			char pixel_str[128] = { 0 };
+			int str_size = snprintf(pixel_str, sizeof(pixel_str), "%d %d %d\n", ir, ig, ib);
+			fwrite(pixel_str, str_size, 1, f);
+		}
+	}
+
+	fclose(f);
+
+	checkCudaErrors(cudaFree(fb));
+
+	return 0;
+}