BufferIndexing.h

Go to the documentation of this file.

 
#ifndef HELIOS_BUFFER_INDEXING_H
#define HELIOS_BUFFER_INDEXING_H
 
#include <cstddef>
#include <climits>
 
// Support both CUDA and CPU compilation
#ifdef __CUDACC__
#define HELIOS_HOST_DEVICE __host__ __device__
#else
#define HELIOS_HOST_DEVICE
#endif
 
class BufferIndexer2D {
private:
    size_t dim1_size;
 
public:
    HELIOS_HOST_DEVICE
    explicit BufferIndexer2D(size_t dim0_size, size_t dim1_size) : dim1_size(dim1_size) {
        // dim0_size parameter exists for self-documentation and future validation
        (void) dim0_size;
    }
 
    HELIOS_HOST_DEVICE inline size_t operator()(size_t i, size_t j) const {
        return i * dim1_size + j;
    }
 
    HELIOS_HOST_DEVICE inline size_t getDim1Size() const {
        return dim1_size;
    }
};
 
class BufferIndexer3D {
private:
    size_t dim1_size;
    size_t dim2_size;
    size_t dim1_dim2_product; // Precomputed for efficiency
 
public:
    HELIOS_HOST_DEVICE
    explicit BufferIndexer3D(size_t dim0_size, size_t dim1_size, size_t dim2_size) : dim1_size(dim1_size), dim2_size(dim2_size), dim1_dim2_product(dim1_size * dim2_size) {
        (void) dim0_size;
    }
 
    HELIOS_HOST_DEVICE inline size_t operator()(size_t i, size_t j, size_t k) const {
        return i * dim1_dim2_product + j * dim2_size + k;
    }
 
    HELIOS_HOST_DEVICE inline size_t getDim1Size() const {
        return dim1_size;
    }
 
    HELIOS_HOST_DEVICE inline size_t getDim2Size() const {
        return dim2_size;
    }
};
 
class BufferIndexer4D {
private:
    size_t dim1_size;
    size_t dim2_size;
    size_t dim3_size;
    size_t dim1_dim2_dim3_product; // dim1 * dim2 * dim3
    size_t dim2_dim3_product; // dim2 * dim3
 
public:
    HELIOS_HOST_DEVICE
    explicit BufferIndexer4D(size_t dim0_size, size_t dim1_size, size_t dim2_size, size_t dim3_size) :
        dim1_size(dim1_size), dim2_size(dim2_size), dim3_size(dim3_size), dim1_dim2_dim3_product(dim1_size * dim2_size * dim3_size), dim2_dim3_product(dim2_size * dim3_size) {
        (void) dim0_size;
    }
 
    HELIOS_HOST_DEVICE inline size_t operator()(size_t i, size_t j, size_t k, size_t l) const {
        return i * dim1_dim2_dim3_product + j * dim2_dim3_product + k * dim3_size + l;
    }
 
    HELIOS_HOST_DEVICE inline size_t getDim1Size() const {
        return dim1_size;
    }
 
    HELIOS_HOST_DEVICE inline size_t getDim2Size() const {
        return dim2_size;
    }
 
    HELIOS_HOST_DEVICE inline size_t getDim3Size() const {
        return dim3_size;
    }
};
 
// ========== Named Type Aliases for Self-Documenting Code ==========
 
using RadiationBufferIndexer = BufferIndexer2D;
 
using SourceFluxIndexer = BufferIndexer2D;
 
using MaterialPropertyIndexer = BufferIndexer3D;
 
using SourceCameraFluxIndexer = BufferIndexer3D;
 
using CameraMaterialIndexer = BufferIndexer4D;
 
using SpecularRadiationIndexer = BufferIndexer4D;
 
// ========== UUID / Position Lookup Helper ==========
 
class UUIDLookupHelper {
private:
    const uint *primitive_positions_; 
    size_t primitive_count_; 
 
public:
    HELIOS_HOST_DEVICE
    UUIDLookupHelper(const uint *prim_positions, size_t prim_count) : primitive_positions_(prim_positions), primitive_count_(prim_count) {
    }
 
    HELIOS_HOST_DEVICE inline bool toPosition(uint UUID, uint &position) const {
        // Lookup in sparse table - returns UINT_MAX if UUID doesn't exist
        uint pos = primitive_positions_[UUID];
        if (pos == UINT_MAX || pos >= primitive_count_) {
            return false; // Invalid UUID or out-of-bounds position
        }
        position = pos;
        return true;
    }
};
 
// ========== Camera Pixel Coordinate Helper ==========
 
struct PixelCoordinate {
    uint x; 
    uint y; 
 
    HELIOS_HOST_DEVICE
    PixelCoordinate(uint x_coord, uint y_coord) : x(x_coord), y(y_coord) {
    }
 
#ifdef __CUDACC__
    // CUDA version: use int2 (built-in CUDA type)
    __host__ __device__ inline size_t toFlatIndex(const int2 &full_resolution) const {
        return static_cast<size_t>(y) * full_resolution.x + x;
    }
#else
    // CPU version: use helios::int2
    inline size_t toFlatIndex(const helios::int2 &full_resolution) const {
        return static_cast<size_t>(y) * full_resolution.x + x;
    }
#endif
 
// CUDA-specific factory methods (use OptiX types)
#ifdef __CUDACC__
    __device__ static PixelCoordinate fromTiledLaunch(const optix::uint3 &launch_idx, const optix::int2 &tile_offset, const optix::int2 &full_resolution) {
        (void) full_resolution; // Unused but included for API clarity
        return PixelCoordinate(tile_offset.x + launch_idx.y, // Global x from local y-index
                               tile_offset.y + launch_idx.z // Global y from local z-index
        );
    }
 
    __device__ static size_t computeFlatIndex(const optix::uint3 &launch_idx, const optix::int2 &tile_offset, const optix::int2 &full_resolution) {
        PixelCoordinate pixel = fromTiledLaunch(launch_idx, tile_offset, full_resolution);
        return pixel.toFlatIndex(full_resolution);
    }
#endif // __CUDACC__
};
 
// ========== Subpatch UUID Calculator ==========
 
class SubpatchUUIDCalculator {
private:
    uint base_UUID_; 
#ifdef __CUDACC__
    int2 subdivisions_; 
#else
    helios::int2 subdivisions_; 
#endif
 
public:
#ifdef __CUDACC__
    __host__ __device__ SubpatchUUIDCalculator(uint base_UUID, int2 subdivisions) : base_UUID_(base_UUID), subdivisions_(subdivisions) {
    }
#else
    SubpatchUUIDCalculator(uint base_UUID, helios::int2 subdivisions) : base_UUID_(base_UUID), subdivisions_(subdivisions) {
    }
#endif
 
    HELIOS_HOST_DEVICE inline uint getUUID(int col, int row) const {
        return base_UUID_ + row * subdivisions_.x + col;
    }
 
    HELIOS_HOST_DEVICE inline int getSubpatchCount() const {
        return subdivisions_.x * subdivisions_.y;
    }
 
#ifdef __CUDACC__
    __host__ __device__ int2 getSubdivisions() const {
        return subdivisions_;
    }
#else
    helios::int2 getSubdivisions() const {
        return subdivisions_;
    }
#endif
};
 
#endif // HELIOS_BUFFER_INDEXING_H