i commit 2HEADmaster

6 files changed, 244 insertions, 105 deletions

diff --git a/CMakeLists.txt b/CMakeLists.txt
index da2c461..a4ab4ee 100644
--- a/CMakeLists.txt
+++ b/CMakeLists.txt
@@ -1,12 +1,14 @@
 cmake_minimum_required(VERSION 3.13)
-project(opengltest)
+set(ENV{CUDAHOSTCXX} "g++-5")
+project(opengltest LANGUAGES C CXX CUDA)
 
 set(CMAKE_CXX_STANDARD 14)
 set(LIB_DIR "${CMAKE_CURRENT_SOURCE_DIR}/libraries")
 
 find_package(OpenGL REQUIRED)
+find_package(CUDA)
 
-add_executable(${PROJECT_NAME} main.cpp point.h Canvas.h)
+add_executable(${PROJECT_NAME} main.cpp point.h Canvas/Canvas.h)
 
 # GLFW
 find_package(PkgConfig REQUIRED)
@@ -16,10 +18,11 @@ target_link_libraries(${PROJECT_NAME} ${GLFW_STATIC_LIBRARIES})
 
 # GLAD
 set(GLAD_DIR "${LIB_DIR}/glad")
-add_library("glad" "${GLAD_DIR}/src/glad.c")
-target_include_directories("glad" PRIVATE "${GLAD_DIR}/include")
+set(GLAD_NAME "glad")
+add_library(${GLAD_NAME} "${GLAD_DIR}/src/glad.c")
+target_include_directories(${GLAD_NAME} PRIVATE "${GLAD_DIR}/include")
 target_include_directories(${PROJECT_NAME} PRIVATE "${GLAD_DIR}/include")
-target_link_libraries(${PROJECT_NAME} "glad" "${CMAKE_DL_LIBS}")
+target_link_libraries(${PROJECT_NAME} ${GLAD_NAME} "${CMAKE_DL_LIBS}")
 
 # stb_image
 target_include_directories(${PROJECT_NAME} PRIVATE "${LIB_DIR}/stb_image")
@@ -27,4 +30,10 @@ target_include_directories(${PROJECT_NAME} PRIVATE "${LIB_DIR}/stb_image")
 # shader
 target_include_directories(${PROJECT_NAME} PRIVATE "${CMAKE_CURRENT_SOURCE_DIR}/learnopengl")
 
-# Canvas
+# CUDA
+set(CUDA_LIB "cudalib")
+#target_include_directories(${PROJECT_NAME} PRIVATE "Canvas/")
+cuda_add_library(${CUDA_LIB} STATIC Canvas/cuda/CalculateCanvasGPU.cu Canvas/cuda/CalculateCanvasGPU.h OPTIONS -std=c++11)
+target_compile_features(${CUDA_LIB} PUBLIC cxx_std_11)
+set_target_properties(${CUDA_LIB} PROPERTIES CUDA_SEPARABLE_COMPILATION ON)
+target_link_libraries(${PROJECT_NAME} ${CUDA_LIB})
diff --git a/Canvas.h b/Canvas.h
deleted file mode 100644
index b2f943d..0000000
--- a/Canvas.h
+++ /dev/null
@@ -1,62 +0,0 @@
-//
-// Created by saintruler on 04.05.19.
-//
-
-#ifndef OPENGLTEST_CANVAS_H
-#define OPENGLTEST_CANVAS_H
-
-#include <glad/glad.h>
-#include <GLFW/glfw3.h>
-
-#include <iostream>
-
-struct Color
-{
-    unsigned int r = 0, g = 0, b = 0;
-};
-
-class Canvas
-{
-public:
-    Canvas(int width, int height)
-    {
-        this->width = width;
-        this->height = height;
-
-        canvas = (unsigned char*) malloc(width * height * 3);
-        glGenTextures(1, &texture);
-    }
-
-    void SetPixel(int x, int y, Color color)
-    {
-        glBindTexture(GL_TEXTURE_2D, texture); // all upcoming GL_TEXTURE_2D operations now have effect on this texture object
-
-        // set texture wrapping to GL_REPEAT (default wrapping method)
-        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_REPEAT);
-        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_REPEAT);
-
-        // set texture filtering parameters
-        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR);
-        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR);
-
-        canvas[y * width * 3 + x * 3 + 0] = color.r;
-        canvas[y * width * 3 + x * 3 + 1] = color.g;
-        canvas[y * width * 3 + x * 3 + 2] = color.b;
-
-
-        glTexImage2D(GL_TEXTURE_2D, 0, GL_RGB, width, width, 0, GL_RGB, GL_UNSIGNED_BYTE, canvas);
-        glGenerateMipmap(GL_TEXTURE_2D);
-    }
-
-    unsigned int* GetTexture()
-    {
-        return &(this->texture);
-    }
-
-private:
-    int width, height;
-    unsigned char* canvas;
-    unsigned int texture;
-};
-
-#endif //OPENGLTEST_CANVAS_H
diff --git a/Canvas/Canvas.h b/Canvas/Canvas.h
new file mode 100644
index 0000000..84dbe56
--- /dev/null
+++ b/Canvas/Canvas.h
@@ -0,0 +1,117 @@
+//
+// Created by saintruler on 04.05.19.
+//
+
+#ifndef OPENGLTEST_CANVAS_H
+#define OPENGLTEST_CANVAS_H
+
+#include <glad/glad.h>
+#include <GLFW/glfw3.h>
+
+#include <iostream>
+
+struct Color
+{
+    unsigned int r = 0, g = 0, b = 0;
+};
+
+class Canvas
+{
+public:
+    Canvas(int width, int height)
+    {
+//        vertices = {
+//            // positions        // colors           // texture coords
+//            1.f,  1.f, 0.0f,    1.0f, 0.0f, 0.0f,   1.0f, 1.0f, // top right
+//            1.f, -1.f, 0.0f,    0.0f, 1.0f, 0.0f,   1.0f, 0.0f, // bottom right
+//            -1.f, -1.f, 0.0f,   0.0f, 0.0f, 1.0f,   0.0f, 0.0f, // bottom left
+//            -1.f,  1.f, 0.0f,   1.0f, 1.0f, 0.0f,   0.0f, 1.0f  // top left
+//        };
+//
+//        indices = {
+//            0, 1, 3, // first triangle
+//            1, 2, 3  // second triangle
+//        };
+
+        glGenVertexArrays(1, &VAO);
+        glGenBuffers(1, &VBO);
+        glGenBuffers(1, &EBO);
+
+        glBindVertexArray(VAO);
+
+        glBindBuffer(GL_ARRAY_BUFFER, VBO);
+        glBufferData(GL_ARRAY_BUFFER, sizeof(vertices), vertices, GL_STATIC_DRAW);
+
+        glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, EBO);
+        glBufferData(GL_ELEMENT_ARRAY_BUFFER, sizeof(indices), indices, GL_STATIC_DRAW);
+
+        // position attribute
+        glVertexAttribPointer(0, 3, GL_FLOAT, GL_FALSE, 8 * sizeof(float), (void*)0);
+        glEnableVertexAttribArray(0);
+        // color attribute
+        glVertexAttribPointer(1, 3, GL_FLOAT, GL_FALSE, 8 * sizeof(float), (void*)(3 * sizeof(float)));
+        glEnableVertexAttribArray(1);
+        // texture coord attribute
+        glVertexAttribPointer(2, 2, GL_FLOAT, GL_FALSE, 8 * sizeof(float), (void*)(6 * sizeof(float)));
+        glEnableVertexAttribArray(2);
+
+
+        this->width = width;
+        this->height = height;
+
+        canvas = (unsigned char*) malloc(width * height * 3);
+        glGenTextures(1, &texture);
+    }
+
+    void SetPixel(int x, int y, Color color)
+    {
+        canvas[y * width * 3 + x * 3 + 0] = color.r;
+        canvas[y * width * 3 + x * 3 + 1] = color.g;
+        canvas[y * width * 3 + x * 3 + 2] = color.b;
+
+        UpdateTexture();
+    }
+
+    void UpdateTexture()
+    {
+        // all upcoming GL_TEXTURE_2D operations now have effect on this texture object
+        glBindTexture(GL_TEXTURE_2D, texture);
+
+        // set texture wrapping to GL_REPEAT (default wrapping method)
+        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_REPEAT);
+        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_REPEAT);
+
+        // set texture filtering parameters
+        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR);
+        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR);
+
+        glTexImage2D(GL_TEXTURE_2D, 0, GL_RGB, width, width, 0, GL_RGB, GL_UNSIGNED_BYTE, canvas);
+        glGenerateMipmap(GL_TEXTURE_2D);
+    }
+
+    void DrawTexture()
+    {
+        glBindTexture(GL_TEXTURE_2D, texture);
+        // render container
+//        ourShader.use();
+        glBindVertexArray(VAO);
+        glDrawElements(GL_TRIANGLES, 6, GL_UNSIGNED_INT, 0);
+    }
+
+    unsigned int* GetTexture()
+    {
+        return &(this->texture);
+    }
+
+private:
+    int width, height;
+    unsigned char* canvas;
+    unsigned int texture;
+
+    unsigned int VBO, VAO, EBO;
+
+    float* vertices;
+    unsigned int* indices;
+};
+
+#endif //OPENGLTEST_CANVAS_H
diff --git a/Canvas/cuda/CalculateCanvasGPU.cu b/Canvas/cuda/CalculateCanvasGPU.cu
new file mode 100644
index 0000000..c8c6fad
--- /dev/null
+++ b/Canvas/cuda/CalculateCanvasGPU.cu
@@ -0,0 +1,52 @@
+#include "CalculateCanvasGPU.h"
+
+__global__
+void calculate(unsigned char* canvas, int width, cuda_point* points, int nPoints)
+{
+    int canvasX = blockIdx.x * blockDim.x + threadIdx.x;
+    int canvasY = blockIdx.y * blockDim.y + threadIdx.y;
+
+    int pathX, pathY;
+    int length_squared;
+    double f = 0;
+
+    for (int i = 0; i < nPoints; i++)
+    {
+        pathX = points[i].x - canvasX;
+        pathY = points[i].y - canvasY;
+
+        length_squared = pathX * pathX + pathY * pathY;
+        f += (1e5 * 20) / (float) length_squared;
+    }
+
+    int l = (int) f;
+    if (l > 255) l = 255;
+
+    canvas[canvasY * width * 3 + canvasX * 3 + 0] = l;
+    canvas[canvasY * width * 3 + canvasX * 3 + 1] = 0;
+    canvas[canvasY * width * 3 + canvasX * 3 + 2] = 0;
+}
+
+void generate_canvas(int width, int height, unsigned char* canvas, cuda_point* points, int nPoints)
+{
+    const int canvasSize = width * height * 3 * sizeof(unsigned char);
+    const int pointsSize = nPoints * sizeof(cuda_point);
+
+    unsigned char* gpuCanvas;
+    cuda_point* gpuPoints;
+
+    cudaMalloc((void**) &gpuCanvas, canvasSize);
+    cudaMalloc((void**) &gpuPoints, pointsSize);
+
+    cudaMemcpy(gpuCanvas, canvas, canvasSize, cudaMemcpyHostToDevice);
+    cudaMemcpy(gpuPoints, points, pointsSize, cudaMemcpyHostToDevice);
+
+    dim3 threadsPerBlock(16, 16);
+    dim3 numBlocks(width / threadsPerBlock.x, height / threadsPerBlock.y);
+    calculate<<<numBlocks, threadsPerBlock>>>(gpuCanvas, width, gpuPoints, nPoints);
+
+    cudaMemcpy(canvas, gpuCanvas, canvasSize, cudaMemcpyDeviceToHost);
+
+    cudaFree(gpuCanvas);
+    cudaFree(gpuPoints);
+}
diff --git a/Canvas/cuda/CalculateCanvasGPU.h b/Canvas/cuda/CalculateCanvasGPU.h
new file mode 100644
index 0000000..46f54c5
--- /dev/null
+++ b/Canvas/cuda/CalculateCanvasGPU.h
@@ -0,0 +1,20 @@
+//
+// Created by saintruler on 05.05.19.
+//
+
+#ifndef OPENGLTEST_CALCULATECANVASGPU_H
+#define OPENGLTEST_CALCULATECANVASGPU_H
+
+#include <iostream>
+#include <vector>
+#include <fstream>
+
+typedef struct
+{
+    int x, y;
+} cuda_point;
+
+
+void generate_canvas(int width, int height, unsigned char* canvas, cuda_point* points, int nPoints);
+
+#endif //OPENGLTEST_CALCULATECANVASGPU_H
diff --git a/main.cpp b/main.cpp
index 2d4ecce..f281df8 100644
--- a/main.cpp
+++ b/main.cpp
@@ -5,8 +5,9 @@
 #include <stb_image.h>
 
 #include <shader.h>
-#include "Canvas.h"
+#include "Canvas/Canvas.h"
 #include "point.h"
+#include "Canvas/cuda/CalculateCanvasGPU.h"
 
 #include <iostream>
 #include <cmath>
@@ -17,15 +18,17 @@ void processInput(GLFWwindow *window);
 
 void loadTexture(std::string path, unsigned int* texture, int* width, int* height);
 void generateCanvasTexture(unsigned int *texture, unsigned char color);
+void window_size_callback(GLFWwindow* window, int width, int height);
 
 // settings
-const unsigned int SCR_WIDTH = 800;
-const unsigned int SCR_HEIGHT = 600;
-const std::string PROJECT_PATH = "/home/saintruler/CLionProjects/opengltest/";
+unsigned int SCR_WIDTH = 768;
+unsigned int SCR_HEIGHT = 768;
+const std::string PROJECT_PATH = "/home/saintruler/Documents/opengl_gravity/";
 unsigned char* canvas = (unsigned char*) malloc(SCR_HEIGHT * SCR_WIDTH * 3);
 
 std::vector<point> points;
 const double c = 1e5;
+cuda_point* gpuPoints = (cuda_point*) malloc(3 * sizeof(cuda_point));
 
 int main()
 {
@@ -62,13 +65,6 @@ int main()
 
     // set up vertex data (and buffer(s)) and configure vertex attributes
     // ------------------------------------------------------------------
-    float vertices1[] = {
-            // positions          // colors           // texture coords
-            0.5f,  0.5f, 0.0f,    1.0f, 0.0f, 0.0f,   1.0f, 1.0f, // top right
-            0.5f, -0.5f, 0.0f,    0.0f, 1.0f, 0.0f,   1.0f, 0.0f, // bottom right
-            -0.5f, -0.5f, 0.0f,   0.0f, 0.0f, 1.0f,   0.0f, 0.0f, // bottom left
-            -0.5f,  0.5f, 0.0f,   1.0f, 1.0f, 0.0f,   0.0f, 1.0f  // top left
-    };
 
     float vertices[] = {
             // positions        // colors           // texture coords
@@ -115,6 +111,10 @@ int main()
     points.emplace_back(point(500, 500));
     points.emplace_back(point(200, 375));
 
+    gpuPoints[0] = cuda_point{.x=100, .y=100};
+    gpuPoints[1] = cuda_point{.x=500, .y=500};
+    gpuPoints[2] = cuda_point{.x=200, .y=200};
+
 
     // render loop
     // -----------
@@ -196,29 +196,31 @@ void generateCanvasTexture(unsigned int *texture, unsigned char color)
     // load image, create texture and generate mipmaps
 
 
-    for (int x = 0; x < SCR_WIDTH; x++)
-    {
-        for (int y = 0; y < SCR_HEIGHT; y++)
-        {
-
-            point canvas_pt(x, y);
-            double f = 0;
-            for (point pt : points)
-            {
-                point path = pt - canvas_pt;
-
-                f += (c * 20 * 1) / path.length_sqared();
-            }
-
-            int l = (int) f;
-            if (l > 255) l = 255;
-
-
-            canvas[y * SCR_WIDTH * 3 + x * 3 + 0] = l;
-            canvas[y * SCR_WIDTH * 3 + x * 3 + 1] = 0;
-            canvas[y * SCR_WIDTH * 3 + x * 3 + 2] = 0;
-        }
-    }
+//    for (int x = 0; x < SCR_WIDTH; x++)
+//    {
+//        for (int y = 0; y < SCR_HEIGHT; y++)
+//        {
+//
+//            point canvas_pt(x, y);
+//            double f = 0;
+//            for (point pt : points)
+//            {
+//                point path = pt - canvas_pt;
+//
+//                f += (c * 20 * 1) / path.length_sqared();
+//            }
+//
+//            int l = (int) f;
+//            if (l > 255) l = 255;
+//
+//
+//            canvas[y * SCR_WIDTH * 3 + x * 3 + 0] = l;
+//            canvas[y * SCR_WIDTH * 3 + x * 3 + 1] = 0;
+//            canvas[y * SCR_WIDTH * 3 + x * 3 + 2] = 0;
+//        }
+//    }
+
+    generate_canvas(SCR_WIDTH, SCR_HEIGHT, canvas, gpuPoints, 3);
 
     glTexImage2D(GL_TEXTURE_2D, 0, GL_RGB, SCR_WIDTH, SCR_HEIGHT, 0, GL_RGB, GL_UNSIGNED_BYTE, (unsigned char*) canvas);
     glGenerateMipmap(GL_TEXTURE_2D);
@@ -245,8 +247,10 @@ void processInput(GLFWwindow *window)
     double xpos, ypos;
     //getting cursor position
     glfwGetCursorPos(window, &xpos, &ypos);
-    points[0].x = xpos;
-    points[0].y = SCR_HEIGHT - ypos;
+//    points[0].x = xpos;
+//    points[0].y = SCR_HEIGHT - ypos;
+    gpuPoints[0].x = xpos;
+    gpuPoints[0].y = SCR_HEIGHT - ypos;
 }
 
 // glfw: whenever the window size changed (by OS or user resize) this callback function executes
@@ -257,4 +261,3 @@ void framebuffer_size_callback(GLFWwindow* window, int width, int height)
     // height will be significantly larger than specified on retina displays.
     glViewport(0, 0, width, height);
 }
-