diff --git a/.gitignore b/.gitignore index 1c3a1df..48c406d 100644 --- a/.gitignore +++ b/.gitignore @@ -3,6 +3,7 @@ # Build build/ +build_*/ *.o # Data diff --git a/.gitmodules b/.gitmodules new file mode 100644 index 0000000..76e461a --- /dev/null +++ b/.gitmodules @@ -0,0 +1,3 @@ +[submodule "include/imgui"] + path = include/imgui + url = https://github.com/ocornut/imgui.git diff --git a/CMakeLists.txt b/CMakeLists.txt index cacca9f..79cf298 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -14,11 +14,19 @@ file(GLOB_RECURSE SOURCE_FILES ${CMAKE_SOURCE_DIR}/src/*.cu) # header files -file(GLOB_RECURSE HEADER_FILES +file(GLOB HEADER_FILES ${CMAKE_SOURCE_DIR}/src/*.h ${CMAKE_SOURCE_DIR}/src/*.hpp) -add_executable(${PROJECT_NAME} ${HEADER_FILES} ${SOURCE_FILES}) +# imgui - note: could be done as a library as well but this was easier for now +file(GLOB IMGUI_FILES + ${CMAKE_SOURCE_DIR}/include/imgui/*.h + ${CMAKE_SOURCE_DIR}/include/imgui/*.cpp + ${CMAKE_SOURCE_DIR}/include/imgui/backends/imgui_impl_glfw.* + ${CMAKE_SOURCE_DIR}/include/imgui/backends/imgui_impl_opengl3.*) +include_directories(${CMAKE_SOURCE_DIR}/include/imgui) + +add_executable(${PROJECT_NAME} ${HEADER_FILES} ${SOURCE_FILES} ${IMGUI_FILES}) # CUDA has specific architectures - set it to the system's architecture if available (or 70 by default) set_target_properties(${PROJECT_NAME} PROPERTIES CUDA_ARCHITECTURES 70) diff --git a/include/imgui b/include/imgui new file mode 160000 index 0000000..0b8ff4b --- /dev/null +++ b/include/imgui @@ -0,0 +1 @@ +Subproject commit 0b8ff4b2382d4bbd5c168b1a5373f86ea3145957 diff --git a/src/consts.cu b/src/consts.cu index b432f6d..fd5f731 100644 --- a/src/consts.cu +++ b/src/consts.cu @@ -4,15 +4,3 @@ __device__ Point3 d_cameraPos; __device__ Vec3 d_cameraDir; __device__ Vec3 d_cameraUp; __device__ Point3 d_lightPos; - -Point3 h_cameraPos = Point3::init(-0.7, -1.0, -2.0); -Vec3 h_cameraDir = Vec3::init(0.4, 0.6, 1.0).normalize(); -Vec3 h_cameraUp = Vec3::init(0.0, 1.0, 0.0).normalize(); -Point3 h_lightPos = Point3::init(1.5, 2.0, -1.0); - -void copyConstantsToDevice() { - cudaMemcpyToSymbol(d_cameraPos, &h_cameraPos, sizeof(Point3)); - cudaMemcpyToSymbol(d_cameraDir, &h_cameraDir, sizeof(Vec3)); - cudaMemcpyToSymbol(d_cameraUp, &h_cameraUp, sizeof(Vec3)); - cudaMemcpyToSymbol(d_lightPos, &h_lightPos, sizeof(Point3)); -} diff --git a/src/consts.h b/src/consts.h index 0c1e8d4..6685ffc 100644 --- a/src/consts.h +++ b/src/consts.h @@ -9,6 +9,7 @@ const int VOLUME_WIDTH = 49; const int VOLUME_HEIGHT = 51; const int VOLUME_DEPTH = 42; +// TODO: replace with window->w and window->h const int IMAGE_WIDTH = 800; const int IMAGE_HEIGHT = 600; @@ -39,6 +40,6 @@ extern __device__ Vec3 d_cameraUp; extern __device__ Point3 d_lightPos; // --------------------------- Functions for handling external constants --------------------------- -void copyConstantsToDevice(); +// void copyConstantsToDevice(); #endif // CONSTS_H diff --git a/src/gui/MainWindow.cpp b/src/gui/MainWindow.cpp index a487328..9c2db1f 100644 --- a/src/gui/MainWindow.cpp +++ b/src/gui/MainWindow.cpp @@ -5,6 +5,7 @@ #include #include "Shader.h" +#include "input/Widget.h" Window::Window(unsigned int w, unsigned int h) { this->w = w; @@ -28,9 +29,6 @@ int Window::init(float* data) { this->window = glfwCreateWindow(this->w, this->h, "CUDA ray tracing", NULL, NULL); - //hide cursor // TODO: switch from this style input to something more resembling an actual gui - glfwSetInputMode(this->window, GLFW_CURSOR, GLFW_CURSOR_DISABLED); - glfwSetWindowUserPointer(this->window, reinterpret_cast(this)); if (this->window == NULL) { std::cout << "Failed to create window\n"; @@ -53,6 +51,10 @@ int Window::init(float* data) { if (init_quad(data)) return -1; this->last_frame = std::chrono::steady_clock::now(); + // init imGUI + this->widget = new Widget(this->window); + + while (!glfwWindowShouldClose(window)) { Window::tick(); } @@ -79,8 +81,10 @@ void Window::free(float* data) { // To preserve the proper destruction order we forcefully set the quads to null (calling their destructor in the process) // Not strictly necessary, but i saw some weird errors on exit without this so best to keep it in. this->quad = nullptr; + this->widget = nullptr; cudaFree(data); + glfwDestroyWindow(window); glfwTerminate(); } @@ -95,23 +99,23 @@ void Window::tick() { // TODO: remove debug line at some point std::cout << 1000.0/diff << " fps\n"; - // TODO: code input logic and class/struct and stuff - // ticking input probably involves 4? steps: - // * check if window needs to be closed (escape/q pressed) - // * check if camera moved (wasd/hjkl pressed) - // (phase 3/do later): check if we switched from realtime tracing to that other option - maybe a pause function? (p pressed?) - // * if moved -> update camera (raytracing will involve some logic here too? see when i get there) + // input + this->widget->tick(); + // tick render glBindFramebuffer(GL_FRAMEBUFFER, 0); glDisable(GL_DEPTH_TEST); - this->quad->render(); + if (!this->widget->paused) this->quad->render(); this->shader->use(); glBindVertexArray(this->quad->VAO); glBindTexture(GL_TEXTURE_2D, this->quad->tex); glDrawArrays(GL_TRIANGLES, 0, 6); // draw current frame to texture + // render ImGui context + this->widget->render(); + // check for events glfwSwapBuffers(this->window); glfwPollEvents(); diff --git a/src/gui/MainWindow.h b/src/gui/MainWindow.h index 96f490a..f4ed0f1 100644 --- a/src/gui/MainWindow.h +++ b/src/gui/MainWindow.h @@ -6,6 +6,7 @@ #include #include #include +#include "input/Widget.h" class Window { @@ -23,6 +24,7 @@ public: private: GLFWwindow* window; std::unique_ptr quad; + Widget* widget; std::chrono::steady_clock::time_point last_frame; diff --git a/src/gui/input/Widget.cpp b/src/gui/input/Widget.cpp new file mode 100644 index 0000000..7f3a87d --- /dev/null +++ b/src/gui/input/Widget.cpp @@ -0,0 +1,77 @@ +#include "Widget.h" +#include "linalg/linalg.h" +#include "consts.h" + + +Widget::Widget(GLFWwindow* window) { + IMGUI_CHECKVERSION(); + ImGui::CreateContext(); + this->io = ImGui::GetIO(); + + ImGui_ImplGlfw_InitForOpenGL(window, true); + ImGui_ImplOpenGL3_Init(); + + this->cameraDir = Vec3::init(0.4, 0.6, 1.0).normalize(); + this->cameraPos = Point3::init(-0.7, -1.0, -2.0); + this->cameraUp = Vec3::init(0.0, 1.0, 0.0).normalize(); + this->lightPos = Point3::init(1.5, 2.0, -1.0); + this->paused = true; + this->renderOnce = false; +}; + +void Widget::tick() { + if (this->renderOnce) { + this->renderOnce = false; + this->paused = true; + } + + ImGui_ImplOpenGL3_NewFrame(); + ImGui_ImplGlfw_NewFrame(); + ImGui::NewFrame(); + + // input widgets + float min = -1, max = 1; + + ImGui::Begin("Light Controls"); + ImGui::DragScalar("X coordinate", ImGuiDataType_Double, &this->lightPos.x, 0.005f, &min, &max, "%.3f"); + ImGui::DragScalar("Y coordinate", ImGuiDataType_Double, &this->lightPos.y, 0.005f, &min, &max, "%.3f"); + ImGui::DragScalar("Z coordinate", ImGuiDataType_Double, &this->lightPos.z, 0.005f, &min, &max, "%.3f"); + ImGui::End(); + + ImGui::Begin("Pause"); + if (ImGui::Button(this->paused ? "Unpause" : "Pause")) this->paused = !this->paused; + if (ImGui::Button("render Once")) { + this->paused = !this->paused; + this->renderOnce = true; + } + ImGui::End(); + + ImGui::Begin("Camera Controls"); + ImGui::DragScalar("X position", ImGuiDataType_Double, &this->cameraPos.x, 0.005f, &min, &max, "%.3f"); + ImGui::DragScalar("Y position", ImGuiDataType_Double, &this->cameraPos.y, 0.005f, &min, &max, "%.3f"); + ImGui::DragScalar("Z position", ImGuiDataType_Double, &this->cameraPos.z, 0.005f, &min, &max, "%.3f"); + ImGui::DragScalar("X direction", ImGuiDataType_Double, &this->cameraDir.x, 0.005f, &min, &max, "%.3f"); + ImGui::DragScalar("Y direction", ImGuiDataType_Double, &this->cameraDir.y, 0.005f, &min, &max, "%.3f"); + ImGui::DragScalar("Z direction", ImGuiDataType_Double, &this->cameraDir.z, 0.005f, &min, &max, "%.3f"); + ImGui::End(); + + copyToDevice(); +} + +void Widget::render() { + ImGui::Render(); + ImGui_ImplOpenGL3_RenderDrawData(ImGui::GetDrawData()); +} + +void Widget::copyToDevice() { + cudaMemcpyToSymbol(&d_cameraPos, &this->cameraPos, sizeof(Point3)); + cudaMemcpyToSymbol(&d_cameraDir, &this->cameraDir, sizeof(Vec3)); + cudaMemcpyToSymbol(&d_cameraUp, &this->cameraUp, sizeof(Vec3)); + cudaMemcpyToSymbol(&d_lightPos, &this->lightPos, sizeof(Point3)); +} + +Widget::~Widget() { + ImGui_ImplOpenGL3_Shutdown(); + ImGui_ImplGlfw_Shutdown(); + ImGui::DestroyContext(); +} diff --git a/src/gui/input/Widget.h b/src/gui/input/Widget.h new file mode 100644 index 0000000..9eb243d --- /dev/null +++ b/src/gui/input/Widget.h @@ -0,0 +1,31 @@ +#ifndef WIDGET_H +#define WIDGET_H + +#include "../include/imgui/imgui.h" +#include "../include/imgui/backends/imgui_impl_glfw.h" +#include "../include/imgui/backends/imgui_impl_opengl3.h" +#include +#include "linalg/linalg.h" + +class Widget { +public: + Point3 cameraDir; + Vec3 cameraPos; + Vec3 cameraUp; + Point3 lightPos; + bool paused; + bool renderOnce; + + ImGuiIO io; + + void tick(); + void render(); + void copyToDevice(); + + Widget(GLFWwindow* window); + ~Widget(); +}; + + + +#endif // WIDGET_H diff --git a/src/gui/input/rendererinfo.h b/src/gui/input/rendererinfo.h new file mode 100644 index 0000000..7c46315 --- /dev/null +++ b/src/gui/input/rendererinfo.h @@ -0,0 +1,14 @@ +#ifndef RENDERERINFO_H +#define RENDERERINFO_H + +class rendererInfo { +public: + Vec3 cameraDir; + Vec3 cameraPos; + Vec3 lightPos; + + RendererInfo(); + void copyToDevice(); +}; + +#endif // RENDERERINFO_H diff --git a/src/gui/shaders/fragshader.glsl b/src/gui/shaders/fragshader.glsl index 23016a3..63dfaf7 100644 --- a/src/gui/shaders/fragshader.glsl +++ b/src/gui/shaders/fragshader.glsl @@ -9,4 +9,4 @@ void main() { vec3 col = texture(screenTexture, TexCoords).rgb; FragColor = vec4(col, 1.0); -} \ No newline at end of file +} diff --git a/src/illumination/Raycaster.cu b/src/illumination/Raycaster.cu index a405e25..d530f0c 100644 --- a/src/illumination/Raycaster.cu +++ b/src/illumination/Raycaster.cu @@ -10,7 +10,6 @@ #include #include "objs/sphere.h" - // TODO: instead of IMAGEWIDTH and IMAGEHEIGHT this should reflect the windowSize; __global__ void raycastKernel(float* volumeData, FrameBuffer framebuffer) { int px = blockIdx.x * blockDim.x + threadIdx.x; @@ -148,9 +147,9 @@ void Raycaster::render() { check_cuda_errors(cudaGraphicsMapResources(1, &this->resources)); check_cuda_errors(cudaGraphicsResourceGetMappedPointer((void**)&(this->fb->buffer), &(this->fb->buffer_size), resources)); - // FIXME: might not be the best parallelization configuraiton - int tx = 16; - int ty = 16; + // FIXME: might not be the best parallelization configuration + int tx = 8; + int ty = 8; dim3 threadSize(this->w / tx + 1, this->h / ty + 1); dim3 blockSize(tx, ty); diff --git a/src/illumination/Raycaster.h b/src/illumination/Raycaster.h index 44ce2a6..3327066 100644 --- a/src/illumination/Raycaster.h +++ b/src/illumination/Raycaster.h @@ -1,7 +1,6 @@ #ifndef RAYCASTER_H #define RAYCASTER_H -// #include "Camera.h" #include "cuda_runtime.h" #include "FrameBuffer.h" #include "linalg/linalg.h" @@ -12,9 +11,6 @@ __global__ void raycastKernel(float* volumeData, unsigned char* framebuffer); struct Raycaster { - // thrust::device_ptr d_camera; - // CameraInfo camera_info; - cudaGraphicsResource_t resources; FrameBuffer* fb; float* data; diff --git a/src/main.cu b/src/main.cu index 10a3dfc..a8ec392 100644 --- a/src/main.cu +++ b/src/main.cu @@ -14,10 +14,15 @@ static float* d_volume = nullptr; +// FIXME: segfaults on window resize - the raycasting function should work with window->w and window-h instead of constants. + // TODO: general -// * pass camera_info to the raycasting function - updated according to glfw. -// * on that note, code for handling input (mouse movement certainly, possibly free input / 4 pre-coded views, q/esc to quit, space for pause (would be were the 'simple' render idea would come in)) -// * very similarly - actual code for loading new data as the simulation progresses - right now its effectively a static image loader +// * very similarly - actual code for loading new data as the simulation progresses - right now its effectively a static image loader * pause button once that dataloading is implemented + +// * save frames to file while running program -> then export to gif on close. +// * time controls - arbitrary skipping to specified point (would require some changes to gpubuffer) (could have) + +// * transfer function -> move the code in raycastkernel to its own class and add silhouette detection here as well. void getTemperature(std::vector& temperatureData, int idx = 0) { std::string path = "data/trimmed"; @@ -78,46 +83,12 @@ int main() { cudaMalloc((void**)&d_volume, volumeSize); cudaMemcpy(d_volume, hostVolume, volumeSize, cudaMemcpyHostToDevice); - // Allocate framebuffer - // unsigned char* d_framebuffer; - // size_t fbSize = IMAGE_WIDTH * IMAGE_HEIGHT * 3 * sizeof(unsigned char); - // cudaMalloc((void**)&d_framebuffer, fbSize); - // cudaMemset(d_framebuffer, 0, fbSize); - - // Copy external constants from consts.h to cuda - copyConstantsToDevice(); - - // NOTE: this is done within the rayTracer class - // // Launch kernel - // dim3 blockSize(16, 16); - // dim3 gridSize((IMAGE_WIDTH + blockSize.x - 1)/blockSize.x, - // (IMAGE_HEIGHT + blockSize.y - 1)/blockSize.y); - // - // raycastKernel<<>>( - // d_volume, - // d_framebuffer - // ); - // cudaDeviceSynchronize(); - + // Create the GUI Window window(IMAGE_WIDTH, IMAGE_HEIGHT); int out = window.init(d_volume); + // memory management cudaFree(d_volume); - - // // Copy framebuffer back to CPU - // unsigned char* hostFramebuffer = new unsigned char[IMAGE_WIDTH * IMAGE_HEIGHT * 3]; - // cudaMemcpy(hostFramebuffer, d_framebuffer, fbSize, cudaMemcpyDeviceToHost); - // - // // Export image - // saveImage("output.ppm", hostFramebuffer, IMAGE_WIDTH, IMAGE_HEIGHT); - // - // // Cleanup //TODO: cleanup properly delete[] hostVolume; - // delete[] hostFramebuffer; - // cudaFree(d_volume); - // cudaFree(d_framebuffer); - // - // std::cout << "Phong-DVR rendering done. Image saved to output.ppm" << std::endl; - // return 0; return out; }