From 032295a62b37ee13ca3d2f36287d5ff07b3bbc0d Mon Sep 17 00:00:00 2001 From: Tobias Reiter Date: Sun, 21 Dec 2025 00:47:19 +0100 Subject: [PATCH 01/11] Test new CUDA setup --- CMakeLists.txt | 3 +- cmake/gpuUtil.cmake | 6 +++ gpu/CMakeLists.txt | 59 ++++++++++++++++------------ gpu/examples/CMakeLists.txt | 18 ++++++--- gpu/examples/trenchTriangles.cpp | 2 +- gpu/tests/boundaries/CMakeLists.txt | 5 ++- gpu/tests/reflections/CMakeLists.txt | 8 ++-- gpu/tests/rngSeed/CMakeLists.txt | 3 +- gpu/tests/tracer/CMakeLists.txt | 3 +- 9 files changed, 65 insertions(+), 42 deletions(-) create mode 100644 cmake/gpuUtil.cmake diff --git a/CMakeLists.txt b/CMakeLists.txt index d7c6ffb..be4f592 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -99,6 +99,7 @@ include("cmake/cpm.cmake") CPMAddPackage( NAME ViennaCore VERSION 1.8.0 + GIT_TAG msvc GIT_REPOSITORY "https://github.com/ViennaTools/ViennaCore" OPTIONS "VIENNACORE_USE_GPU ${VIENNARAY_USE_GPU}") @@ -153,8 +154,6 @@ endif() if(VIENNARAY_USE_GPU AND VIENNACORE_PTX_DIR) message(STATUS "[ViennaRay] Enabling GPU support") add_subdirectory(gpu) - target_include_directories(${PROJECT_NAME} - INTERFACE $) endif() # -------------------------------------------------------------------------------------------------------- diff --git a/cmake/gpuUtil.cmake b/cmake/gpuUtil.cmake new file mode 100644 index 0000000..78fb492 --- /dev/null +++ b/cmake/gpuUtil.cmake @@ -0,0 +1,6 @@ +# In CMake, functions have their own scope, whereas macros use the scope of the caller. +function(add_gpu_deps target_name) + + add_dependencies(${target_name} disk_pipeline triangle_pipeline line_pipeline callable_wrapper + norm_kernels) +endfunction() diff --git a/gpu/CMakeLists.txt b/gpu/CMakeLists.txt index ec72fb5..f806bdb 100644 --- a/gpu/CMakeLists.txt +++ b/gpu/CMakeLists.txt @@ -1,35 +1,44 @@ -project(ViennaRay-GPU) - -option(VIENNARAY_ENABLE_OPTIXIR_SUPPORT - "Enable support for generating OptiX-IR targeted input files" ON) - -if(CUDA_VERSION VERSION_LESS 11.7) - if(VIENNARAY_ENABLE_OPTIXIR_SUPPORT) - message( - SEND_ERROR "VIENNARAY_ENABLE_OPTIXIR_SUPPORT is not supported in CUDA versions less than 11.7" - ) - endif() -endif() - -if(VIENNARAY_ENABLE_OPTIXIR_SUPPORT) - option(VIENNARAY_GENERATE_OPTIXIR "Generate Optix-IR OptiX shaders" ON) - option(VIENNARAY_GENERATE_PTX "Generate PTX OptiX shaders" OFF) -else() - option(VIENNARAY_GENERATE_OPTIXIR "Generate Optix-IR OptiX shaders" OFF) - option(VIENNARAY_GENERATE_PTX "Generate PTX OptiX shaders" ON) -endif() - #### Set variables set(VIENNARAY_GPU_INCLUDE - "${PROJECT_SOURCE_DIR}/include" + "${PROJECT_SOURCE_DIR}/gpu/include" CACHE STRING "ViennaRay GPU headers.") set(VIENNARAY_CUDA_KERNELS - "${PROJECT_SOURCE_DIR}/kernels/normKernels.cu" + "${PROJECT_SOURCE_DIR}/gpu/kernels/normKernels.cu" CACHE STRING "ViennaRay CUDA kernel source files.") set(VIENNARAY_PIPELINE_DIR - "${PROJECT_SOURCE_DIR}/pipelines" + "${PROJECT_SOURCE_DIR}/gpu/pipelines" CACHE STRING "ViennaRay pipeline directory.") -include("../cmake/generate_ptx.cmake") + +target_include_directories(${PROJECT_NAME} + INTERFACE $) + +if(NOT DEFINED VIENNACORE_PTX_INCLUDE_DIRS) + set(VIENNACORE_PTX_INCLUDE_DIRS + "" + CACHE STRING "ViennaCore PTX include directories") +endif() +list(APPEND VIENNACORE_PTX_INCLUDE_DIRS "${VIENNARAY_GPU_INCLUDE}") + +if(VIENNARAY_GPU_DOUBLE_PRECISION) + if(NOT DEFINED VIENNACORE_PTX_DEFINES) + set(VIENNACORE_PTX_DEFINES + "" + CACHE STRING "ViennaCore PTX compile definitions") + endif() + list(APPEND VIENNACORE_PTX_DEFINES "VIENNARAY_GPU_DOUBLE_PRECISION=1") +endif() + +# Add the general pipelines and callable wrapper +viennacore_add_optixir(disk_pipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineDisk.cu) +viennacore_add_optixir(triangle_pipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineTriangle.cu) +viennacore_add_optixir(line_pipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineLine.cu) +viennacore_add_optixir(callable_wrapper ${VIENNARAY_PIPELINE_DIR}/CallableWrapper.cu) + +# Add the norm kernels +viennacore_add_ptx(norm_kernels ${VIENNARAY_CUDA_KERNELS}) + +set(VIENNARAY_GPU_DEPENDENCIES disk_pipeline triangle_pipeline line_pipeline callable_wrapper + norm_kernels) if(VIENNARAY_BUILD_EXAMPLES) message(STATUS "[ViennaRay] Adding GPU Examples") diff --git a/gpu/examples/CMakeLists.txt b/gpu/examples/CMakeLists.txt index aca59da..3c3fa78 100644 --- a/gpu/examples/CMakeLists.txt +++ b/gpu/examples/CMakeLists.txt @@ -2,17 +2,23 @@ project(ViennaRay-GPU_Examples) add_custom_target(ViennaRay-GPU_Examples ALL) -add_gpu_executable(GPU_trenchTriangles target_name trenchTriangles.cpp) +add_executable(GPU_trenchTriangles trenchTriangles.cpp) +target_link_libraries(GPU_trenchTriangles PRIVATE ViennaRay) +add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) configure_file(${CMAKE_SOURCE_DIR}/examples/triangle3D/trenchMesh.dat ${CMAKE_CURRENT_BINARY_DIR}/trenchMesh.dat COPYONLY) -add_dependencies(ViennaRay-GPU_Examples ${target_name}) +add_dependencies(ViennaRay-GPU_Examples GPU_trenchTriangles) -add_gpu_executable(GPU_trenchDisks target_name trenchDisks.cpp) +add_executable(GPU_trenchDisks trenchDisks.cpp) +target_link_libraries(GPU_trenchDisks PRIVATE ViennaRay) +add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) configure_file(${CMAKE_SOURCE_DIR}/examples/disk3D/trenchGrid3D.dat ${CMAKE_CURRENT_BINARY_DIR}/trenchGrid3D.dat COPYONLY) -add_dependencies(ViennaRay-GPU_Examples ${target_name}) +add_dependencies(ViennaRay-GPU_Examples GPU_trenchDisks) -add_gpu_executable(GPU_trenchLines target_name trenchLines.cpp) +add_executable(GPU_trenchLines trenchLines.cpp) +target_link_libraries(GPU_trenchLines PRIVATE ViennaRay) +add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) configure_file(${CMAKE_SOURCE_DIR}/examples/triangle2D/lineMesh.dat ${CMAKE_CURRENT_BINARY_DIR}/lineMesh.dat COPYONLY) -add_dependencies(ViennaRay-GPU_Examples ${target_name}) +add_dependencies(ViennaRay-GPU_Examples GPU_trenchLines) diff --git a/gpu/examples/trenchTriangles.cpp b/gpu/examples/trenchTriangles.cpp index aee4afb..ccb0362 100644 --- a/gpu/examples/trenchTriangles.cpp +++ b/gpu/examples/trenchTriangles.cpp @@ -1,4 +1,4 @@ -#include +#include "raygTraceTriangle.hpp" #include #include diff --git a/gpu/tests/boundaries/CMakeLists.txt b/gpu/tests/boundaries/CMakeLists.txt index e388be4..d398ebd 100644 --- a/gpu/tests/boundaries/CMakeLists.txt +++ b/gpu/tests/boundaries/CMakeLists.txt @@ -1,7 +1,8 @@ project(boundaries LANGUAGES CXX) -generate_pipeline(${PROJECT_NAME} generated_files TestPipelineTriangle.cu) -add_executable(${PROJECT_NAME} ${PROJECT_NAME}.cpp ${generated_files}) +viennacore_add_optixir(viennaray_test_pipeline ${CMAKE_CURRENT_SOURCE_DIR}/TestPipelineTriangle.cu) +add_executable(${PROJECT_NAME} ${PROJECT_NAME}.cpp) +add_dependencies(${PROJECT_NAME} viennaray_test_pipeline) target_link_libraries(${PROJECT_NAME} PRIVATE ViennaRay) add_dependencies(ViennaRay-GPU_Tests ${PROJECT_NAME}) diff --git a/gpu/tests/reflections/CMakeLists.txt b/gpu/tests/reflections/CMakeLists.txt index 63e83b7..ba38fdc 100644 --- a/gpu/tests/reflections/CMakeLists.txt +++ b/gpu/tests/reflections/CMakeLists.txt @@ -1,11 +1,11 @@ project(testReflections LANGUAGES CXX) # wrap CUDA kernel only -generate_kernel(generated_files "testReflections.cu") +viennacore_add_ptx(viennaray_test_reflections ${CMAKE_CURRENT_SOURCE_DIR}/testReflections.cu) -add_executable(${PROJECT_NAME} "${PROJECT_NAME}.cpp" ${generated_files}) -target_include_directories(${PROJECT_NAME} PRIVATE ${OptiX_INCLUDE_DIR}) -target_link_libraries(${PROJECT_NAME} PRIVATE ViennaRay ${VIENNACORE_GPU_LIBS}) +add_executable(${PROJECT_NAME} "${PROJECT_NAME}.cpp") +add_dependencies(${PROJECT_NAME} viennaray_test_reflections) +target_link_libraries(${PROJECT_NAME} PRIVATE ViennaRay) add_dependencies(ViennaRay-GPU_Tests ${PROJECT_NAME}) add_test(NAME ${PROJECT_NAME} COMMAND $) diff --git a/gpu/tests/rngSeed/CMakeLists.txt b/gpu/tests/rngSeed/CMakeLists.txt index d47d5d5..db68896 100644 --- a/gpu/tests/rngSeed/CMakeLists.txt +++ b/gpu/tests/rngSeed/CMakeLists.txt @@ -1,6 +1,7 @@ project(GPU_rngSeed LANGUAGES CXX) -add_gpu_executable(${PROJECT_NAME} target_name rngSeed.cpp) +add_executable(${PROJECT_NAME} rngSeed.cpp) +add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) add_dependencies(ViennaRay-GPU_Tests ${PROJECT_NAME}) add_test(NAME ${PROJECT_NAME} COMMAND $) diff --git a/gpu/tests/tracer/CMakeLists.txt b/gpu/tests/tracer/CMakeLists.txt index 4480781..5fac0f5 100644 --- a/gpu/tests/tracer/CMakeLists.txt +++ b/gpu/tests/tracer/CMakeLists.txt @@ -1,6 +1,7 @@ project(tracer LANGUAGES CXX) -add_gpu_executable(${PROJECT_NAME} target_name ${PROJECT_NAME}.cpp) +add_executable(${PROJECT_NAME} ${PROJECT_NAME}.cpp) +add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) add_dependencies(ViennaRay-GPU_Tests ${PROJECT_NAME}) add_test(NAME ${PROJECT_NAME} COMMAND $) From f944c6af32c764a2626a4f203a5a3423ce418853 Mon Sep 17 00:00:00 2001 From: reiter Date: Sun, 21 Dec 2025 23:43:25 +0100 Subject: [PATCH 02/11] Fixes, rename GPU targets --- gpu/CMakeLists.txt | 14 +++++++------- gpu/tests/rngSeed/CMakeLists.txt | 1 + gpu/tests/tracer/CMakeLists.txt | 1 + 3 files changed, 9 insertions(+), 7 deletions(-) diff --git a/gpu/CMakeLists.txt b/gpu/CMakeLists.txt index f806bdb..ee40ba1 100644 --- a/gpu/CMakeLists.txt +++ b/gpu/CMakeLists.txt @@ -29,16 +29,16 @@ if(VIENNARAY_GPU_DOUBLE_PRECISION) endif() # Add the general pipelines and callable wrapper -viennacore_add_optixir(disk_pipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineDisk.cu) -viennacore_add_optixir(triangle_pipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineTriangle.cu) -viennacore_add_optixir(line_pipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineLine.cu) -viennacore_add_optixir(callable_wrapper ${VIENNARAY_PIPELINE_DIR}/CallableWrapper.cu) +viennacore_add_optixir(diskPipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineDisk.cu) +viennacore_add_optixir(trianglePipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineTriangle.cu) +viennacore_add_optixir(linePipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineLine.cu) +viennacore_add_optixir(ViennaRayCallableWrapper ${VIENNARAY_PIPELINE_DIR}/CallableWrapper.cu) # Add the norm kernels -viennacore_add_ptx(norm_kernels ${VIENNARAY_CUDA_KERNELS}) +viennacore_add_ptx(normKernels ${VIENNARAY_CUDA_KERNELS}) -set(VIENNARAY_GPU_DEPENDENCIES disk_pipeline triangle_pipeline line_pipeline callable_wrapper - norm_kernels) +set(VIENNARAY_GPU_DEPENDENCIES diskPipeline trianglePipeline linePipeline ViennaRayCallableWrapper + normKernels) if(VIENNARAY_BUILD_EXAMPLES) message(STATUS "[ViennaRay] Adding GPU Examples") diff --git a/gpu/tests/rngSeed/CMakeLists.txt b/gpu/tests/rngSeed/CMakeLists.txt index db68896..803fd99 100644 --- a/gpu/tests/rngSeed/CMakeLists.txt +++ b/gpu/tests/rngSeed/CMakeLists.txt @@ -1,6 +1,7 @@ project(GPU_rngSeed LANGUAGES CXX) add_executable(${PROJECT_NAME} rngSeed.cpp) +target_link_libraries(${PROJECT_NAME} PRIVATE ViennaRay) add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) add_dependencies(ViennaRay-GPU_Tests ${PROJECT_NAME}) diff --git a/gpu/tests/tracer/CMakeLists.txt b/gpu/tests/tracer/CMakeLists.txt index 5fac0f5..9b1aa1e 100644 --- a/gpu/tests/tracer/CMakeLists.txt +++ b/gpu/tests/tracer/CMakeLists.txt @@ -1,6 +1,7 @@ project(tracer LANGUAGES CXX) add_executable(${PROJECT_NAME} ${PROJECT_NAME}.cpp) +target_link_libraries(${PROJECT_NAME} PRIVATE ViennaRay) add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) add_dependencies(ViennaRay-GPU_Tests ${PROJECT_NAME}) From 71a9fb36ae7df6011c68d0808cbacc66d2261a93 Mon Sep 17 00:00:00 2001 From: reiter Date: Tue, 23 Dec 2025 10:44:52 +0100 Subject: [PATCH 03/11] Force override PTX cache --- gpu/CMakeLists.txt | 20 +++++++------------- 1 file changed, 7 insertions(+), 13 deletions(-) diff --git a/gpu/CMakeLists.txt b/gpu/CMakeLists.txt index ee40ba1..d16afda 100644 --- a/gpu/CMakeLists.txt +++ b/gpu/CMakeLists.txt @@ -12,20 +12,14 @@ set(VIENNARAY_PIPELINE_DIR target_include_directories(${PROJECT_NAME} INTERFACE $) -if(NOT DEFINED VIENNACORE_PTX_INCLUDE_DIRS) - set(VIENNACORE_PTX_INCLUDE_DIRS - "" - CACHE STRING "ViennaCore PTX include directories") -endif() -list(APPEND VIENNACORE_PTX_INCLUDE_DIRS "${VIENNARAY_GPU_INCLUDE}") - +# Incldue directories for PTX and optix-ir compilation (force cache to update with new entries) +set(VIENNACORE_PTX_INCLUDE_DIRS + ${VIENNACORE_PTX_INCLUDE_DIRS} ${VIENNARAY_GPU_INCLUDE} + CACHE STRING "PTX include directories" FORCE) if(VIENNARAY_GPU_DOUBLE_PRECISION) - if(NOT DEFINED VIENNACORE_PTX_DEFINES) - set(VIENNACORE_PTX_DEFINES - "" - CACHE STRING "ViennaCore PTX compile definitions") - endif() - list(APPEND VIENNACORE_PTX_DEFINES "VIENNARAY_GPU_DOUBLE_PRECISION=1") + set(VIENNACORE_PTX_DEFINES + ${VIENNACORE_PTX_DEFINES} "VIENNARAY_GPU_DOUBLE_PRECISION=1" + CACHE STRING "PTX compile definitions" FORCE) endif() # Add the general pipelines and callable wrapper From e4583c4c167cf0b73886541b171d1f9eb201e7ed Mon Sep 17 00:00:00 2001 From: reiter Date: Tue, 23 Dec 2025 11:32:07 +0100 Subject: [PATCH 04/11] Fix GPU targets --- gpu/CMakeLists.txt | 11 +++++------ gpu/examples/CMakeLists.txt | 5 ++--- gpu/examples/trenchDisks.cpp | 5 ++--- gpu/examples/trenchLines.cpp | 7 +++---- gpu/examples/trenchTriangles.cpp | 5 ++--- 5 files changed, 14 insertions(+), 19 deletions(-) diff --git a/gpu/CMakeLists.txt b/gpu/CMakeLists.txt index d16afda..0638657 100644 --- a/gpu/CMakeLists.txt +++ b/gpu/CMakeLists.txt @@ -9,8 +9,7 @@ set(VIENNARAY_PIPELINE_DIR "${PROJECT_SOURCE_DIR}/gpu/pipelines" CACHE STRING "ViennaRay pipeline directory.") -target_include_directories(${PROJECT_NAME} - INTERFACE $) +target_include_directories(${PROJECT_NAME} INTERFACE $) # Incldue directories for PTX and optix-ir compilation (force cache to update with new entries) set(VIENNACORE_PTX_INCLUDE_DIRS @@ -23,15 +22,15 @@ if(VIENNARAY_GPU_DOUBLE_PRECISION) endif() # Add the general pipelines and callable wrapper -viennacore_add_optixir(diskPipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineDisk.cu) -viennacore_add_optixir(trianglePipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineTriangle.cu) -viennacore_add_optixir(linePipeline ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineLine.cu) +viennacore_add_optixir(GeneralPipelineDisk ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineDisk.cu) +viennacore_add_optixir(GeneralPipelineTriangle ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineTriangle.cu) +viennacore_add_optixir(GeneralPipelineLine ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineLine.cu) viennacore_add_optixir(ViennaRayCallableWrapper ${VIENNARAY_PIPELINE_DIR}/CallableWrapper.cu) # Add the norm kernels viennacore_add_ptx(normKernels ${VIENNARAY_CUDA_KERNELS}) -set(VIENNARAY_GPU_DEPENDENCIES diskPipeline trianglePipeline linePipeline ViennaRayCallableWrapper +set(VIENNARAY_GPU_DEPENDENCIES GeneralPipelineDisk GeneralPipelineTriangle GeneralPipelineLine normKernels) if(VIENNARAY_BUILD_EXAMPLES) diff --git a/gpu/examples/CMakeLists.txt b/gpu/examples/CMakeLists.txt index 3c3fa78..62d5a3f 100644 --- a/gpu/examples/CMakeLists.txt +++ b/gpu/examples/CMakeLists.txt @@ -4,21 +4,20 @@ add_custom_target(ViennaRay-GPU_Examples ALL) add_executable(GPU_trenchTriangles trenchTriangles.cpp) target_link_libraries(GPU_trenchTriangles PRIVATE ViennaRay) -add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) configure_file(${CMAKE_SOURCE_DIR}/examples/triangle3D/trenchMesh.dat ${CMAKE_CURRENT_BINARY_DIR}/trenchMesh.dat COPYONLY) add_dependencies(ViennaRay-GPU_Examples GPU_trenchTriangles) add_executable(GPU_trenchDisks trenchDisks.cpp) target_link_libraries(GPU_trenchDisks PRIVATE ViennaRay) -add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) configure_file(${CMAKE_SOURCE_DIR}/examples/disk3D/trenchGrid3D.dat ${CMAKE_CURRENT_BINARY_DIR}/trenchGrid3D.dat COPYONLY) add_dependencies(ViennaRay-GPU_Examples GPU_trenchDisks) add_executable(GPU_trenchLines trenchLines.cpp) target_link_libraries(GPU_trenchLines PRIVATE ViennaRay) -add_dependencies(${PROJECT_NAME} ${VIENNARAY_GPU_DEPENDENCIES}) configure_file(${CMAKE_SOURCE_DIR}/examples/triangle2D/lineMesh.dat ${CMAKE_CURRENT_BINARY_DIR}/lineMesh.dat COPYONLY) add_dependencies(ViennaRay-GPU_Examples GPU_trenchLines) + +add_dependencies(ViennaRay-GPU_Examples ${VIENNARAY_GPU_DEPENDENCIES} ViennaRayCallableWrapper) diff --git a/gpu/examples/trenchDisks.cpp b/gpu/examples/trenchDisks.cpp index 1c456b9..a11c1ce 100644 --- a/gpu/examples/trenchDisks.cpp +++ b/gpu/examples/trenchDisks.cpp @@ -11,8 +11,7 @@ int main() { using NumericType = float; Logger::setLogLevel(LogLevel::DEBUG); - auto context = DeviceContext::createContext("../../lib/ptx", 0); - // relative to build directory + auto context = DeviceContext::createContext(); // Read stored geometry grid NumericType gridDelta; @@ -43,7 +42,7 @@ int main() { gpu::TraceDisk tracer(context); tracer.setGeometry(mesh); tracer.setMaterialIds(materialIds); - tracer.setCallables("CallableWrapper", context->modulePath); + tracer.setCallables("ViennaRayCallableWrapper", context->modulePath); tracer.setParticleCallableMap({pMap, cMap}); tracer.setNumberOfRaysPerPoint(1000); tracer.insertNextParticle(particle); diff --git a/gpu/examples/trenchLines.cpp b/gpu/examples/trenchLines.cpp index 53a8efb..0f805e4 100644 --- a/gpu/examples/trenchLines.cpp +++ b/gpu/examples/trenchLines.cpp @@ -12,8 +12,7 @@ int main() { using NumericType = float; Logger::setLogLevel(LogLevel::DEBUG); - auto context = DeviceContext::createContext("../../lib/ptx", 0); - // relative to build directory + auto context = DeviceContext::createContext(); // Read stored geometry grid std::vector> points; @@ -46,7 +45,7 @@ int main() { gpu::TraceLine tracer(context); tracer.setGeometry(mesh); tracer.setMaterialIds(materialIds); - tracer.setCallables("CallableWrapper", context->modulePath); + tracer.setCallables("ViennaRayCallableWrapper", context->modulePath); tracer.setParticleCallableMap({pMap, cMap}); tracer.setNumberOfRaysPerPoint(5000); tracer.insertNextParticle(particle); @@ -72,7 +71,7 @@ int main() { gpu::TraceTriangle triangleTracer(context); triangleTracer.setGeometry(triMesh); triangleTracer.setMaterialIds(materialIds); - triangleTracer.setCallables("CallableWrapper", context->modulePath); + triangleTracer.setCallables("ViennaRayCallableWrapper", context->modulePath); triangleTracer.setParticleCallableMap({pMap, cMap}); triangleTracer.setNumberOfRaysPerPoint(5000); triangleTracer.insertNextParticle(particle); diff --git a/gpu/examples/trenchTriangles.cpp b/gpu/examples/trenchTriangles.cpp index ccb0362..97b338e 100644 --- a/gpu/examples/trenchTriangles.cpp +++ b/gpu/examples/trenchTriangles.cpp @@ -14,8 +14,7 @@ int main(int argc, char **argv) { using NumericType = double; Logger::setLogLevel(LogLevel::DEBUG); - auto context = DeviceContext::createContext("../../lib/ptx", 0); - // relative to build directory + auto context = DeviceContext::createContext(); std::vector> points; std::vector> triangles; @@ -44,7 +43,7 @@ int main(int argc, char **argv) { gpu::TraceTriangle tracer(context); tracer.setGeometry(mesh); tracer.setMaterialIds(materialIds); - tracer.setCallables("CallableWrapper", context->modulePath); + tracer.setCallables("ViennaRayCallableWrapper", context->modulePath); tracer.setParticleCallableMap({pMap, cMap}); tracer.setNumberOfRaysPerPoint(5000); tracer.insertNextParticle(particle); From 2c99a7b720d8dbb41cb75f4d2496851194a338ea Mon Sep 17 00:00:00 2001 From: Tobias Reiter Date: Tue, 23 Dec 2025 11:42:00 +0100 Subject: [PATCH 05/11] Cache GPU dependencies --- gpu/CMakeLists.txt | 5 +++-- 1 file changed, 3 insertions(+), 2 deletions(-) diff --git a/gpu/CMakeLists.txt b/gpu/CMakeLists.txt index 0638657..3df06e8 100644 --- a/gpu/CMakeLists.txt +++ b/gpu/CMakeLists.txt @@ -30,8 +30,9 @@ viennacore_add_optixir(ViennaRayCallableWrapper ${VIENNARAY_PIPELINE_DIR}/Callab # Add the norm kernels viennacore_add_ptx(normKernels ${VIENNARAY_CUDA_KERNELS}) -set(VIENNARAY_GPU_DEPENDENCIES GeneralPipelineDisk GeneralPipelineTriangle GeneralPipelineLine - normKernels) +set(VIENNARAY_GPU_DEPENDENCIES + GeneralPipelineDisk GeneralPipelineTriangle GeneralPipelineLine normKernels + CACHE STRING "ViennaRay GPU dependencies.") if(VIENNARAY_BUILD_EXAMPLES) message(STATUS "[ViennaRay] Adding GPU Examples") From 6258ed0796f5cf7e2626cefec02b707be412af8b Mon Sep 17 00:00:00 2001 From: Tobias Reiter Date: Fri, 2 Jan 2026 17:12:21 +0100 Subject: [PATCH 06/11] MSCV fixes --- examples/disk2D/disk2D.cpp | 3 ++- examples/disk3D/disk3D.cpp | 3 ++- examples/triangle2D/triangle2D.cpp | 3 ++- examples/triangle3D/triangle3D.cpp | 3 ++- gpu/CMakeLists.txt | 18 +++++++++++------- gpu/include/raygReflection.hpp | 1 + gpu/include/raygTrace.hpp | 10 ++++++---- include/viennaray/rayParticle.hpp | 2 +- include/viennaray/rayTraceKernel.hpp | 2 +- include/viennaray/rayUtil.hpp | 1 - tests/diskAreas/diskAreas.cpp | 3 ++- tests/particle/particle.cpp | 13 ++++++++----- tests/rngSeed/rngSeed.cpp | 3 ++- tests/trace2D/trace2D.cpp | 4 ++-- tests/traceInterface/traceInterface.cpp | 4 ++-- 15 files changed, 44 insertions(+), 29 deletions(-) diff --git a/examples/disk2D/disk2D.cpp b/examples/disk2D/disk2D.cpp index 9834bd0..7840a6c 100644 --- a/examples/disk2D/disk2D.cpp +++ b/examples/disk2D/disk2D.cpp @@ -36,8 +36,9 @@ int main() { // defined, but has to interface the rayParticle class and // provide the functions: initNew(...), surfaceCollision(...), // surfaceReflection(...). + NumericType stickingProbability = 0.1; auto particle = - std::make_unique>(0.5, "flux"); + std::make_unique>(stickingProbability, "flux"); TraceDisk rayTracer; rayTracer.setGeometry(points, normals, gridDelta); diff --git a/examples/disk3D/disk3D.cpp b/examples/disk3D/disk3D.cpp index bf62e41..8e5aeb4 100644 --- a/examples/disk3D/disk3D.cpp +++ b/examples/disk3D/disk3D.cpp @@ -39,8 +39,9 @@ int main() { // defined, but has to interface the rayParticle class and // provide the functions: initNew(...), surfaceCollision(...), // surfaceReflection(...). + NumericType stickingProbability = 0.1; auto particle = - std::make_unique>(0.1, "flux"); + std::make_unique>(stickingProbability, "flux"); TraceDisk rayTracer; rayTracer.setGeometry(points, normals, gridDelta); diff --git a/examples/triangle2D/triangle2D.cpp b/examples/triangle2D/triangle2D.cpp index 1a7a26a..4f7a62f 100644 --- a/examples/triangle2D/triangle2D.cpp +++ b/examples/triangle2D/triangle2D.cpp @@ -23,8 +23,9 @@ int main() { TraceTriangle tracer; tracer.setGeometry(lineMesh); + NumericType stickingProbability = 0.1; auto particle = - std::make_unique>(0.1, "flux"); + std::make_unique>(stickingProbability, "flux"); tracer.setParticleType(particle); tracer.setNumberOfRaysPerPoint(5000); diff --git a/examples/triangle3D/triangle3D.cpp b/examples/triangle3D/triangle3D.cpp index a416845..f3046ae 100644 --- a/examples/triangle3D/triangle3D.cpp +++ b/examples/triangle3D/triangle3D.cpp @@ -23,8 +23,9 @@ int main() { TraceTriangle tracer; tracer.setGeometry(mesh); + NumericType stickingProbability = 0.1; auto particle = - std::make_unique>(0.1, "flux"); + std::make_unique>(stickingProbability, "flux"); tracer.setParticleType(particle); tracer.setNumberOfRaysPerPoint(2000); diff --git a/gpu/CMakeLists.txt b/gpu/CMakeLists.txt index 3df06e8..3116769 100644 --- a/gpu/CMakeLists.txt +++ b/gpu/CMakeLists.txt @@ -12,13 +12,15 @@ set(VIENNARAY_PIPELINE_DIR target_include_directories(${PROJECT_NAME} INTERFACE $) # Incldue directories for PTX and optix-ir compilation (force cache to update with new entries) -set(VIENNACORE_PTX_INCLUDE_DIRS - ${VIENNACORE_PTX_INCLUDE_DIRS} ${VIENNARAY_GPU_INCLUDE} - CACHE STRING "PTX include directories" FORCE) -if(VIENNARAY_GPU_DOUBLE_PRECISION) - set(VIENNACORE_PTX_DEFINES - ${VIENNACORE_PTX_DEFINES} "VIENNARAY_GPU_DOUBLE_PRECISION=1" - CACHE STRING "PTX compile definitions" FORCE) +if (NOT VIENNARAY_PASSED_FIRST_CONFIGURE) + set(VIENNACORE_PTX_INCLUDE_DIRS + ${VIENNACORE_PTX_INCLUDE_DIRS} ${VIENNARAY_GPU_INCLUDE} + CACHE STRING "PTX include directories" FORCE) + if(VIENNARAY_GPU_DOUBLE_PRECISION) + set(VIENNACORE_PTX_DEFINES + ${VIENNACORE_PTX_DEFINES} "VIENNARAY_GPU_DOUBLE_PRECISION=1" + CACHE STRING "PTX compile definitions" FORCE) + endif() endif() # Add the general pipelines and callable wrapper @@ -43,3 +45,5 @@ if(VIENNARAY_BUILD_TESTS) message(STATUS "[ViennaRay] Adding GPU Tests") add_subdirectory(tests) endif() + +set(VIENNARAY_PASSED_FIRST_CONFIGURE ON CACHE INTERNAL "Indicates whether the first configure has been passed.") diff --git a/gpu/include/raygReflection.hpp b/gpu/include/raygReflection.hpp index f2eebab..64fbcac 100644 --- a/gpu/include/raygReflection.hpp +++ b/gpu/include/raygReflection.hpp @@ -49,6 +49,7 @@ __device__ __inline__ Vec3Df computeNormal(const void *sbtData, printf("Unknown geometry type in computeNormal\n"); } break; } + return Vec3Df{0.f, 0.f, 0.f}; } __device__ __forceinline__ Vec3Df getNormal(const void *sbtData, diff --git a/gpu/include/raygTrace.hpp b/gpu/include/raygTrace.hpp index 93b321c..1f96af5 100644 --- a/gpu/include/raygTrace.hpp +++ b/gpu/include/raygTrace.hpp @@ -260,13 +260,13 @@ template class Trace { std::vector materialIdsMapped(launchParams_.numElements); #pragma omp parallel for - for (size_t i = 0; i < launchParams_.numElements; i++) { + for (int i = 0; i < launchParams_.numElements; i++) { materialIdsMapped[i] = materialIdMap[materialIds[i]]; } materialIdsBuffer_.allocUpload(materialIdsMapped); } else { std::vector materialIdsMapped(launchParams_.numElements); - for (size_t i = 0; i < launchParams_.numElements; i++) { + for (int i = 0; i < launchParams_.numElements; i++) { materialIdsMapped[i] = static_cast(materialIds[i]); } materialIdsBuffer_.allocUpload(materialIdsMapped); @@ -538,7 +538,8 @@ template class Trace { " not found."); } - auto pipelineInput = getInputData(pipelinePath.c_str(), inputSize); + std::string path_str = pipelinePath.string(); // explicit conversion + auto pipelineInput = getInputData(path_str.c_str(), inputSize); if (!pipelineInput) { VIENNACORE_LOG_ERROR("Pipeline file " + pipelinePath.string() + " not found."); @@ -550,7 +551,8 @@ template class Trace { VIENNACORE_LOG_WARNING("No callable file set."); return; } - auto callableInput = getInputData(callableFile_.c_str(), inputSize); + std::string callable_path_str = callableFile_.string(); // explicit conversion + auto callableInput = getInputData(callable_path_str.c_str(), inputSize); if (!callableInput) { VIENNACORE_LOG_ERROR("Callable file " + callableFile_.string() + " not found."); diff --git a/include/viennaray/rayParticle.hpp b/include/viennaray/rayParticle.hpp index ce09021..8824f29 100644 --- a/include/viennaray/rayParticle.hpp +++ b/include/viennaray/rayParticle.hpp @@ -11,7 +11,7 @@ #define VIENNARAY_PARTICLE_STOP \ std::pair> { \ - 1., Vec3D { 0., 0., 0. } \ + NumericType(1), Vec3D{} \ } namespace viennaray { diff --git a/include/viennaray/rayTraceKernel.hpp b/include/viennaray/rayTraceKernel.hpp index 3111336..f47700a 100644 --- a/include/viennaray/rayTraceKernel.hpp +++ b/include/viennaray/rayTraceKernel.hpp @@ -78,7 +78,7 @@ template class TraceKernel { data = dataLog_; assert(data.data.size() == dataLog_.data.size()); for (auto &d : data.data) { - std::fill(d.begin(), d.end(), 0.); + std::fill(d.begin(), d.end(), NumericType(0)); } } diff --git a/include/viennaray/rayUtil.hpp b/include/viennaray/rayUtil.hpp index ea38e98..08f7f9f 100644 --- a/include/viennaray/rayUtil.hpp +++ b/include/viennaray/rayUtil.hpp @@ -293,7 +293,6 @@ template if (len2 == T(0)) { VIENNACORE_LOG_ERROR( "Cannot build orthonormal basis for zero-length vector"); - return B; } const T invLen = T(1) / std::sqrt(len2); u = u * invLen; diff --git a/tests/diskAreas/diskAreas.cpp b/tests/diskAreas/diskAreas.cpp index 4ee20de..75ff44a 100644 --- a/tests/diskAreas/diskAreas.cpp +++ b/tests/diskAreas/diskAreas.cpp @@ -40,7 +40,8 @@ int main() { traceSettings); std::array, 3> orthoBasis; auto raySource = std::make_unique>( - boundingBox, 1., traceSettings, geometry.getNumPrimitives(), false, + boundingBox, NumericType(1), traceSettings, geometry.getNumPrimitives(), + false, orthoBasis); geometry.computeDiskAreas(boundary); diff --git a/tests/particle/particle.cpp b/tests/particle/particle.cpp index e3bda31..8fe0a82 100644 --- a/tests/particle/particle.cpp +++ b/tests/particle/particle.cpp @@ -10,8 +10,9 @@ using namespace viennaray; template void RunTest() { { - auto particle = - std::make_unique>(1., "test"); + NumericType stickingProbability = 1.0; + auto particle = std::make_unique>( + stickingProbability, "test"); NumericType sourcePower = particle->getSourceDistributionPower(); VC_TEST_ASSERT(sourcePower == 1.); @@ -25,11 +26,13 @@ template void RunTest() { } { + NumericType stickingProbability = 1.0; + NumericType sourcePower = 50.; auto particle = - std::make_unique>(1., 100., "test"); + std::make_unique>(stickingProbability, sourcePower, "test"); - NumericType sourcePower = particle->getSourceDistributionPower(); - VC_TEST_ASSERT(sourcePower == 100.); + NumericType sourcePowerTest = particle->getSourceDistributionPower(); + VC_TEST_ASSERT(sourcePowerTest == 100.); auto labels = particle->getLocalDataLabels(); VC_TEST_ASSERT(labels.size() == 1); diff --git a/tests/rngSeed/rngSeed.cpp b/tests/rngSeed/rngSeed.cpp index 5a92919..777651d 100644 --- a/tests/rngSeed/rngSeed.cpp +++ b/tests/rngSeed/rngSeed.cpp @@ -15,8 +15,9 @@ int main() { std::vector> normals; rayInternal::createPlaneGrid(gridDelta, extent, {0, 1, 2}, points, normals); + NumericType stickingProbability = 1.0; auto particle = - std::make_unique>(1.0, "hitFlux"); + std::make_unique>(stickingProbability, "hitFlux"); std::vector flux1, flux2; diff --git a/tests/trace2D/trace2D.cpp b/tests/trace2D/trace2D.cpp index 0b8d25e..a0eb974 100644 --- a/tests/trace2D/trace2D.cpp +++ b/tests/trace2D/trace2D.cpp @@ -22,8 +22,8 @@ int main() { BoundaryCondition boundaryConds[D]; boundaryConds[0] = BoundaryCondition::REFLECTIVE; boundaryConds[1] = BoundaryCondition::REFLECTIVE; - auto particle = - std::make_unique>(1.0, "hitFlux"); + auto particle = std::make_unique>( + NumericType(1), "hitFlux"); TraceDisk rayTracer; rayTracer.setParticleType(particle); diff --git a/tests/traceInterface/traceInterface.cpp b/tests/traceInterface/traceInterface.cpp index 6e2464f..8587ac9 100644 --- a/tests/traceInterface/traceInterface.cpp +++ b/tests/traceInterface/traceInterface.cpp @@ -38,8 +38,8 @@ int main() { boundaryConds[0] = BoundaryCondition::REFLECTIVE; boundaryConds[1] = BoundaryCondition::REFLECTIVE; boundaryConds[2] = BoundaryCondition::REFLECTIVE; - auto particle = - std::make_unique>(1.0, "hitFlux"); + auto particle = std::make_unique>( + NumericType(1), "hitFlux"); TraceDisk rayTracer; rayTracer.setParticleType(particle); From 7cdf6d9c7bd93ed96f646295c7fc11acb9e94295 Mon Sep 17 00:00:00 2001 From: reiter Date: Fri, 2 Jan 2026 17:34:19 +0100 Subject: [PATCH 07/11] Format --- examples/disk2D/disk2D.cpp | 4 ++-- examples/disk3D/disk3D.cpp | 4 ++-- examples/triangle2D/triangle2D.cpp | 4 ++-- examples/triangle3D/triangle3D.cpp | 4 ++-- gpu/CMakeLists.txt | 22 ++++++++++++---------- gpu/include/raygTrace.hpp | 3 ++- include/viennaray/rayParticle.hpp | 2 +- tests/diskAreas/diskAreas.cpp | 3 +-- tests/particle/particle.cpp | 4 ++-- tests/rngSeed/rngSeed.cpp | 4 ++-- 10 files changed, 28 insertions(+), 26 deletions(-) diff --git a/examples/disk2D/disk2D.cpp b/examples/disk2D/disk2D.cpp index 7840a6c..b5579bb 100644 --- a/examples/disk2D/disk2D.cpp +++ b/examples/disk2D/disk2D.cpp @@ -37,8 +37,8 @@ int main() { // provide the functions: initNew(...), surfaceCollision(...), // surfaceReflection(...). NumericType stickingProbability = 0.1; - auto particle = - std::make_unique>(stickingProbability, "flux"); + auto particle = std::make_unique>( + stickingProbability, "flux"); TraceDisk rayTracer; rayTracer.setGeometry(points, normals, gridDelta); diff --git a/examples/disk3D/disk3D.cpp b/examples/disk3D/disk3D.cpp index 8e5aeb4..2ebfc90 100644 --- a/examples/disk3D/disk3D.cpp +++ b/examples/disk3D/disk3D.cpp @@ -40,8 +40,8 @@ int main() { // provide the functions: initNew(...), surfaceCollision(...), // surfaceReflection(...). NumericType stickingProbability = 0.1; - auto particle = - std::make_unique>(stickingProbability, "flux"); + auto particle = std::make_unique>( + stickingProbability, "flux"); TraceDisk rayTracer; rayTracer.setGeometry(points, normals, gridDelta); diff --git a/examples/triangle2D/triangle2D.cpp b/examples/triangle2D/triangle2D.cpp index 4f7a62f..45cd968 100644 --- a/examples/triangle2D/triangle2D.cpp +++ b/examples/triangle2D/triangle2D.cpp @@ -24,8 +24,8 @@ int main() { tracer.setGeometry(lineMesh); NumericType stickingProbability = 0.1; - auto particle = - std::make_unique>(stickingProbability, "flux"); + auto particle = std::make_unique>( + stickingProbability, "flux"); tracer.setParticleType(particle); tracer.setNumberOfRaysPerPoint(5000); diff --git a/examples/triangle3D/triangle3D.cpp b/examples/triangle3D/triangle3D.cpp index f3046ae..0cec1aa 100644 --- a/examples/triangle3D/triangle3D.cpp +++ b/examples/triangle3D/triangle3D.cpp @@ -24,8 +24,8 @@ int main() { tracer.setGeometry(mesh); NumericType stickingProbability = 0.1; - auto particle = - std::make_unique>(stickingProbability, "flux"); + auto particle = std::make_unique>( + stickingProbability, "flux"); tracer.setParticleType(particle); tracer.setNumberOfRaysPerPoint(2000); diff --git a/gpu/CMakeLists.txt b/gpu/CMakeLists.txt index 3116769..097f02d 100644 --- a/gpu/CMakeLists.txt +++ b/gpu/CMakeLists.txt @@ -12,15 +12,15 @@ set(VIENNARAY_PIPELINE_DIR target_include_directories(${PROJECT_NAME} INTERFACE $) # Incldue directories for PTX and optix-ir compilation (force cache to update with new entries) -if (NOT VIENNARAY_PASSED_FIRST_CONFIGURE) - set(VIENNACORE_PTX_INCLUDE_DIRS - ${VIENNACORE_PTX_INCLUDE_DIRS} ${VIENNARAY_GPU_INCLUDE} - CACHE STRING "PTX include directories" FORCE) - if(VIENNARAY_GPU_DOUBLE_PRECISION) - set(VIENNACORE_PTX_DEFINES - ${VIENNACORE_PTX_DEFINES} "VIENNARAY_GPU_DOUBLE_PRECISION=1" - CACHE STRING "PTX compile definitions" FORCE) - endif() +if(NOT VIENNARAY_PASSED_FIRST_CONFIGURE) + set(VIENNACORE_PTX_INCLUDE_DIRS + ${VIENNACORE_PTX_INCLUDE_DIRS} ${VIENNARAY_GPU_INCLUDE} + CACHE STRING "PTX include directories" FORCE) + if(VIENNARAY_GPU_DOUBLE_PRECISION) + set(VIENNACORE_PTX_DEFINES + ${VIENNACORE_PTX_DEFINES} "VIENNARAY_GPU_DOUBLE_PRECISION=1" + CACHE STRING "PTX compile definitions" FORCE) + endif() endif() # Add the general pipelines and callable wrapper @@ -46,4 +46,6 @@ if(VIENNARAY_BUILD_TESTS) add_subdirectory(tests) endif() -set(VIENNARAY_PASSED_FIRST_CONFIGURE ON CACHE INTERNAL "Indicates whether the first configure has been passed.") +set(VIENNARAY_PASSED_FIRST_CONFIGURE + ON + CACHE INTERNAL "Indicates whether the first configure has been passed.") diff --git a/gpu/include/raygTrace.hpp b/gpu/include/raygTrace.hpp index 1f96af5..2bb4a86 100644 --- a/gpu/include/raygTrace.hpp +++ b/gpu/include/raygTrace.hpp @@ -551,7 +551,8 @@ template class Trace { VIENNACORE_LOG_WARNING("No callable file set."); return; } - std::string callable_path_str = callableFile_.string(); // explicit conversion + std::string callable_path_str = + callableFile_.string(); // explicit conversion auto callableInput = getInputData(callable_path_str.c_str(), inputSize); if (!callableInput) { VIENNACORE_LOG_ERROR("Callable file " + callableFile_.string() + diff --git a/include/viennaray/rayParticle.hpp b/include/viennaray/rayParticle.hpp index 8824f29..51a6b18 100644 --- a/include/viennaray/rayParticle.hpp +++ b/include/viennaray/rayParticle.hpp @@ -11,7 +11,7 @@ #define VIENNARAY_PARTICLE_STOP \ std::pair> { \ - NumericType(1), Vec3D{} \ + NumericType(1), Vec3D {} \ } namespace viennaray { diff --git a/tests/diskAreas/diskAreas.cpp b/tests/diskAreas/diskAreas.cpp index 75ff44a..459a8d3 100644 --- a/tests/diskAreas/diskAreas.cpp +++ b/tests/diskAreas/diskAreas.cpp @@ -41,8 +41,7 @@ int main() { std::array, 3> orthoBasis; auto raySource = std::make_unique>( boundingBox, NumericType(1), traceSettings, geometry.getNumPrimitives(), - false, - orthoBasis); + false, orthoBasis); geometry.computeDiskAreas(boundary); DiffuseParticle particle(1.0, "hitFlux"); diff --git a/tests/particle/particle.cpp b/tests/particle/particle.cpp index 8fe0a82..41293c0 100644 --- a/tests/particle/particle.cpp +++ b/tests/particle/particle.cpp @@ -28,8 +28,8 @@ template void RunTest() { { NumericType stickingProbability = 1.0; NumericType sourcePower = 50.; - auto particle = - std::make_unique>(stickingProbability, sourcePower, "test"); + auto particle = std::make_unique>( + stickingProbability, sourcePower, "test"); NumericType sourcePowerTest = particle->getSourceDistributionPower(); VC_TEST_ASSERT(sourcePowerTest == 100.); diff --git a/tests/rngSeed/rngSeed.cpp b/tests/rngSeed/rngSeed.cpp index 777651d..956ff69 100644 --- a/tests/rngSeed/rngSeed.cpp +++ b/tests/rngSeed/rngSeed.cpp @@ -16,8 +16,8 @@ int main() { rayInternal::createPlaneGrid(gridDelta, extent, {0, 1, 2}, points, normals); NumericType stickingProbability = 1.0; - auto particle = - std::make_unique>(stickingProbability, "hitFlux"); + auto particle = std::make_unique>( + stickingProbability, "hitFlux"); std::vector flux1, flux2; From d6dabe7f960596e71bc9959bbfbbb8c31d4e4e6e Mon Sep 17 00:00:00 2001 From: Tobias Reiter Date: Fri, 2 Jan 2026 21:00:48 +0100 Subject: [PATCH 08/11] Rename BoundaryCondition enum --- examples/disk2D/disk2D.cpp | 4 ++-- examples/disk3D/disk3D.cpp | 6 +++--- include/viennaray/rayBoundary.hpp | 18 +++++++++--------- include/viennaray/rayGeometryDisk.hpp | 12 ++++++++---- tests/boundaryHit/boundaryHit.cpp | 18 +++++++++--------- tests/boundaryHit2D/boundaryHit2D.cpp | 8 ++++---- tests/trace2D/trace2D.cpp | 4 ++-- tests/traceInterface/traceInterface.cpp | 6 +++--- 8 files changed, 40 insertions(+), 36 deletions(-) diff --git a/examples/disk2D/disk2D.cpp b/examples/disk2D/disk2D.cpp index b5579bb..48517f7 100644 --- a/examples/disk2D/disk2D.cpp +++ b/examples/disk2D/disk2D.cpp @@ -28,8 +28,8 @@ int main() { // however the boundary condition in direction of the tracing direction will // not be used. Possible choices are: PERIODIC, REFLECTIVE, IGNORE BoundaryCondition boundaryConds[D]; - boundaryConds[0] = BoundaryCondition::PERIODIC; // x - boundaryConds[1] = BoundaryCondition::PERIODIC; // y + boundaryConds[0] = BoundaryCondition::PERIODIC_BOUNDARY; // x + boundaryConds[1] = BoundaryCondition::PERIODIC_BOUNDARY; // y // ParticleType: The particle types provides the sticking probability and // the reflection process for each surface hit. This class can be user diff --git a/examples/disk3D/disk3D.cpp b/examples/disk3D/disk3D.cpp index 2ebfc90..ac19bb7 100644 --- a/examples/disk3D/disk3D.cpp +++ b/examples/disk3D/disk3D.cpp @@ -30,9 +30,9 @@ int main() { // however the boundary condition in direction of the tracing direction will // not be used. Possible choices are: PERIODIC, REFLECTIVE, IGNORE BoundaryCondition boundaryConds[D]; - boundaryConds[0] = BoundaryCondition::PERIODIC; // x - boundaryConds[1] = BoundaryCondition::PERIODIC; // y - boundaryConds[2] = BoundaryCondition::PERIODIC; // z + boundaryConds[0] = BoundaryCondition::PERIODIC_BOUNDARY; // x + boundaryConds[1] = BoundaryCondition::PERIODIC_BOUNDARY; // y + boundaryConds[2] = BoundaryCondition::PERIODIC_BOUNDARY; // z // ParticleType: The particle types provides the sticking probability and // the reflection process for each surface hit. This class can be user diff --git a/include/viennaray/rayBoundary.hpp b/include/viennaray/rayBoundary.hpp index 731d97e..b0e5605 100644 --- a/include/viennaray/rayBoundary.hpp +++ b/include/viennaray/rayBoundary.hpp @@ -8,9 +8,9 @@ namespace viennaray { using namespace viennacore; enum class BoundaryCondition : unsigned { - REFLECTIVE = 0, - PERIODIC = 1, - IGNORE = 2 + REFLECTIVE_BOUNDARY = 0, + PERIODIC_BOUNDARY = 1, + IGNORE_BOUNDARY = 2 }; template class Boundary { @@ -45,11 +45,11 @@ template class Boundary { if constexpr (D == 2) { assert((primID == 0 || primID == 1 || primID == 2 || primID == 3) && "Assumption"); - if (boundaryConds_[0] == BoundaryCondition::REFLECTIVE) { + if (boundaryConds_[0] == BoundaryCondition::REFLECTIVE_BOUNDARY) { reflectRay(rayHit); reflect = true; return; - } else if (boundaryConds_[0] == BoundaryCondition::PERIODIC) { + } else if (boundaryConds_[0] == BoundaryCondition::PERIODIC_BOUNDARY) { auto impactCoords = getNewOrigin(rayHit.ray); // periodically move ray origin if (primID == 0 || primID == 1) { @@ -71,12 +71,12 @@ template class Boundary { assert(false && "Correctness Assumption"); } else { if (primID <= 3) { - if (boundaryConds_[0] == BoundaryCondition::REFLECTIVE) { + if (boundaryConds_[0] == BoundaryCondition::REFLECTIVE_BOUNDARY) { // use specular reflection reflectRay(rayHit); reflect = true; return; - } else if (boundaryConds_[0] == BoundaryCondition::PERIODIC) { + } else if (boundaryConds_[0] == BoundaryCondition::PERIODIC_BOUNDARY) { auto impactCoords = getNewOrigin(rayHit.ray); // periodically move ray origin if (primID == 0 || primID == 1) { @@ -95,12 +95,12 @@ template class Boundary { return; } } else if (primID <= 7) { - if (boundaryConds_[1] == BoundaryCondition::REFLECTIVE) { + if (boundaryConds_[1] == BoundaryCondition::REFLECTIVE_BOUNDARY) { // use specular reflection reflectRay(rayHit); reflect = true; return; - } else if (boundaryConds_[1] == BoundaryCondition::PERIODIC) { + } else if (boundaryConds_[1] == BoundaryCondition::PERIODIC_BOUNDARY) { auto impactCoords = getNewOrigin(rayHit.ray); // periodically move ray origin if (primID == 4 || primID == 5) { diff --git a/include/viennaray/rayGeometryDisk.hpp b/include/viennaray/rayGeometryDisk.hpp index 87706eb..040a62d 100644 --- a/include/viennaray/rayGeometryDisk.hpp +++ b/include/viennaray/rayGeometryDisk.hpp @@ -279,8 +279,10 @@ class GeometryDisk : public Geometry { if constexpr (D == 3) { diskAreas_[idx] = disk[3] * disk[3] * M_PI; // full disk area - if (boundaryConds[boundaryDirs[0]] == BoundaryCondition::IGNORE && - boundaryConds[boundaryDirs[1]] == BoundaryCondition::IGNORE) { + if (boundaryConds[boundaryDirs[0]] == + BoundaryCondition::IGNORE_BOUNDARY && + boundaryConds[boundaryDirs[1]] == + BoundaryCondition::IGNORE_BOUNDARY) { // no boundaries continue; } @@ -316,7 +318,8 @@ class GeometryDisk : public Geometry { auto normal = getNormalRef(idx); // test min boundary - if ((boundaryConds[boundaryDirs[0]] != BoundaryCondition::IGNORE) && + if ((boundaryConds[boundaryDirs[0]] != + BoundaryCondition::IGNORE_BOUNDARY) && (std::abs(disk[boundaryDirs[0]] - bdBox[0][boundaryDirs[0]]) < disk[3])) { NumericType insideTest = @@ -332,7 +335,8 @@ class GeometryDisk : public Geometry { } // test max boundary - if ((boundaryConds[boundaryDirs[0]] != BoundaryCondition::IGNORE) && + if ((boundaryConds[boundaryDirs[0]] != + BoundaryCondition::IGNORE_BOUNDARY) && (std::abs(disk[boundaryDirs[0]] - bdBox[1][boundaryDirs[0]]) < disk[3])) { NumericType insideTest = diff --git a/tests/boundaryHit/boundaryHit.cpp b/tests/boundaryHit/boundaryHit.cpp index 3bd800f..8f32712 100644 --- a/tests/boundaryHit/boundaryHit.cpp +++ b/tests/boundaryHit/boundaryHit.cpp @@ -26,9 +26,9 @@ int main() { geometry.initGeometry(device, points, normals, gridDelta); BoundaryCondition boundCons[D]; { - boundCons[0] = BoundaryCondition::REFLECTIVE; - boundCons[1] = BoundaryCondition::PERIODIC; - boundCons[2] = BoundaryCondition::PERIODIC; + boundCons[0] = BoundaryCondition::REFLECTIVE_BOUNDARY; + boundCons[1] = BoundaryCondition::PERIODIC_BOUNDARY; + boundCons[2] = BoundaryCondition::PERIODIC_BOUNDARY; auto boundingBox = geometry.getBoundingBox(); auto traceSetting = rayInternal::getTraceSettings(TraceDirection::POS_Z); rayInternal::adjustBoundingBox( @@ -86,9 +86,9 @@ int main() { geometry.initGeometry(device, points, normals, gridDelta); BoundaryCondition boundCons[D]; { - boundCons[0] = BoundaryCondition::PERIODIC; - boundCons[1] = BoundaryCondition::PERIODIC; - boundCons[2] = BoundaryCondition::REFLECTIVE; + boundCons[0] = BoundaryCondition::PERIODIC_BOUNDARY; + boundCons[1] = BoundaryCondition::PERIODIC_BOUNDARY; + boundCons[2] = BoundaryCondition::REFLECTIVE_BOUNDARY; auto boundingBox = geometry.getBoundingBox(); auto traceSetting = rayInternal::getTraceSettings(TraceDirection::POS_Y); rayInternal::adjustBoundingBox( @@ -146,9 +146,9 @@ int main() { geometry.initGeometry(device, points, normals, gridDelta); BoundaryCondition boundCons[D]; { - boundCons[0] = BoundaryCondition::PERIODIC; - boundCons[1] = BoundaryCondition::PERIODIC; - boundCons[2] = BoundaryCondition::PERIODIC; + boundCons[0] = BoundaryCondition::PERIODIC_BOUNDARY; + boundCons[1] = BoundaryCondition::PERIODIC_BOUNDARY; + boundCons[2] = BoundaryCondition::PERIODIC_BOUNDARY; auto boundingBox = geometry.getBoundingBox(); auto traceSetting = rayInternal::getTraceSettings(TraceDirection::POS_Z); rayInternal::adjustBoundingBox( diff --git a/tests/boundaryHit2D/boundaryHit2D.cpp b/tests/boundaryHit2D/boundaryHit2D.cpp index d003ae1..2d214ec 100644 --- a/tests/boundaryHit2D/boundaryHit2D.cpp +++ b/tests/boundaryHit2D/boundaryHit2D.cpp @@ -30,7 +30,7 @@ int main() { GeometryDisk geometry; geometry.initGeometry(device, points, normals, gridDelta); - boundCons[1] = BoundaryCondition::REFLECTIVE; + boundCons[1] = BoundaryCondition::REFLECTIVE_BOUNDARY; { // build reflective boundary in y and z directions auto dir = TraceDirection::POS_X; @@ -79,7 +79,7 @@ int main() { VC_TEST_ASSERT_ISCLOSE(rayHit.ray.dir_z, direction[2], eps) } - boundCons[1] = BoundaryCondition::PERIODIC; + boundCons[1] = BoundaryCondition::PERIODIC_BOUNDARY; { // build periodic boundary in y and z directions auto dir = TraceDirection::POS_X; @@ -142,7 +142,7 @@ int main() { GeometryDisk geometry2; geometry2.initGeometry(device, points, normals, gridDelta); - boundCons[0] = BoundaryCondition::REFLECTIVE; + boundCons[0] = BoundaryCondition::REFLECTIVE_BOUNDARY; { // build periodic boundary in x and z directions auto dir = TraceDirection::POS_Y; @@ -191,7 +191,7 @@ int main() { VC_TEST_ASSERT_ISCLOSE(rayHit.ray.dir_z, direction[2], eps) } - boundCons[0] = BoundaryCondition::PERIODIC; + boundCons[0] = BoundaryCondition::PERIODIC_BOUNDARY; { // build periodic boundary in x and z directions auto dir = TraceDirection::POS_Y; diff --git a/tests/trace2D/trace2D.cpp b/tests/trace2D/trace2D.cpp index a0eb974..fdfa5fc 100644 --- a/tests/trace2D/trace2D.cpp +++ b/tests/trace2D/trace2D.cpp @@ -20,8 +20,8 @@ int main() { std::vector materialIds(points.size(), 0); BoundaryCondition boundaryConds[D]; - boundaryConds[0] = BoundaryCondition::REFLECTIVE; - boundaryConds[1] = BoundaryCondition::REFLECTIVE; + boundaryConds[0] = BoundaryCondition::REFLECTIVE_BOUNDARY; + boundaryConds[1] = BoundaryCondition::REFLECTIVE_BOUNDARY; auto particle = std::make_unique>( NumericType(1), "hitFlux"); diff --git a/tests/traceInterface/traceInterface.cpp b/tests/traceInterface/traceInterface.cpp index 8587ac9..eb8a15d 100644 --- a/tests/traceInterface/traceInterface.cpp +++ b/tests/traceInterface/traceInterface.cpp @@ -35,9 +35,9 @@ int main() { std::vector matIds(points.size(), 0); BoundaryCondition boundaryConds[D]; - boundaryConds[0] = BoundaryCondition::REFLECTIVE; - boundaryConds[1] = BoundaryCondition::REFLECTIVE; - boundaryConds[2] = BoundaryCondition::REFLECTIVE; + boundaryConds[0] = BoundaryCondition::REFLECTIVE_BOUNDARY; + boundaryConds[1] = BoundaryCondition::REFLECTIVE_BOUNDARY; + boundaryConds[2] = BoundaryCondition::REFLECTIVE_BOUNDARY; auto particle = std::make_unique>( NumericType(1), "hitFlux"); From 1d2b37d1109df69237871503997007b389b01bba Mon Sep 17 00:00:00 2001 From: Tobias Reiter Date: Mon, 5 Jan 2026 09:55:37 +0100 Subject: [PATCH 09/11] Fix tests --- cmake/generate_ptx.cmake | 151 ---------------------- cmake/gpuUtil.cmake | 6 - gpu/include/raygTrace.hpp | 2 - gpu/tests/boundaries/CMakeLists.txt | 4 +- gpu/tests/boundaries/boundaries.cpp | 2 +- gpu/tests/reflections/CMakeLists.txt | 4 +- gpu/tests/reflections/testReflections.cpp | 12 +- tests/particle/particle.cpp | 2 +- 8 files changed, 12 insertions(+), 171 deletions(-) delete mode 100644 cmake/generate_ptx.cmake delete mode 100644 cmake/gpuUtil.cmake diff --git a/cmake/generate_ptx.cmake b/cmake/generate_ptx.cmake deleted file mode 100644 index 3abd48b..0000000 --- a/cmake/generate_ptx.cmake +++ /dev/null @@ -1,151 +0,0 @@ -function(generate_pipeline target_name generated_files) - - # Separate the sources from the CMake and CUDA options fed to the macro. This code - # comes from the CUDA_COMPILE_PTX macro found in FindCUDA.cmake. - cuda_get_sources_and_options(cu_optix_source_files cmake_options options ${ARGN}) - - # Add the path to the OptiX headers to our include paths. - cuda_include_directories(${OptiX_INCLUDE_DIR}) - - # Include ViennaRay headers which are used in pipelines - cuda_include_directories(${VIENNARAY_GPU_INCLUDE}) - cuda_include_directories(${ViennaCore_SOURCE_DIR}/include/viennacore) - add_compile_definitions(VIENNACORE_COMPILE_GPU) - if(VIENNARAY_GPU_DOUBLE_PRECISION) - add_compile_definitions(VIENNARAY_GPU_DOUBLE_PRECISION) - endif() - - # Generate OptiX IR files if enabled - if(VIENNARAY_GENERATE_OPTIXIR) - cuda_wrap_srcs( - ${target_name} - OPTIXIR - generated_optixir_files - ${cu_optix_source_files} - ${cmake_options} - OPTIONS - ${options}) - list(APPEND generated_files_local ${generated_optixir_files}) - endif() - - # Generate PTX files if enabled - if(VIENNARAY_GENERATE_PTX) - cuda_wrap_srcs( - ${target_name} - PTX - generated_ptx_files - ${cu_optix_source_files} - ${cmake_options} - OPTIONS - ${options}) - list(APPEND generated_files_local ${generated_ptx_files}) - endif() - - set(${generated_files} - ${generated_files_local} - PARENT_SCOPE) -endfunction() - -function(generate_kernel generated_files) - - # Separate the sources from the CMake and CUDA options fed to the macro. This code - # comes from the CUDA_COMPILE_PTX macro found in FindCUDA.cmake. - cuda_get_sources_and_options(cu_source_files cmake_options options ${ARGN}) - - cuda_include_directories(${ViennaCore_SOURCE_DIR}/include/viennacore) - cuda_include_directories(${VIENNARAY_GPU_INCLUDE}) - cuda_include_directories(${OptiX_INCLUDE_DIR}) - add_compile_definitions(VIENNACORE_COMPILE_GPU) - if(VIENNARAY_GPU_DOUBLE_PRECISION) - add_compile_definitions(VIENNARAY_GPU_DOUBLE_PRECISION) - endif() - - cuda_compile_ptx(generated_ptx_files ${cu_source_files} ${cmake_options} ${options}) - - set(${generated_files} - ${generated_ptx_files} - PARENT_SCOPE) -endfunction() - -# In CMake, functions have their own scope, whereas macros use the scope of the caller. -function(add_gpu_executable target_name_base target_name_var) - set(target_name ${target_name_base}) - set(${target_name_var} - ${target_name} - PARENT_SCOPE) - - # Separate the sources from the CMake and CUDA options fed to the macro. This code - # comes from the CUDA_COMPILE_PTX macro found in FindCUDA.cmake. We are copying the - # code here, so that we can use our own name for the target. target_name is used in the - # creation of the output file names, and we want this to be unique for each target in - # the SDK. - cuda_get_sources_and_options(source_files cmake_options options ${ARGN}) - - # Isolate OBJ target files. NVCC should only process these files and leave PTX targets for NVRTC - set(cu_obj_source_files) - set(cu_optix_source_files) - foreach(file ${source_files}) - get_filename_component(_file_extension ${file} EXT) - if(${_file_extension} MATCHES "cu") - list(APPEND cu_optix_source_files ${file}) - endif() - endforeach() - - # Add the general pipelines and callable wrapper - list(APPEND cu_optix_source_files ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineDisk.cu) - list(APPEND cu_optix_source_files ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineTriangle.cu) - list(APPEND cu_optix_source_files ${VIENNARAY_PIPELINE_DIR}/GeneralPipelineLine.cu) - list(APPEND cu_optix_source_files ${VIENNARAY_PIPELINE_DIR}/CallableWrapper.cu) - - # Add the path to the OptiX headers to our include paths. - cuda_include_directories(${OptiX_INCLUDE_DIR}) - - # Include ViennaRay headers which are used in pipelines - cuda_include_directories(${VIENNARAY_GPU_INCLUDE}) - cuda_include_directories(${ViennaCore_SOURCE_DIR}/include/viennacore) - add_compile_definitions(VIENNACORE_COMPILE_GPU) - if(VIENNARAY_GPU_DOUBLE_PRECISION) - add_compile_definitions(VIENNARAY_GPU_DOUBLE_PRECISION) - endif() - - # Create CUDA kernels - cuda_wrap_srcs( - ${target_name} - PTX - generated_files - ${VIENNARAY_CUDA_KERNELS} - ${cmake_options} - OPTIONS - ${options}) - - # Create the rules to build the PTX and/or OPTIX files. - if(VIENNARAY_GENERATE_OPTIXIR) - cuda_wrap_srcs( - ${target_name} - OPTIXIR - generated_optixir_files - ${cu_optix_source_files} - ${cmake_options} - OPTIONS - ${options}) - list(APPEND generated_files ${generated_optixir_files}) - endif() - if(VIENNARAY_GENERATE_PTX) - cuda_wrap_srcs( - ${target_name} - PTX - generated_ptx_files - ${cu_optix_source_files} - ${cmake_options} - OPTIONS - ${options}) - list(APPEND generated_files ${generated_ptx_files}) - endif() - - # Here is where we create the rule to make the executable. We define a target name and - # list all the source files used to create the target. In addition we also pass along - # the cmake_options parsed out of the arguments. - message(STATUS "Adding target: ${target_name}") - add_executable(${target_name} ${source_files} ${generated_files} ${cmake_options}) - target_link_libraries(${target_name} PRIVATE ViennaRay) -endfunction() diff --git a/cmake/gpuUtil.cmake b/cmake/gpuUtil.cmake deleted file mode 100644 index 78fb492..0000000 --- a/cmake/gpuUtil.cmake +++ /dev/null @@ -1,6 +0,0 @@ -# In CMake, functions have their own scope, whereas macros use the scope of the caller. -function(add_gpu_deps target_name) - - add_dependencies(${target_name} disk_pipeline triangle_pipeline line_pipeline callable_wrapper - norm_kernels) -endfunction() diff --git a/gpu/include/raygTrace.hpp b/gpu/include/raygTrace.hpp index 2bb4a86..5fb345b 100644 --- a/gpu/include/raygTrace.hpp +++ b/gpu/include/raygTrace.hpp @@ -727,8 +727,6 @@ template class Trace { shaderBindingTable_.callablesRecordStrideInBytes = sizeof(CallableRecord); shaderBindingTable_.callablesRecordCount = static_cast(directCallablePGs_.size()); - } else { - assert(false && "No direct callables found."); } } diff --git a/gpu/tests/boundaries/CMakeLists.txt b/gpu/tests/boundaries/CMakeLists.txt index d398ebd..05795db 100644 --- a/gpu/tests/boundaries/CMakeLists.txt +++ b/gpu/tests/boundaries/CMakeLists.txt @@ -1,8 +1,8 @@ project(boundaries LANGUAGES CXX) -viennacore_add_optixir(viennaray_test_pipeline ${CMAKE_CURRENT_SOURCE_DIR}/TestPipelineTriangle.cu) +viennacore_add_optixir(TestPipelineTriangle ${CMAKE_CURRENT_SOURCE_DIR}/TestPipelineTriangle.cu) add_executable(${PROJECT_NAME} ${PROJECT_NAME}.cpp) -add_dependencies(${PROJECT_NAME} viennaray_test_pipeline) +add_dependencies(${PROJECT_NAME} TestPipelineTriangle) target_link_libraries(${PROJECT_NAME} PRIVATE ViennaRay) add_dependencies(ViennaRay-GPU_Tests ${PROJECT_NAME}) diff --git a/gpu/tests/boundaries/boundaries.cpp b/gpu/tests/boundaries/boundaries.cpp index 2ccdf38..1646e23 100644 --- a/gpu/tests/boundaries/boundaries.cpp +++ b/gpu/tests/boundaries/boundaries.cpp @@ -43,7 +43,7 @@ template TriangleMesh createGeometry() { int main() { - auto context = DeviceContext::createContext("../../../lib/ptx", 0); + auto context = DeviceContext::createContext(); gpu::Particle particle; particle.name = "Particle"; diff --git a/gpu/tests/reflections/CMakeLists.txt b/gpu/tests/reflections/CMakeLists.txt index ba38fdc..3b5ec67 100644 --- a/gpu/tests/reflections/CMakeLists.txt +++ b/gpu/tests/reflections/CMakeLists.txt @@ -1,10 +1,10 @@ project(testReflections LANGUAGES CXX) # wrap CUDA kernel only -viennacore_add_ptx(viennaray_test_reflections ${CMAKE_CURRENT_SOURCE_DIR}/testReflections.cu) +viennacore_add_ptx(testReflectionsKernel ${CMAKE_CURRENT_SOURCE_DIR}/testReflections.cu) add_executable(${PROJECT_NAME} "${PROJECT_NAME}.cpp") -add_dependencies(${PROJECT_NAME} viennaray_test_reflections) +add_dependencies(${PROJECT_NAME} testReflectionsKernel) target_link_libraries(${PROJECT_NAME} PRIVATE ViennaRay) add_dependencies(ViennaRay-GPU_Tests ${PROJECT_NAME}) diff --git a/gpu/tests/reflections/testReflections.cpp b/gpu/tests/reflections/testReflections.cpp index efa441b..c6d46ca 100644 --- a/gpu/tests/reflections/testReflections.cpp +++ b/gpu/tests/reflections/testReflections.cpp @@ -15,10 +15,9 @@ using namespace viennaray; int main() { Logger::setLogLevel(LogLevel::DEBUG); - DeviceContext context; - context.create("../../../lib/ptx"); // relative to build directory - const std::string moduleName = "testReflections.ptx"; - context.addModule(moduleName); + auto context = DeviceContext::createContext(); + const std::string moduleName = "testReflectionsKernel.ptx"; + context->addModule(moduleName); { unsigned numResults = 10000; CudaBuffer resultBuffer; @@ -31,7 +30,7 @@ int main() { CUdeviceptr d_data = resultBuffer.dPointer(); void *kernel_args[] = {&inDir, &normal, &d_data, &numResults}; - LaunchKernel::launch(moduleName, "test_diffuse", kernel_args, context); + LaunchKernel::launch(moduleName, "test_diffuse", kernel_args, *context); resultBuffer.download(results.data(), numResults); @@ -65,7 +64,8 @@ int main() { CUdeviceptr d_data = resultBuffer.dPointer(); void *kernel_args[] = {&inDir, &normal, &coneAngle, &d_data, &numResults}; - LaunchKernel::launch(moduleName, "test_coned_cosine", kernel_args, context); + LaunchKernel::launch(moduleName, "test_coned_cosine", kernel_args, + *context); resultBuffer.download(results.data(), numResults); #ifdef WRITE_TO_FILE diff --git a/tests/particle/particle.cpp b/tests/particle/particle.cpp index 41293c0..62239b5 100644 --- a/tests/particle/particle.cpp +++ b/tests/particle/particle.cpp @@ -32,7 +32,7 @@ template void RunTest() { stickingProbability, sourcePower, "test"); NumericType sourcePowerTest = particle->getSourceDistributionPower(); - VC_TEST_ASSERT(sourcePowerTest == 100.); + VC_TEST_ASSERT(sourcePowerTest == 50.); auto labels = particle->getLocalDataLabels(); VC_TEST_ASSERT(labels.size() == 1); From ab8fc24e9ce70db021de6787129fc839eaa8633d Mon Sep 17 00:00:00 2001 From: Tobias Reiter Date: Mon, 5 Jan 2026 11:27:15 +0100 Subject: [PATCH 10/11] Move GPU includes into include folder --- CMakeLists.txt | 6 +++--- gpu/CMakeLists.txt | 16 +++++++------- gpu/examples/trenchDisks.cpp | 2 +- gpu/examples/trenchLines.cpp | 4 ++-- gpu/examples/trenchTriangles.cpp | 2 +- gpu/pipelines/GeneralPipelineDisk.cu | 18 ++++++++-------- gpu/pipelines/GeneralPipelineLine.cu | 18 ++++++++-------- gpu/pipelines/GeneralPipelineTriangle.cu | 21 ++++++++++--------- gpu/pipelines/Particle.cuh | 4 ++-- gpu/tests/boundaries/TestPipelineTriangle.cu | 18 ++++++++-------- gpu/tests/boundaries/boundaries.cpp | 2 +- gpu/tests/deviceStructs/deviceStructs.cpp | 4 ++-- gpu/tests/reflections/testReflections.cu | 4 ++-- gpu/tests/rngSeed/rngSeed.cpp | 15 +++++++------ gpu/tests/tracer/tracer.cpp | 6 +++--- .../viennaray/gpu}/raygBoundary.hpp | 0 .../viennaray/gpu}/raygCallableConfig.hpp | 0 .../viennaray/gpu}/raygDiskGeometry.hpp | 0 .../viennaray/gpu}/raygLaunchParams.hpp | 0 .../viennaray/gpu}/raygLineGeometry.hpp | 0 .../viennaray/gpu}/raygPerRayData.hpp | 0 .../viennaray/gpu}/raygReflection.hpp | 0 .../viennaray/gpu}/raygSBTRecords.hpp | 0 .../viennaray/gpu}/raygSource.hpp | 0 .../viennaray/gpu}/raygTrace.hpp | 0 .../viennaray/gpu}/raygTraceDisk.hpp | 0 .../viennaray/gpu}/raygTraceLine.hpp | 0 .../viennaray/gpu}/raygTraceTriangle.hpp | 0 .../viennaray/gpu}/raygTriangleGeometry.hpp | 0 29 files changed, 69 insertions(+), 71 deletions(-) rename {gpu/include => include/viennaray/gpu}/raygBoundary.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygCallableConfig.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygDiskGeometry.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygLaunchParams.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygLineGeometry.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygPerRayData.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygReflection.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygSBTRecords.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygSource.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygTrace.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygTraceDisk.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygTraceLine.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygTraceTriangle.hpp (100%) rename {gpu/include => include/viennaray/gpu}/raygTriangleGeometry.hpp (100%) diff --git a/CMakeLists.txt b/CMakeLists.txt index be4f592..1fb52ce 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -2,7 +2,7 @@ cmake_minimum_required(VERSION 3.20 FATAL_ERROR) project( ViennaRay LANGUAGES CXX - VERSION 3.9.1) + VERSION 3.10.0) # -------------------------------------------------------------------------------------------------------- # Library switches @@ -74,6 +74,7 @@ if(VIENNARAY_GPU_DOUBLE_PRECISION) endif() if(VIENNARAY_PRINT_PROGRESS) + message(STATUS "[ViennaRay] Enabling progress printing") target_compile_definitions(${PROJECT_NAME} INTERFACE VIENNARAY_PRINT_PROGRESS) endif() @@ -150,8 +151,7 @@ endif() # Setup GPU # -------------------------------------------------------------------------------------------------------- -# If CUDA or OptiX is not found in ViennaCore, VIENNACORE_PTX_DIR is not set -if(VIENNARAY_USE_GPU AND VIENNACORE_PTX_DIR) +if(VIENNARAY_USE_GPU) message(STATUS "[ViennaRay] Enabling GPU support") add_subdirectory(gpu) endif() diff --git a/gpu/CMakeLists.txt b/gpu/CMakeLists.txt index 097f02d..880d9f4 100644 --- a/gpu/CMakeLists.txt +++ b/gpu/CMakeLists.txt @@ -1,7 +1,4 @@ #### Set variables -set(VIENNARAY_GPU_INCLUDE - "${PROJECT_SOURCE_DIR}/gpu/include" - CACHE STRING "ViennaRay GPU headers.") set(VIENNARAY_CUDA_KERNELS "${PROJECT_SOURCE_DIR}/gpu/kernels/normKernels.cu" CACHE STRING "ViennaRay CUDA kernel source files.") @@ -9,16 +6,14 @@ set(VIENNARAY_PIPELINE_DIR "${PROJECT_SOURCE_DIR}/gpu/pipelines" CACHE STRING "ViennaRay pipeline directory.") -target_include_directories(${PROJECT_NAME} INTERFACE $) - # Incldue directories for PTX and optix-ir compilation (force cache to update with new entries) if(NOT VIENNARAY_PASSED_FIRST_CONFIGURE) - set(VIENNACORE_PTX_INCLUDE_DIRS - ${VIENNACORE_PTX_INCLUDE_DIRS} ${VIENNARAY_GPU_INCLUDE} + set(VIENNACORE_NVCC_INCLUDE_DIRS + ${VIENNACORE_NVCC_INCLUDE_DIRS} ${PROJECT_SOURCE_DIR}/include/viennaray/gpu CACHE STRING "PTX include directories" FORCE) if(VIENNARAY_GPU_DOUBLE_PRECISION) - set(VIENNACORE_PTX_DEFINES - ${VIENNACORE_PTX_DEFINES} "VIENNARAY_GPU_DOUBLE_PRECISION=1" + set(VIENNACORE_NVCC_DEFINES + ${VIENNACORE_NVCC_DEFINES} "VIENNARAY_GPU_DOUBLE_PRECISION=1" CACHE STRING "PTX compile definitions" FORCE) endif() endif() @@ -32,9 +27,12 @@ viennacore_add_optixir(ViennaRayCallableWrapper ${VIENNARAY_PIPELINE_DIR}/Callab # Add the norm kernels viennacore_add_ptx(normKernels ${VIENNARAY_CUDA_KERNELS}) +# CMake target to collect all GPU dependencies for tests and examples set(VIENNARAY_GPU_DEPENDENCIES GeneralPipelineDisk GeneralPipelineTriangle GeneralPipelineLine normKernels CACHE STRING "ViennaRay GPU dependencies.") +# Install PTX files +install(DIRECTORY "${VIENNACORE_NVCC_PTX_DIR}/" DESTINATION "lib/ptx") if(VIENNARAY_BUILD_EXAMPLES) message(STATUS "[ViennaRay] Adding GPU Examples") diff --git a/gpu/examples/trenchDisks.cpp b/gpu/examples/trenchDisks.cpp index a11c1ce..214e3ac 100644 --- a/gpu/examples/trenchDisks.cpp +++ b/gpu/examples/trenchDisks.cpp @@ -1,4 +1,4 @@ -#include +#include #include diff --git a/gpu/examples/trenchLines.cpp b/gpu/examples/trenchLines.cpp index 0f805e4..c38adb5 100644 --- a/gpu/examples/trenchLines.cpp +++ b/gpu/examples/trenchLines.cpp @@ -1,5 +1,5 @@ -#include -#include +#include +#include #include diff --git a/gpu/examples/trenchTriangles.cpp b/gpu/examples/trenchTriangles.cpp index 97b338e..63b25a5 100644 --- a/gpu/examples/trenchTriangles.cpp +++ b/gpu/examples/trenchTriangles.cpp @@ -1,4 +1,4 @@ -#include "raygTraceTriangle.hpp" +#include "gpu/raygTraceTriangle.hpp" #include #include diff --git a/gpu/pipelines/GeneralPipelineDisk.cu b/gpu/pipelines/GeneralPipelineDisk.cu index f5a3d1e..776da5c 100644 --- a/gpu/pipelines/GeneralPipelineDisk.cu +++ b/gpu/pipelines/GeneralPipelineDisk.cu @@ -4,15 +4,15 @@ #define __CUDACC__ #endif -#include -#include -#include -#include -#include -#include -#include - -#include +#include "raygBoundary.hpp" +#include "raygCallableConfig.hpp" +#include "raygLaunchParams.hpp" +#include "raygPerRayData.hpp" +#include "raygReflection.hpp" +#include "raygSBTRecords.hpp" +#include "raygSource.hpp" + +#include "vcContext.hpp" using namespace viennaray::gpu; diff --git a/gpu/pipelines/GeneralPipelineLine.cu b/gpu/pipelines/GeneralPipelineLine.cu index ddb97db..0e0d258 100644 --- a/gpu/pipelines/GeneralPipelineLine.cu +++ b/gpu/pipelines/GeneralPipelineLine.cu @@ -4,15 +4,15 @@ #define __CUDACC__ #endif -#include -#include -#include -#include -#include -#include -#include - -#include +#include "raygBoundary.hpp" +#include "raygCallableConfig.hpp" +#include "raygLaunchParams.hpp" +#include "raygPerRayData.hpp" +#include "raygReflection.hpp" +#include "raygSBTRecords.hpp" +#include "raygSource.hpp" + +#include "vcContext.hpp" using namespace viennaray::gpu; diff --git a/gpu/pipelines/GeneralPipelineTriangle.cu b/gpu/pipelines/GeneralPipelineTriangle.cu index 7e293ff..ebee29d 100644 --- a/gpu/pipelines/GeneralPipelineTriangle.cu +++ b/gpu/pipelines/GeneralPipelineTriangle.cu @@ -4,15 +4,15 @@ #define __CUDACC__ #endif -#include -#include -#include -#include -#include -#include -#include +#include "raygBoundary.hpp" +#include "raygCallableConfig.hpp" +#include "raygLaunchParams.hpp" +#include "raygPerRayData.hpp" +#include "raygReflection.hpp" +#include "raygSBTRecords.hpp" +#include "raygSource.hpp" -#include +#include "vcContext.hpp" // #define COUNT_RAYS @@ -104,8 +104,9 @@ extern "C" __global__ void __raygen__() { hint |= (1 << 0); } if (optixHitObjectIsHit()) { - const HitSBTDataTriangle *hitData = reinterpret_cast( - optixHitObjectGetSbtDataPointer()); + const HitSBTDataTriangle *hitData = + reinterpret_cast( + optixHitObjectGetSbtDataPointer()); hint |= hitData->base.isBoundary << 1; } optixReorder(hint, hintBitLength); diff --git a/gpu/pipelines/Particle.cuh b/gpu/pipelines/Particle.cuh index ab9705f..7af9e64 100644 --- a/gpu/pipelines/Particle.cuh +++ b/gpu/pipelines/Particle.cuh @@ -3,8 +3,8 @@ #include #include -#include -#include +#include "raygLaunchParams.hpp" +#include "raygReflection.hpp" extern "C" __constant__ viennaray::gpu::LaunchParams launchParams; diff --git a/gpu/tests/boundaries/TestPipelineTriangle.cu b/gpu/tests/boundaries/TestPipelineTriangle.cu index 2eb5f3d..aa6ae55 100644 --- a/gpu/tests/boundaries/TestPipelineTriangle.cu +++ b/gpu/tests/boundaries/TestPipelineTriangle.cu @@ -4,15 +4,15 @@ #define __CUDACC__ #endif -#include -#include -#include -#include -#include -#include -#include - -#include +#include "raygBoundary.hpp" +#include "raygCallableConfig.hpp" +#include "raygLaunchParams.hpp" +#include "raygPerRayData.hpp" +#include "raygReflection.hpp" +#include "raygSBTRecords.hpp" +#include "raygSource.hpp" + +#include "vcContext.hpp" // #define COUNT_RAYS diff --git a/gpu/tests/boundaries/boundaries.cpp b/gpu/tests/boundaries/boundaries.cpp index 1646e23..3194186 100644 --- a/gpu/tests/boundaries/boundaries.cpp +++ b/gpu/tests/boundaries/boundaries.cpp @@ -1,5 +1,5 @@ +#include #include -#include #include diff --git a/gpu/tests/deviceStructs/deviceStructs.cpp b/gpu/tests/deviceStructs/deviceStructs.cpp index 60a80b8..ca9c51a 100644 --- a/gpu/tests/deviceStructs/deviceStructs.cpp +++ b/gpu/tests/deviceStructs/deviceStructs.cpp @@ -1,5 +1,5 @@ -#include -#include +#include +#include #include diff --git a/gpu/tests/reflections/testReflections.cu b/gpu/tests/reflections/testReflections.cu index 813166b..7198e27 100644 --- a/gpu/tests/reflections/testReflections.cu +++ b/gpu/tests/reflections/testReflections.cu @@ -1,8 +1,8 @@ #include #include -#include -#include +#include "raygReflection.hpp" +#include "vcVectorType.hpp" #define VIENNARAY_TEST diff --git a/gpu/tests/rngSeed/rngSeed.cpp b/gpu/tests/rngSeed/rngSeed.cpp index 7c9ce98..413d4f2 100644 --- a/gpu/tests/rngSeed/rngSeed.cpp +++ b/gpu/tests/rngSeed/rngSeed.cpp @@ -1,6 +1,6 @@ +#include +#include #include -#include -#include #include using namespace viennaray; @@ -9,8 +9,7 @@ int main() { constexpr int D = 3; using NumericType = float; - auto context = DeviceContext::createContext("../../../lib/ptx", - 0); // relative to build directory + auto context = DeviceContext::createContext(); gpu::Particle particle; particle.name = "Particle"; @@ -35,7 +34,7 @@ int main() { { gpu::TraceDisk rayTracer; rayTracer.insertNextParticle(particle); - rayTracer.setCallables("CallableWrapper", context->modulePath); + rayTracer.setCallables("ViennaRayCallableWrapper", context->modulePath); rayTracer.setParticleCallableMap({pMap, cMap}); rayTracer.setGeometry(mesh); rayTracer.setNumberOfRaysPerPoint(100); @@ -50,7 +49,7 @@ int main() { { gpu::TraceDisk rayTracer; rayTracer.insertNextParticle(particle); - rayTracer.setCallables("CallableWrapper", context->modulePath); + rayTracer.setCallables("ViennaRayCallableWrapper", context->modulePath); rayTracer.setParticleCallableMap({pMap, cMap}); rayTracer.setGeometry(mesh); rayTracer.setNumberOfRaysPerPoint(100); @@ -83,7 +82,7 @@ int main() { { gpu::TraceTriangle rayTracer; rayTracer.insertNextParticle(particle); - rayTracer.setCallables("CallableWrapper", context->modulePath); + rayTracer.setCallables("ViennaRayCallableWrapper", context->modulePath); rayTracer.setParticleCallableMap({pMap, cMap}); rayTracer.setGeometry(mesh); rayTracer.setNumberOfRaysPerPoint(100); @@ -98,7 +97,7 @@ int main() { { gpu::TraceTriangle rayTracer; rayTracer.insertNextParticle(particle); - rayTracer.setCallables("CallableWrapper", context->modulePath); + rayTracer.setCallables("ViennaRayCallableWrapper", context->modulePath); rayTracer.setParticleCallableMap({pMap, cMap}); rayTracer.setGeometry(mesh); rayTracer.setNumberOfRaysPerPoint(100); diff --git a/gpu/tests/tracer/tracer.cpp b/gpu/tests/tracer/tracer.cpp index f544cae..5e14483 100644 --- a/gpu/tests/tracer/tracer.cpp +++ b/gpu/tests/tracer/tracer.cpp @@ -1,6 +1,6 @@ -#include -#include -#include +#include +#include +#include #include diff --git a/gpu/include/raygBoundary.hpp b/include/viennaray/gpu/raygBoundary.hpp similarity index 100% rename from gpu/include/raygBoundary.hpp rename to include/viennaray/gpu/raygBoundary.hpp diff --git a/gpu/include/raygCallableConfig.hpp b/include/viennaray/gpu/raygCallableConfig.hpp similarity index 100% rename from gpu/include/raygCallableConfig.hpp rename to include/viennaray/gpu/raygCallableConfig.hpp diff --git a/gpu/include/raygDiskGeometry.hpp b/include/viennaray/gpu/raygDiskGeometry.hpp similarity index 100% rename from gpu/include/raygDiskGeometry.hpp rename to include/viennaray/gpu/raygDiskGeometry.hpp diff --git a/gpu/include/raygLaunchParams.hpp b/include/viennaray/gpu/raygLaunchParams.hpp similarity index 100% rename from gpu/include/raygLaunchParams.hpp rename to include/viennaray/gpu/raygLaunchParams.hpp diff --git a/gpu/include/raygLineGeometry.hpp b/include/viennaray/gpu/raygLineGeometry.hpp similarity index 100% rename from gpu/include/raygLineGeometry.hpp rename to include/viennaray/gpu/raygLineGeometry.hpp diff --git a/gpu/include/raygPerRayData.hpp b/include/viennaray/gpu/raygPerRayData.hpp similarity index 100% rename from gpu/include/raygPerRayData.hpp rename to include/viennaray/gpu/raygPerRayData.hpp diff --git a/gpu/include/raygReflection.hpp b/include/viennaray/gpu/raygReflection.hpp similarity index 100% rename from gpu/include/raygReflection.hpp rename to include/viennaray/gpu/raygReflection.hpp diff --git a/gpu/include/raygSBTRecords.hpp b/include/viennaray/gpu/raygSBTRecords.hpp similarity index 100% rename from gpu/include/raygSBTRecords.hpp rename to include/viennaray/gpu/raygSBTRecords.hpp diff --git a/gpu/include/raygSource.hpp b/include/viennaray/gpu/raygSource.hpp similarity index 100% rename from gpu/include/raygSource.hpp rename to include/viennaray/gpu/raygSource.hpp diff --git a/gpu/include/raygTrace.hpp b/include/viennaray/gpu/raygTrace.hpp similarity index 100% rename from gpu/include/raygTrace.hpp rename to include/viennaray/gpu/raygTrace.hpp diff --git a/gpu/include/raygTraceDisk.hpp b/include/viennaray/gpu/raygTraceDisk.hpp similarity index 100% rename from gpu/include/raygTraceDisk.hpp rename to include/viennaray/gpu/raygTraceDisk.hpp diff --git a/gpu/include/raygTraceLine.hpp b/include/viennaray/gpu/raygTraceLine.hpp similarity index 100% rename from gpu/include/raygTraceLine.hpp rename to include/viennaray/gpu/raygTraceLine.hpp diff --git a/gpu/include/raygTraceTriangle.hpp b/include/viennaray/gpu/raygTraceTriangle.hpp similarity index 100% rename from gpu/include/raygTraceTriangle.hpp rename to include/viennaray/gpu/raygTraceTriangle.hpp diff --git a/gpu/include/raygTriangleGeometry.hpp b/include/viennaray/gpu/raygTriangleGeometry.hpp similarity index 100% rename from gpu/include/raygTriangleGeometry.hpp rename to include/viennaray/gpu/raygTriangleGeometry.hpp From b225f238972be98379559b9fd27aebf6a9c3ab3a Mon Sep 17 00:00:00 2001 From: Tobias Reiter Date: Mon, 5 Jan 2026 20:00:36 +0100 Subject: [PATCH 11/11] Bump version --- CMakeLists.txt | 3 +-- README.md | 2 +- 2 files changed, 2 insertions(+), 3 deletions(-) diff --git a/CMakeLists.txt b/CMakeLists.txt index 1fb52ce..a79f2aa 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -99,8 +99,7 @@ include("cmake/cpm.cmake") CPMAddPackage( NAME ViennaCore - VERSION 1.8.0 - GIT_TAG msvc + VERSION 1.9.0 GIT_REPOSITORY "https://github.com/ViennaTools/ViennaCore" OPTIONS "VIENNACORE_USE_GPU ${VIENNARAY_USE_GPU}") diff --git a/README.md b/README.md index 4952224..ffb2ba2 100644 --- a/README.md +++ b/README.md @@ -63,7 +63,7 @@ We recommend using [CPM.cmake](https://github.com/cpm-cmake/CPM.cmake) to consum * Installation with CPM ```cmake - CPMAddPackage("gh:viennatools/viennaray@3.8.5") # Use the latest release version + CPMAddPackage("gh:viennatools/viennaray@3.10.0") # Use the latest release version ``` * With a local installation