Add Cuda OpenGL interop

SPH-CUDA/main.cu

@@ -11,6 +11,8 @@
 #include "integrators.cuh"
 
 #include "visualizer.h"
+#include "particlePositionCopy.cu"
+#include <cuda_gl_interop.h>
 
 void initializeParticles(std::vector<Particle>&, Parameters&);
 
@@ -34,7 +36,6 @@ int main() {
 	size_t bytes_struct = sizeof(Particle) * params.particle_num;
 	size_t bytes_particle_list = sizeof(int) * params.particle_num;
 	size_t bytes_cell_list = sizeof(int) * params.cell_num;
-
 	checkError(cudaMalloc((void**)&d_particle_list, bytes_particle_list));
 	checkError(cudaMalloc((void**)&d_cell_list, bytes_cell_list));
 	checkError(cudaMalloc((void**)&d_particles, bytes_struct));
@@ -47,10 +48,11 @@ int main() {
 	checkError(cudaMemcpy(d_cell_list, cell_list.data(), bytes_cell_list, cudaMemcpyHostToDevice));
 
 	/* Visualization init */
-	float* translations = new float[params.particle_num * 3];
-	Visualizer vis(params.particle_radius, params.min_box_bound.x, params.min_box_bound.y, params.min_box_bound.z,
+	Visualizer vis(params.particle_num, params.particle_radius, params.min_box_bound.x, params.min_box_bound.y, params.min_box_bound.z,
 		params.max_box_bound.x, params.max_box_bound.y, params.max_box_bound.z);
 
+	struct cudaGraphicsResource* positionsVBO_CUDA = NULL;
+	checkError(cudaGraphicsGLRegisterBuffer(&positionsVBO_CUDA, vis.vertexArray, cudaGraphicsMapFlagsWriteDiscard));
 
 	std::cout << "Simulation started" << std::endl;
 	while (!glfwWindowShouldClose(vis.window)) {
@@ -85,13 +87,15 @@ int main() {
 		checkError(cudaDeviceSynchronize());
 
 		/* Visualization update */
-		checkError(cudaMemcpy(particles.data(), d_particles, bytes_struct, cudaMemcpyDeviceToHost));
-		for (int i = 0; i < particles.size(); i++) {
-			translations[i*3 + 0] = particles[i].pos.x;
-			translations[i*3 + 1] = particles[i].pos.y;
-			translations[i*3 + 2] = particles[i].pos.z;
-		}
-		vis.draw(translations, params.particle_num);
+		void* vertexPointer = nullptr;
+		// Map the buffer to CUDA
+		cudaGLMapBufferObject((void**)vertexPointer, vis.vertexArray);
+		// Run kernel
+		copy_particle_positions<<<params.thread_groups_part, params.threads_per_group>>>((float*)vertexPointer, d_particles, params.particle_num);
+		// Unmap the buffer
+		cudaGLUnmapBufferObject(vis.vertexArray);
+
+		vis.draw(params.particle_num);
 	}
 
 	std::cout << "Simulation finished" << std::endl;

SPH-CUDA/particlePositionCopy.cu

@@ -0,0 +1,14 @@
+#include "cuda_runtime.h"
+#include "device_launch_parameters.h"
+
+#include "helper_structs.h"
+#include "helper_math.h"
+
+__global__ void copy_particle_positions(float* translations, Particle* particles, int N) {
+	int tid = (blockIdx.x * blockDim.x) + threadIdx.x;
+	if (tid < N) {
+		translations[tid*3 + 0] = particles[tid].pos.x;
+        translations[tid*3 + 1] = particles[tid].pos.y;
+        translations[tid*3 + 2] = particles[tid].pos.z;
+	}
+}

SPH-CUDA/visualizer.cpp

@@ -173,7 +173,7 @@ bool inFOV(float x, float y) {
     return abs(getAngle2(-denormalizeRadians(glm::radians(camera.Yaw-90)), x-getCamPos().x, y-getCamPos().y)) <= (getFOVrad()/2.0 + glm::radians(10.0));
 }
 
-Visualizer::Visualizer(float radius, float minBoundX, float minBoundY, float minBoundZ, float maxBoundX, float maxBoundY, float maxBoundZ) {
+Visualizer::Visualizer(int objectNum, float radius, float minBoundX, float minBoundY, float minBoundZ, float maxBoundX, float maxBoundY, float maxBoundZ) {
     // DEBUG
     #ifdef DEBUG
         glfwSetErrorCallback(error_callback);
@@ -276,6 +276,7 @@ Visualizer::Visualizer(float radius, float minBoundX, float minBoundY, float min
 
     // version Info
     std::cout << glGetString(GL_VERSION) << std::endl;
+    std::cout << glGetString(GL_VENDOR) << "\n" << glGetString(GL_RENDERER) << std::endl;
     consoleMessage();
 
     // text init
@@ -285,6 +286,10 @@ Visualizer::Visualizer(float radius, float minBoundX, float minBoundY, float min
 		exit(EXIT_FAILURE);
 	}
 
+    glGenBuffers(1, &vertexArray);
+    glBindBuffer(GL_ARRAY_BUFFER, vertexArray);
+    glBufferData(GL_ARRAY_BUFFER, objectNum*3*sizeof(float), NULL, GL_DYNAMIC_COPY);
+
     renderer = new Renderer();
     // Shader stuff
     shader = new Shader("src/Shaders/shader");
@@ -294,7 +299,7 @@ Visualizer::Visualizer(float radius, float minBoundX, float minBoundY, float min
         glm::vec3(maxBoundX - radius, maxBoundY + radius, maxBoundZ - radius));
 }
 
-void Visualizer::draw(float* translations, int objectNum) {
+void Visualizer::draw(int objectNum) {
     // Loop until the user closes the window 
     //while (!glfwWindowShouldClose(window)) {
     float currentFrame = glfwGetTime();
@@ -351,18 +356,15 @@ void Visualizer::draw(float* translations, int objectNum) {
 
     // draw spheres
     sphere->bind();
-    VertexBuffer *instanceVBO = new VertexBuffer(translations, objectNum*3*sizeof(translations[0]));
+    glBindBuffer(GL_ARRAY_BUFFER, vertexArray);
     glEnableVertexAttribArray(3);
-    instanceVBO->bind();
     glVertexAttribPointer(3, 3, GL_FLOAT, GL_FALSE, 3 * sizeof(float), (void*)0);
     glVertexAttribDivisor(3, 1);  
 
     IndexBuffer* ibo = sphere->ibo;
     ibo->bind();
     glDrawElementsInstanced(GL_TRIANGLES, ibo->getCount(), GL_UNSIGNED_INT, (void*)0, objectNum);
 
-    delete instanceVBO;
-
     // draw box
     shader->setFloat("alpha", 0.2f);
     box->bind();

SPH-CUDA/visualizer.h

@@ -35,9 +35,11 @@
 
 class Visualizer {
 public:
-    Visualizer(float radius, float minBoundX, float minBoundY, float minBoundZ, float maxBoundX, float maxBoundY, float maxBoundZ);
-    void draw(float* translations, int objectNum);
+    Visualizer(int objectNum,float radius, float minBoundX, float minBoundY, float minBoundZ, float maxBoundX, float maxBoundY, float maxBoundZ);
+    void draw(int objectNum);
     void end();
+
+    GLuint vertexArray;
 
     GLFWwindow* window;