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 36a2a59..7ed979a 100644 --- a/src/consts.cu +++ b/src/consts.cu @@ -32,13 +32,7 @@ __device__ Vec3 d_cameraUp; __device__ Point3 d_lightPos; __device__ Color3 d_backgroundColor; -// Point3 h_cameraPos = Point3::init(300.0f, 200.0f, -700.0f); // Camera for full data set -Point3 h_cameraPos = Point3::init(50.0f, -50.0f, -75.0f); // Camera for partially trimmed data set -Vec3 center = Vec3::init((float)VOLUME_WIDTH/2.0f, (float)VOLUME_HEIGHT/2.0f, (float)VOLUME_DEPTH/2.0f); -Vec3 h_cameraDir = (center - h_cameraPos).normalize(); Vec3 h_cameraUp = Vec3::init(0.0, 1.0, 0.0).normalize(); -Point3 h_lightPos = Point3::init(1.5, 2.0, -1.0); -Color3 h_backgroundColor = Color3::init(0.1f, 0.1f, 0.1f); // Copy the above values to the device @@ -48,11 +42,6 @@ void copyConstantsToDevice() { cudaMemcpyToSymbol(d_stopsGrayscale, h_stopsGrayscale, sizeof(h_stopsGrayscale)); cudaMemcpyToSymbol(d_stopsBluePurleRed, h_stopsBluePurleRed, sizeof(h_stopsBluePurleRed)); - // ----------------------- Camera and Light ----------------------- - 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)); - cudaMemcpyToSymbol(d_backgroundColor, &h_backgroundColor, sizeof(Color3)); } diff --git a/src/consts.h b/src/consts.h index 2686450..70d6ccb 100644 --- a/src/consts.h +++ b/src/consts.h @@ -11,8 +11,9 @@ const int VOLUME_WIDTH = 97; // lon const int VOLUME_HEIGHT = 71; // lat const int VOLUME_DEPTH = 42; // lev -const int IMAGE_WIDTH = 1600; -const int IMAGE_HEIGHT = 1200; +// TODO: replace with window->w and window->h +const int IMAGE_WIDTH = 800; +const int IMAGE_HEIGHT = 600; const double epsilon = 1e-10f; const double infty = 1e15f; // This value is used to represent missing values in data diff --git a/src/gui/MainWindow.cpp b/src/gui/MainWindow.cpp index f7700a9..eaa8b07 100644 --- a/src/gui/MainWindow.cpp +++ b/src/gui/MainWindow.cpp @@ -5,6 +5,7 @@ #include #include "Shader.h" +#include "input/Widget.h" // TODO: Delete @@ -39,9 +40,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"; @@ -64,6 +62,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(); } @@ -95,9 +97,10 @@ int Window::init_quad(float* data) { void Window::free(float* data) { - // To preserve the proper destruction order we forcefully set the quads to null (calling their destructor in the process) + // To preserve the proper destruction order we forcefully set pointers 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); @@ -111,26 +114,22 @@ void Window::tick() { float diff = (float) std::chrono::duration_cast(now - this->last_frame).count(); this->last_frame = now; - // 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(1000.0/diff); // 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..8684db3 --- /dev/null +++ b/src/gui/input/Widget.cpp @@ -0,0 +1,89 @@ +#include "Widget.h" +#include "linalg/linalg.h" +#include "consts.h" +#include +#include + + +Widget::Widget(GLFWwindow* window) { + IMGUI_CHECKVERSION(); + ImGui::CreateContext(); + this->io = ImGui::GetIO(); + + ImGui_ImplGlfw_InitForOpenGL(window, true); + ImGui_ImplOpenGL3_Init(); + + // this->cameraPos = Point3::init(50.0f, -50.0f, -75.0f); // Camera for partially trimmed data set + this->cameraPos = Point3::init(300.0f, 200.0f, -700.0f); // Camera for full data set + // + Vec3 h_center = Vec3::init((float)VOLUME_WIDTH/2.0f, (float)VOLUME_HEIGHT/2.0f, (float)VOLUME_DEPTH/2.0f); + this->cameraDir = (h_center - this->cameraPos).normalize(); + + this->bgColor = Color3::init(0.1f, 0.1f, 0.1f); + this->lightPos = Point3::init(1.5, 2.0, -1.0); + + this->fps = (char*)malloc(512*sizeof(char)); + this->paused = true; + this->renderOnce = false; +}; + +void Widget::tick(double fps) { + 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("Miscellaneous"); + if (ImGui::Button(this->paused ? "Unpause" : "Pause")) this->paused = !this->paused; + ImGui::SameLine(); + if (ImGui::Button("render Once")) { + this->paused = !this->paused; + this->renderOnce = true; + } + sprintf(this->fps, "%.3f fps\n", fps); + ImGui::Text(this->fps); + 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_lightPos, &this->lightPos, sizeof(Point3)); + cudaMemcpyToSymbol(&d_backgroundColor, &this->bgColor, sizeof(Color3)); +} + +Widget::~Widget() { + ImGui_ImplOpenGL3_Shutdown(); + ImGui_ImplGlfw_Shutdown(); + ImGui::DestroyContext(); + free(this->fps); +} diff --git a/src/gui/input/Widget.h b/src/gui/input/Widget.h new file mode 100644 index 0000000..2fb7813 --- /dev/null +++ b/src/gui/input/Widget.h @@ -0,0 +1,33 @@ +#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; + Point3 lightPos; + Color3 bgColor; + + bool paused; + bool renderOnce; + char* fps; + + ImGuiIO io; + + void tick(double fps); + void render(); + void copyToDevice(); + + Widget(GLFWwindow* window); + ~Widget(); +}; + + + +#endif // WIDGET_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 b2d1939..58837ad 100644 --- a/src/illumination/Raycaster.cu +++ b/src/illumination/Raycaster.cu @@ -11,7 +11,6 @@ #include #include - // 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; @@ -160,9 +159,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 0ff59bb..d74f264 100644 --- a/src/main.cu +++ b/src/main.cu @@ -15,10 +15,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"; @@ -109,46 +114,13 @@ 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; }