diff options
| -rw-r--r-- | CMakeLists.txt | 21 | ||||
| -rw-r--r-- | Canvas.h | 62 | ||||
| -rw-r--r-- | Canvas/Canvas.h | 117 | ||||
| -rw-r--r-- | Canvas/cuda/CalculateCanvasGPU.cu | 52 | ||||
| -rw-r--r-- | Canvas/cuda/CalculateCanvasGPU.h | 20 | ||||
| -rw-r--r-- | main.cpp | 77 |
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 @@ -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); } - |