Merge branch 'finding-sany2' into gui

.gitignore

@@ -11,3 +11,4 @@ data/
 
 # Resulting images
 *.ppm
+

CMakeLists.txt

@@ -29,7 +29,7 @@ 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)
+set_target_properties(${PROJECT_NAME} PROPERTIES CUDA_ARCHITECTURES 52)
 set_target_properties(${PROJECT_NAME}
    PROPERTIES
        CUDA_SEPARABLE_COMPILATION ON
@@ -60,7 +60,7 @@ include_directories("${CUDA_INCLUDE_DIRS}")
 
 # netcdf
 find_package(netCDF REQUIRED)
-message(STATUS "Found netcdf in ${NETCDF_INCLUDE_DIR}")
+message(STATUS "Found netcdf in ${NETCDF_LIB}")
 
 execute_process(
     COMMAND nc-config --includedir
@@ -74,9 +74,16 @@ execute_process(
     OUTPUT_STRIP_TRAILING_WHITESPACE
 )
 
+execute_process(
+    COMMAND ncxx4-config --prefix
+    OUTPUT_VARIABLE NETCDFCXX_PREFIX_DIR
+    OUTPUT_STRIP_TRAILING_WHITESPACE
+)
+
 target_include_directories(cuda-raytracer PUBLIC ${netCDF_INCLUDE_DIR})
 
-find_library(NETCDF_LIB NAMES netcdf-cxx4 netcdf_c++4 PATHS ${NETCDFCXX_LIB_DIR} NO_DEFAULT_PATH)
+find_library(NETCDF_LIB NAMES netcdf-cxx4 netcdf_c++4 PATHS ${NETCDFCXX_LIB_DIR} ${NETCDFCXX_PREFIX_DIR}/lib NO_DEFAULT_PATH)
+message(STATUS "Found NetCDFCXX in ${NETCDF_LIB}")
 
 set(LIBS ${GLFW3_LIBRARY} ${OPENGL_LIBRARY} GLAD ${CMAKE_DL_LIBS} ${CUDA_LIBRARIES} ${NETCDF_LIB})
 

default.nix

@@ -0,0 +1,21 @@
+with import <nixpkgs> { config.allowUnfree = true; };
+
+stdenv.mkDerivation {
+	name = "cuda-raytracer";
+	src = ./.;
+	nativeBuildInputs = with cudaPackages; [cmake autoAddDriverRunpath autoPatchelfHook ];
+	buildInputs = with pkgs; with cudaPackages; [ libGL glfw netcdf netcdfcxx4 cuda_nvcc cuda_cudart cuda_cccl libcurand libdecor];
+
+	postConfigure = ''
+		export netCDFCxx_DIR=${netcdfcxx4}/lib/cmake/netCDFCxx
+	'';
+
+	installTargets = "preinstall";
+
+	postInstall = ''
+	    mkdir -p $out/bin
+     	cp cuda-raytracer $out/bin
+	'';
+
+    LD_LIBRARY_PATH="/run/opengl-driver/lib/";
+}

src/consts.cu

@@ -1,16 +1,41 @@
 #include "consts.h"
+#include "cuda_error.h"
 
 // ----------------------- Colour mapping -----------------------
-__constant__ ColorStop d_stopsPythonLike[5];
+// __device__ __constant__ ColorStop d_stopsPythonLike[5];
-__constant__ ColorStop d_stopsGrayscale[2];
+__device__ __constant__ ColorStop d_stopsPythonLike[11];
-__constant__ ColorStop d_stopsBluePurleRed[3];
+__device__ __constant__ ColorStop d_stopsGrayscale[2];
+__device__ __constant__ ColorStop d_stopsBluePurleRed[3];
 
+// const ColorStop h_stopsPythonLike[] = {
+//         { 0.0f, Color3::init(0.2298057f, 0.29871797f, 0.75368315f) }, // Dark Blue
+//         { 0.25f, Color3::init(0.23437708f, 0.30554173f, 0.75967953f) }, // Mid Blue
+//         { 0.5f, Color3::init(0.27582712f, 0.36671692f, 0.81255294f) }, // White
+//         { 0.75f, Color3::init(0.79606387f, 0.84869321f, 0.93347147f) }, // Light Orange
+//         { 1.0f, Color3::init(0.70567316f, 0.01555616f, 0.15023281f) }  // Red
+// };
+
+// const ColorStop h_stopsPythonLike[] = {
+//         { 0.0f, Color3::init(0.2298057f, 0.29871797f, 0.75368315f) }, // Dark Blue
+//         { 0.85f, Color3::init(0.23437708f, 0.30554173f, 0.75967953f) }, // Mid Blue
+//         { 0.90f, Color3::init(0.27582712f, 0.36671692f, 0.81255294f) }, // White
+//         { 0.95f, Color3::init(0.79606387f, 0.84869321f, 0.93347147f) }, // Light Orange
+//         { 1.0f, Color3::init(0.70567316f, 0.01555616f, 0.15023281f) }  // Red
+// };
+
+// Python "jet" colour scheme
 const ColorStop h_stopsPythonLike[] = {
-        { 0.0f, Color3::init(0.2298057f, 0.29871797f, 0.75368315f) }, // Dark Blue
+    { 0.00f, Color3::init(0.00000000f, 0.00000000f, 0.50000000f) },
-        { 0.25f, Color3::init(0.23437708f, 0.30554173f, 0.75967953f) }, // Mid Blue
+    { 0.82f, Color3::init(0.00000000f, 0.00000000f, 0.94563280f) },
-        { 0.5f, Color3::init(0.27582712f, 0.36671692f, 0.81255294f) }, // White
+    { 0.84f, Color3::init(0.00000000f, 0.30000000f, 1.00000000f) },
-        { 0.75f, Color3::init(0.79606387f, 0.84869321f, 0.93347147f) }, // Light Orange
+    { 0.86f, Color3::init(0.00000000f, 0.69215686f, 1.00000000f) },
-        { 1.0f, Color3::init(0.70567316f, 0.01555616f, 0.15023281f) }  // Red
+    { 0.88f, Color3::init(0.16129032f, 1.00000000f, 0.80645161f) },
+    { 0.90f, Color3::init(0.49019608f, 1.00000000f, 0.47754586f) },
+    { 0.92f, Color3::init(0.80645161f, 1.00000000f, 0.16129032f) },
+    { 0.94f, Color3::init(1.00000000f, 0.77051561f, 0.00000000f) },
+    { 0.96f, Color3::init(1.00000000f, 0.40740741f, 0.00000000f) },
+    { 0.98f, Color3::init(0.94563280f, 0.02977487f, 0.00000000f) },
+    { 1.00f, Color3::init(0.50000000f, 0.00000000f, 0.00000000f) },
 };
 
 const ColorStop h_stopsGrayscale[] = {
@@ -37,8 +62,10 @@ Vec3 h_cameraUp = Vec3::init(0.0, 1.0, 0.0).normalize();
 
 // Copy the above values to the device
 void copyConstantsToDevice() {
+    check_cuda_errors(cudaGetLastError());
     // ----------------------- Colour mapping -----------------------
     cudaMemcpyToSymbol(d_stopsPythonLike, h_stopsPythonLike, sizeof(h_stopsPythonLike));
+    check_cuda_errors(cudaGetLastError());
     cudaMemcpyToSymbol(d_stopsGrayscale, h_stopsGrayscale, sizeof(h_stopsGrayscale));
     cudaMemcpyToSymbol(d_stopsBluePurleRed, h_stopsBluePurleRed, sizeof(h_stopsBluePurleRed));
 

src/consts.h

@@ -13,13 +13,15 @@ const double infty   = 1e15f;  // This value is used to represent missing values
 
 // --------------------------- Dataset Constants ---------------------------
 // const int VOLUME_WIDTH  = 576;  // lon
-const int VOLUME_WIDTH  = 97;  // lon
+// const int VOLUME_WIDTH  = 97;  // lon
+const int VOLUME_WIDTH  = 57;  // lon
 // const int VOLUME_HEIGHT = 361;  // lat
-const int VOLUME_HEIGHT = 71;  // lat
+// const int VOLUME_HEIGHT = 71;  // lat
+const int VOLUME_HEIGHT = 121;  // lat
 const int VOLUME_DEPTH  = 42;  // lev
 
-const float DLON = 60.0f / VOLUME_WIDTH;  // 60 for current trimmed data set range
+const float DLON = 35.0f / VOLUME_WIDTH;  // 35 for current trimmed data set range
-const float DLAT = 35.0f / VOLUME_HEIGHT;  // 35 for current trimmed data set range
+const float DLAT = 60.0f / VOLUME_HEIGHT;  // 60 for current trimmed data set range
 const float DLEV = 1000.0f / VOLUME_DEPTH;  // 1000 from max pressure (hPa) but not sure here
 
 const float MIN_TEMP = 210.0f;
@@ -77,10 +79,10 @@ extern __device__ int d_samplesPerPixel;
 extern __device__ bool d_showSilhouettes;
 extern __device__ float d_silhouettesThreshold;
 
-const int lenStopsPythonLike = 5;
+const int lenStopsPythonLike = 11;
 const int lenStopsGrayscale = 2;
 const int lenStopsBluePurpleRed = 3;
-extern __constant__ ColorStop d_stopsPythonLike[5];
+extern __constant__ ColorStop d_stopsPythonLike[11];
 extern __constant__ ColorStop d_stopsGrayscale[2];
 extern __constant__ ColorStop d_stopsBluePurleRed[3];
 

src/gui/MainWindow.cpp

@@ -7,6 +7,7 @@
 
 #include "Shader.h"
 #include "input/Widget.h"
+#include "cuda_error.h"
 
 
 Window::Window(unsigned int w, unsigned int h) {
@@ -98,6 +99,7 @@ void Window::free(float* data) {
 
 
 void Window::tick() {
+  std::cout << "tick tick\n";
   // manually track time diff
 	std::chrono::steady_clock::time_point now = std::chrono::steady_clock::now();
   float diff = (float) std::chrono::duration_cast<std::chrono::milliseconds>(now - this->last_frame).count();
@@ -111,6 +113,7 @@ void Window::tick() {
     this->widget->dateChanged = false;
   }
   
+  std::cout << "setting tick render\n";
   // tick render
 	glBindFramebuffer(GL_FRAMEBUFFER, 0);
 	glDisable(GL_DEPTH_TEST);
@@ -138,11 +141,14 @@ void Window::tick() {
 	glDrawArrays(GL_TRIANGLES, 0, 6); // draw current frame to texture
   
   // render ImGui context
+  std::cout << "rendering\n";
   this->widget->render();
+  std::cout << "done rendering\n";
 
   // check for events
 	glfwSwapBuffers(this->window);
 	glfwPollEvents();
+  std::cout << "done tick tick\n";
 
 }
 

src/gui/Quad.cpp

@@ -83,6 +83,7 @@ void Quad::cuda_init(float* data) {
 
 
 void Quad::render() {
+  check_cuda_errors(cudaGetLastError());
   glBindTexture(GL_TEXTURE_2D, 0);
   this->renderer->render(); 
   glBindTexture(GL_TEXTURE_2D, this->tex);

src/gui/input/Widget.cpp

@@ -4,6 +4,7 @@
 #include <cstdio>
 #include <stdlib.h>
 #include <ctime>
+#include <iostream>
 
 // probably not the cleanest way to do this but it works
 void parseDate(char* string, int dayOfYear) {
@@ -53,7 +54,7 @@ Widget::Widget(GLFWwindow* window) :
   opacityK(0),
   sigmoidShift(0.5f),
   sigmoidExp(-250.0f),
-  tfComboSelected(0),
+  tfComboSelected(2),
   dateChanged(false),
   paused(true),
   renderOnce(false),
@@ -62,8 +63,8 @@ Widget::Widget(GLFWwindow* window) :
   samplesPerPixel(1),
   alphaAcumLimit(0.4f),
   opacityConst(100),
-  showSilhouettes(true),
+  showSilhouettes(false),
-  silhouettesThreshold(0.2f)
+  silhouettesThreshold(0.02f)
 {
   IMGUI_CHECKVERSION();
   ImGui::CreateContext();
@@ -72,6 +73,15 @@ Widget::Widget(GLFWwindow* window) :
   ImGui_ImplGlfw_InitForOpenGL(window, true);
   ImGui_ImplOpenGL3_Init();
 
+  this->cameraPos = Point3::init(59.0f, 77.5f, -18.0f);  // Camera for partially trimmed data set
+  this->pitch = 0.7;
+  this->yaw = 4.85;
+  this->roll = -0.010;
+  this->cameraDir = Vec3::getDirectionFromEuler(pitch, yaw, roll);
+
+  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->dateString = (char*)malloc(512*sizeof(char));
   sprintf(this->dateString, "January 1st");
@@ -84,6 +94,7 @@ Widget::Widget(GLFWwindow* window) :
 // TODO: can be marginally improved by only copying changed values to device - however we're dealing with individual floats here so i dont think the benefit would be all that obvious.
 // TODO: wrap basically all ImGui calls in if statements; better form + allows for checking return values / errors.
 void Widget::tick(double fps) {
+
   if (this->renderOnce) {
     this->renderOnce = false;
     this->paused = true;
@@ -169,12 +180,23 @@ void Widget::tick(double fps) {
   ImGui::DragScalar("X position", ImGuiDataType_Double, &this->cameraPos.x, 0.5f, &min, &max, "%.3f");
   ImGui::DragScalar("Y position", ImGuiDataType_Double, &this->cameraPos.z, 0.5f, &min, &max, "%.3f");
   ImGui::DragScalar("Z position", ImGuiDataType_Double, &this->cameraPos.y, 0.5f, &min, &max, "%.3f");
-  ImGui::DragScalar("X direction", ImGuiDataType_Double, &this->cameraDir.x, 0.005f, &min, &max, "%.3f"); // TODO: should wrap around once fully rotated
+  // ImGui::DragScalar("X direction", ImGuiDataType_Double, &this->cameraDir.x, 0.005f, &min, &max, "%.3f"); // TODO: should wrap around once fully rotated
-  ImGui::DragScalar("Y direction", ImGuiDataType_Double, &this->cameraDir.z, 0.005f, &min, &max, "%.3f");
+  // ImGui::DragScalar("Y direction", ImGuiDataType_Double, &this->cameraDir.z, 0.005f, &min, &max, "%.3f");
-  ImGui::DragScalar("Z direction", ImGuiDataType_Double, &this->cameraDir.y, 0.005f, &min, &max, "%.3f");
+  // ImGui::DragScalar("Z direction", ImGuiDataType_Double, &this->cameraDir.y, 0.005f, &min, &max, "%.3f");
+  ImGui::DragScalar("Pitch", ImGuiDataType_Double, &this->pitch, 0.005f, &min, &max, "%.3f");
+  ImGui::DragScalar("Yaw", ImGuiDataType_Double, &this->yaw, 0.005f, &min, &max, "%.3f");
+  ImGui::DragScalar("Roll", ImGuiDataType_Double, &this->roll, 0.005f, &min, &max, "%.3f");
   if (ImGui::Button("Reset camera")) resetCamera();
   ImGui::End();
 
+  this->cameraDir.setDirectionFromEuler(pitch, yaw, roll);
+
+  // Calculate upCamera
+  Vec3 arbitraryVector = Vec3::init(1, 0, 0);
+  cameraUp = arbitraryVector.cross(cameraDir);
+  cameraUp.normalize();
+  cameraUp.rotateAroundAxis(cameraDir, roll);
+
   copyToDevice();
 }
 
@@ -198,6 +220,7 @@ void Widget::resetLight() {
 
 void Widget::copyToDevice() {
   cudaMemcpyToSymbol(&d_cameraPos, &this->cameraPos, sizeof(Point3));
+  cudaMemcpyToSymbol(&d_cameraUp, &this->cameraUp, sizeof(Point3));
   cudaMemcpyToSymbol(&d_cameraDir, &this->cameraDir, sizeof(Vec3));
   cudaMemcpyToSymbol(&d_lightPos, &this->lightPos, sizeof(Point3));
   cudaMemcpyToSymbol(&d_backgroundColor, &this->bgColor, sizeof(Color3));

src/gui/input/Widget.h

@@ -24,6 +24,8 @@ private:
   // camera controls
   Point3 cameraDir;
   Vec3 cameraPos;
+  double pitch, yaw, roll;
+  Vec3 cameraUp;
 
   // simulation cotnrols
   bool renderOnce;

src/illumination/Raycaster.cu

@@ -38,7 +38,7 @@ __global__ void raycastKernel(float* volumeData, FrameBuffer framebuffer, const
         u *= tanHalfFov;
         v *= tanHalfFov;
 
-        // Find ray direction
+
         Vec3 cameraRight = (d_cameraDir.cross(d_cameraUp)).normalize();
         d_cameraUp = (cameraRight.cross(d_cameraDir)).normalize();
         Vec3 rayDir = (d_cameraDir + cameraRight*u + d_cameraUp*v).normalize();
@@ -156,6 +156,7 @@ Raycaster::Raycaster(cudaGraphicsResource_t resources, int w, int h, float* data
 
 
 void Raycaster::render() {
+  check_cuda_errors(cudaGetLastError());
   check_cuda_errors(cudaGraphicsMapResources(1, &this->resources));
 	check_cuda_errors(cudaGraphicsResourceGetMappedPointer((void**)&(this->fb->buffer), &(this->fb->buffer_size), resources));
 

src/illumination/transferFunction.cu

@@ -111,7 +111,8 @@ __device__ float4 transferFunction(float density, const Vec3& grad, const Point3
     result.x = 0.0f;
     result.y = 0.0f;
     result.z = 0.0f;
-    result.w = alpha;
+    // result.w = alpha;
+    result.w = d_opacityConst;
   }
 
   return result;

src/linalg/vec.h

@@ -31,6 +31,33 @@ struct Vec3 {
     __host__ __device__ Vec3 cross(const Vec3& b) const { return Vec3::init(y * b.z - z * b.y, z * b.x - x * b.z, x * b.y - y * b.x); }
     __host__ __device__ Vec3 normalize() const { double len = sqrt(x * x + y * y + z * z); return Vec3::init(x / len, y / len, z / len); }
     __host__ __device__ double length() const { return sqrt(x * x + y * y + z * z); }
+
+    __host__ __device__ void setDirectionFromEuler(double pitch, double yaw, double roll) {
+        // Compute the direction vector using the Euler angles in radians
+        double cosPitch = cos(pitch);
+        double sinPitch = sin(pitch);
+        double cosYaw = cos(yaw);
+        double sinYaw = sin(yaw);
+
+        // Direction vector components
+        x = cosPitch * cosYaw;
+        y = cosPitch * sinYaw;
+        z = sinPitch;
+    }
+
+    static __host__ __device__ Vec3 getDirectionFromEuler(double pitch, double yaw, double roll) {
+        Vec3 v = Vec3::init(1,0,0);
+        v.setDirectionFromEuler(pitch, yaw, roll);
+        return v;
+    }
+
+    __host__ __device__ void rotateAroundAxis(const Vec3& axis, double angle) {
+        double cosA = cos(angle);
+        double sinA = sin(angle);
+
+        Vec3 rotated = *this * cosA + axis.cross(*this) * sinA + axis * axis.dot(*this) * (1 - cosA);
+        *this = rotated;
+    }
 };
 
 typedef Vec3 Point3;

src/main.cu

@@ -21,7 +21,7 @@ static float* d_volume = nullptr;
 // * time controls - arbitrary skipping to specified point (would require some changes to gpubuffer) (could have)
 
 void getTemperature(std::vector<float>& temperatureData, int idx = 0) {
-    std::string path = "data/trimmed";
+    std::string path = "../../../../hurricane-sandy-data/trimmed";
     // std::string path = "data";
     std::string variable = "T";
     DataReader dataReader(path, variable);
@@ -31,7 +31,7 @@ void getTemperature(std::vector<float>& temperatureData, int idx = 0) {
 }
 
 void getSpeed(std::vector<float>& speedData, int idx = 0) {
-    std::string path = "data/trimmed";
+    std::string path = "../../../../hurricane-sandy-data/trimmed";
     // std::string path = "data";
     std::string varU = "U";
     std::string varV = "V";
@@ -55,7 +55,7 @@ void getSpeed(std::vector<float>& speedData, int idx = 0) {
 
 int main() {
   std::vector<float> data;
-  getTemperature(data, 0);
+  getTemperature(data, 7); // 20121028
   // getSpeed(data, 294);
 
   std::cout << "DATA size: " << data.size() << std::endl;