<commit_before>// This file is distributed under the MIT license.
// See the LICENSE file for details.

#pragma once

#ifndef VSNRAY_BVH_H
#define VSNRAY_BVH_H

#include <cstddef>
#include <stdexcept>
#include <vector>

#ifdef __CUDACC__
#include <cuda.h>
#include <thrust/copy.h>
#include <thrust/device_vector.h>
#endif

#include "detail/aligned_vector.h"
#include "math/math.h"

namespace visionaray
{

struct bvh_node
{
    bvh_node() : num_prims(0) {}

    aabb bbox;
    union
    {
        unsigned first_child;
        unsigned first_prim;
    };
    unsigned num_prims;
};


//-------------------------------------------------------------------------------------------------
//
//

VSNRAY_FUNC
inline bool is_inner(bvh_node const& node)
{
    return node.num_prims == 0;
}

VSNRAY_FUNC
inline bool is_leaf(bvh_node const& node)
{
    return node.num_prims != 0;
}


template <typename P>
class bvh
{
public:

    typedef std::size_t size_type;
    typedef aligned_vector<bvh_node> node_vector_type;
    typedef aligned_vector<unsigned> idx_vector_type;

    size_type num_prims;

#ifdef BVH_WITH_GATHER
    explicit bvh(P const* prims, size_t num_prims)
        : primitives_(prims)
        , num_prims(num_prims)
        , nodes_(node_vector_type(2 * num_prims - 1))
        , prim_indices_(idx_vector_type(num_prims))
    {
    }
#else
    // TODO!!!!!
    explicit bvh(P const* prims, size_t num_prims)
        : primitives_(0)
        , num_prims(num_prims)
        , nodes_(node_vector_type(2 * num_prims - 1))
    {
        VSNRAY_UNUSED(prims);
        primitives_ = new P[num_prims];
    }

   ~bvh()
    {
        // FIXME:
        delete[] primitives_;
    }
#endif

    P const* primitives() const { return primitives_; }

    bvh_node const* nodes() const { return nodes_.data(); }
    bvh_node* nodes() { return nodes_.data(); }

    node_vector_type const& nodes_vector() const { return nodes_; }
    node_vector_type& nodes_vector() { return nodes_; }

    unsigned const* prim_indices() const { return prim_indices_.data(); }
    unsigned* prim_indices() { return prim_indices_.data(); }

    idx_vector_type const& prim_indices_vector() const { return prim_indices_; }
    idx_vector_type& prim_indices_vector() { return prim_indices_; }

private:

    P const* primitives_;
    node_vector_type nodes_;
    idx_vector_type  prim_indices_;

};


#ifdef __CUDACC__

namespace detail
{

//-------------------------------------------------------------------------------------------------
// Wraps a thrust::device_vector and a raw pointer. The latter is accessible from a CUDA kernel.
// Does only implement the vector interface needed in this scope
//

template <typename T>
class kernel_device_vector
{
public:

    typedef typename thrust::device_vector<T>::size_type    size_type;
    typedef typename thrust::device_vector<T>::iterator     iterator;

    kernel_device_vector() = default;

    explicit kernel_device_vector(size_type n)
        : vector_(n)
    {
        ptr_ = thrust::raw_pointer_cast(vector_.data());
    }

    template <typename U, typename Alloc>
    /* implicit */ kernel_device_vector(std::vector<U, Alloc> const& host_vec)
        : vector_(host_vec)
        , ptr_(0)
    {
        ptr_ = thrust::raw_pointer_cast(vector_.data());
    }

    iterator begin() { return vector_.begin(); }
    iterator end()   { return vector_.end(); }

    VSNRAY_GPU_FUNC T const* data() const { return ptr_; }
    VSNRAY_GPU_FUNC T*       data()       { return ptr_; }

private:

    thrust::device_vector<T> vector_;
    T* ptr_;

};

} // detail


//-------------------------------------------------------------------------------------------------
// BVH for traversal on a CUDA device. Can only be edited on the device.
// Can only be constructed from a host BVH, is not copyable or copy assignable
//

template <typename P>
class device_bvh
{
public:

    device_bvh(bvh<P> const& host_bvh)
        : primitives_(host_bvh.num_prims)
    {
        thrust::copy(host_bvh.primitives(), host_bvh.primitives() + host_bvh.num_prims, primitives_.begin());
        nodes_ = host_bvh.nodes_vector();
        prim_indices_ = host_bvh.prim_indices_vector();
    }

    VSNRAY_GPU_FUNC P const*        primitives() const      { return primitives_.data(); }
    VSNRAY_GPU_FUNC P*              primitives()            { return primitives_.data(); }

    VSNRAY_GPU_FUNC bvh_node const* nodes() const           { return nodes_.data(); }
    VSNRAY_GPU_FUNC bvh_node*       nodes()                 { return nodes_.data(); }

    VSNRAY_GPU_FUNC unsigned const* prim_indices() const    { return prim_indices_.data(); }
    VSNRAY_GPU_FUNC unsigned*       prim_indices()          { return prim_indices_.data(); }

private:

    VSNRAY_NOT_COPYABLE(device_bvh)
    device_bvh& operator=(bvh<P> const& host_bvh);

    detail::kernel_device_vector<P>         primitives_;
    detail::kernel_device_vector<bvh_node>  nodes_;
    detail::kernel_device_vector<unsigned>  prim_indices_;

};

#endif // __CUDACC__


//-------------------------------------------------------------------------------------------------
//<commit_msg>Use initializer list and remove default ctor<commit_after>// This file is distributed under the MIT license.
// See the LICENSE file for details.

#pragma once

#ifndef VSNRAY_BVH_H
#define VSNRAY_BVH_H

#include <cstddef>
#include <stdexcept>
#include <vector>

#ifdef __CUDACC__
#include <cuda.h>
#include <thrust/copy.h>
#include <thrust/device_vector.h>
#endif

#include "detail/aligned_vector.h"
#include "math/math.h"

namespace visionaray
{

struct bvh_node
{
    bvh_node() : num_prims(0) {}

    aabb bbox;
    union
    {
        unsigned first_child;
        unsigned first_prim;
    };
    unsigned num_prims;
};


//-------------------------------------------------------------------------------------------------
//
//

VSNRAY_FUNC
inline bool is_inner(bvh_node const& node)
{
    return node.num_prims == 0;
}

VSNRAY_FUNC
inline bool is_leaf(bvh_node const& node)
{
    return node.num_prims != 0;
}


template <typename P>
class bvh
{
public:

    typedef std::size_t size_type;
    typedef aligned_vector<bvh_node> node_vector_type;
    typedef aligned_vector<unsigned> idx_vector_type;

    size_type num_prims;

#ifdef BVH_WITH_GATHER
    explicit bvh(P const* prims, size_t num_prims)
        : primitives_(prims)
        , num_prims(num_prims)
        , nodes_(node_vector_type(2 * num_prims - 1))
        , prim_indices_(idx_vector_type(num_prims))
    {
    }
#else
    // TODO!!!!!
    explicit bvh(P const* prims, size_t num_prims)
        : primitives_(0)
        , num_prims(num_prims)
        , nodes_(node_vector_type(2 * num_prims - 1))
    {
        VSNRAY_UNUSED(prims);
        primitives_ = new P[num_prims];
    }

   ~bvh()
    {
        // FIXME:
        delete[] primitives_;
    }
#endif

    P const* primitives() const { return primitives_; }

    bvh_node const* nodes() const { return nodes_.data(); }
    bvh_node* nodes() { return nodes_.data(); }

    node_vector_type const& nodes_vector() const { return nodes_; }
    node_vector_type& nodes_vector() { return nodes_; }

    unsigned const* prim_indices() const { return prim_indices_.data(); }
    unsigned* prim_indices() { return prim_indices_.data(); }

    idx_vector_type const& prim_indices_vector() const { return prim_indices_; }
    idx_vector_type& prim_indices_vector() { return prim_indices_; }

private:

    P const* primitives_;
    node_vector_type nodes_;
    idx_vector_type  prim_indices_;

};


#ifdef __CUDACC__

namespace detail
{

//-------------------------------------------------------------------------------------------------
// Wraps a thrust::device_vector and a raw pointer. The latter is accessible from a CUDA kernel.
// Does only implement the vector interface needed in this scope
//

template <typename T>
class kernel_device_vector
{
public:

    typedef typename thrust::device_vector<T>::size_type    size_type;
    typedef typename thrust::device_vector<T>::iterator     iterator;

    kernel_device_vector() = default;

    explicit kernel_device_vector(size_type n)
        : vector_(n)
    {
        ptr_ = thrust::raw_pointer_cast(vector_.data());
    }

    template <typename U, typename Alloc>
    /* implicit */ kernel_device_vector(std::vector<U, Alloc> const& host_vec)
        : vector_(host_vec)
        , ptr_(0)
    {
        ptr_ = thrust::raw_pointer_cast(vector_.data());
    }

    iterator begin() { return vector_.begin(); }
    iterator end()   { return vector_.end(); }

    VSNRAY_GPU_FUNC T const* data() const { return ptr_; }
    VSNRAY_GPU_FUNC T*       data()       { return ptr_; }

private:

    thrust::device_vector<T> vector_;
    T* ptr_;

};

} // detail


//-------------------------------------------------------------------------------------------------
// BVH for traversal on a CUDA device. Can only be edited on the device.
// Can only be constructed from a host BVH, is not copyable or copy assignable
//

template <typename P>
class device_bvh
{
public:

    device_bvh(bvh<P> const& host_bvh)
        : primitives_(host_bvh.num_prims)
        , nodes_(host_bvh.nodes_vector())
        , prim_indices_(host_bvh.prim_indices_vector())
    {
        thrust::copy(host_bvh.primitives(), host_bvh.primitives() + host_bvh.num_prims, primitives_.begin());
    }

    VSNRAY_GPU_FUNC P const*        primitives() const      { return primitives_.data(); }
    VSNRAY_GPU_FUNC P*              primitives()            { return primitives_.data(); }

    VSNRAY_GPU_FUNC bvh_node const* nodes() const           { return nodes_.data(); }
    VSNRAY_GPU_FUNC bvh_node*       nodes()                 { return nodes_.data(); }

    VSNRAY_GPU_FUNC unsigned const* prim_indices() const    { return prim_indices_.data(); }
    VSNRAY_GPU_FUNC unsigned*       prim_indices()          { return prim_indices_.data(); }

private:

    VSNRAY_NOT_COPYABLE(device_bvh)
    device_bvh& operator=(bvh<P> const& host_bvh);

    detail::kernel_device_vector<P>         primitives_;
    detail::kernel_device_vector<bvh_node>  nodes_;
    detail::kernel_device_vector<unsigned>  prim_indices_;

};

#endif // __CUDACC__


//-------------------------------------------------------------------------------------------------
//
{'fraction_non_alphanumeric': 0.175075045995933, 'fraction_numerical': 0.0017430037765081825, 'mean_word_length': 2.3241068554876088, 'pattern_counts': {'":': 0, '<': 45, '<?xml version=': 0, '>': 45, 'https://': 0, 'lorem ipsum': 0, 'www.': 0, 'xml': 0}, 'pii_count': 0, 'substrings_counts': 0, 'word_list_counts': {'cursed_substrings.json': 0, 'profanity_word_list.json': 0, 'sexual_word_list.json': 0, 'zh_pornsignals.json': 0}}
{'dir': 'git-commits-cleaned', 'id': '6780802', 'n_tokens_mistral': 3347, 'n_tokens_neox': 3186, 'n_words': 738}