Вычисление ограничивающего прямоугольника сетки в CUDA - PullRequest
2 голосов
/ 02 февраля 2012

Я пытался написать программу Bounding Box для сеток в CUDA

RT_PROGRAM void bounds (int primIdx, float result[6])
{
  int3 v_idx = index_buffer[primIdx];

  float3 v0 = vertex_buffer[ v_idx.x ];
  float3 v1 = vertex_buffer[ v_idx.y ];
  float3 v2 = vertex_buffer[ v_idx.z ];

  float3 min = fminf( fminf( v0, v1), v2 );  
  float3 max = fmaxf( fmaxf( v0, v1), v2 );

  Aabb* aabb = (Aabb*)result;
  aabb->m_min = min;
  aabb->m_max = max;
}

Ошибка не возникает, если я использую следующие строки вместо двух последних:

aabb->m_min = make_float3(0);
aabb->m_max = make_float3(0);

На хосте я использую этот исходный код:

Program mesh_bounds = g_ctx->createProgramFromPTXFile(ptx_path, "bounds");
Program mesh_intersect = g_ctx->createProgramFromPTXFile(ptx_path, "intersect2");

// create buffers
Buffer idx_buffer = g_ctx->createBuffer(RT_BUFFER_INPUT, RT_FORMAT_INT3, mesh->nfaces); // mesh->nfaces should always be 1 for triangles
Buffer vtx_buffer = g_ctx->createBuffer(RT_BUFFER_INPUT, RT_FORMAT_FLOAT3, mesh->nvertices);
Buffer nor_buffer = g_ctx->createBuffer(RT_BUFFER_INPUT, RT_FORMAT_FLOAT3, mesh->nfaces); // should only has 1 normal

// load materials
Material mat = g_ctx->createMaterial();

// create the Geometry
    Geometry geo = g_ctx->createGeometry();
    geo->setPrimitiveCount(mesh->nfaces); // mesh->nfaces should be 1

// set both programs to geometry
geo->setBoundingBoxProgram(mesh_bounds);
geo->setIntersectionProgram(mesh_intersect);

// set buffers
geo["vertex_buffer"]->setBuffer(vtx_buffer);
geo["normal_buffer"]->setBuffer(nor_buffer);
geo["index_buffer"]->setBuffer(idx_buffer);

// create Geometry Instance
GeometryInstance inst = g_ctx->createGeometryInstance();
inst->setGeometry(geo);
inst->setMaterialCount(1);
inst->setMaterial(0, mat);

    GeometryGroup grp = g_ctx->createGeometryGroup();
grp->setChildCount(1);
grp->setChild(0, inst);
    grp->setAcceleration(g_ctx->createAcceleration("Bvh", "Bvh"));
    g_meshes.push_back(grp);

После запуска скрипта я получаю следующее сообщение об ошибке:

***ERROR***: Unknown error (Details: Function "_rtContextLaunch2D" caught exception:
Encountered a CUDA error: driver(). cuEventSynchronize(m_event) returned (999): 
Unknown, [6619195])

У кого-нибудь есть идея?

Ответы [ 2 ]

1 голос
/ 05 апреля 2018

ОП уже ответил на этот вопрос самостоятельно.Тем не менее, было бы хорошо сказать, что Thrust уже имеет пример конструкции Bounding Box, хотя ограничен 2D, см. Пример Thrust Bounding Box .

Ниже я предоставляютрехмерное расширение этого примера, в котором используется float3 вместо настроенной в нем point2d структуры.

#include <thrust/transform_reduce.h>
#include <thrust/device_vector.h>
#include <thrust/pair.h>
#include <thrust/random.h>
#include <thrust/extrema.h>

/***********************/
/* BOUNDING BOX STRUCT */
/***********************/
struct bbox
{
    float3 lower_left, upper_right;

    // --- Empty box constructor
    __host__ __device__ bbox() {}

    // --- Construct a box from a single point
    __host__ __device__ bbox(const float3 &point) : lower_left(point), upper_right(point) {}

    // --- Construct a box from a pair of points
    __host__ __device__ bbox(const float3 &ll, const float3 &ur) : lower_left(ll), upper_right(ur) {}

};

/*********************************/
/* BOUNDING BOX REDUCTION STRUCT */
/*********************************/
// --- Reduce a pair of bounding boxes (a, b) to a bounding box containing a and b
struct bbox_reduction : public thrust::binary_function<bbox, bbox, bbox>
{
    __host__ __device__ bbox operator()(bbox a, bbox b)
    {
        // --- Lower left corner
        float3 ll = make_float3(thrust::min(a.lower_left.x, b.lower_left.x), thrust::min(a.lower_left.y, b.lower_left.y), thrust::min(a.lower_left.z, b.lower_left.z));

        // --- Upper right corner
        float3 ur = make_float3(thrust::max(a.upper_right.x, b.upper_right.x), thrust::max(a.upper_right.y, b.upper_right.y), thrust::max(a.upper_right.z, b.upper_right.z));

        return bbox(ll, ur);
    }
};

/********/
/* MAIN */
/********/
int main(void)
{
    const size_t N = 40;
    thrust::default_random_engine rng;
    thrust::uniform_real_distribution<float> u01(0.0f, 1.0f);

    // --- Allocate space for 3D points
    thrust::device_vector<float3> d_points(N);

    // --- Generate random 3D points in the unit cube
    for (size_t i = 0; i < N; i++)
    {
        float x = u01(rng);
        float y = u01(rng);
        float z = u01(rng);
        d_points[i] = make_float3(x, y, z);
    }

    // --- The initial bounding box contains the first point of the point cloud
    bbox init = bbox(d_points[0], d_points[0]);

    // --- Binary reduction operation
    bbox_reduction binary_op;

    // --- Compute the bounding box for the point set
    bbox result = thrust::reduce(d_points.begin(), d_points.end(), init, binary_op);

    for (int k = 0; k < N; k++) {
        float3 temp = d_points[k];
        printf("%d %f %f %f\n", k, temp.x, temp.y, temp.z);
    }

    // --- Print output
    std::cout << "bounding box " << std::fixed;
    std::cout << "(" << result.lower_left.x << "," << result.lower_left.y << "," << result.lower_left.z << ") ";
    std::cout << "(" << result.upper_right.x << "," << result.upper_right.y << "," << result.upper_right.z << ")" << std::endl;

    return 0;
}
1 голос
/ 02 февраля 2012

Извините за мое невежество, но эта строка выделяется.

Aabb* aabb = (Aabb*)result;

Вы уверены, что это хорошо? Вы приводите указатель с плавающей точкой в ​​указатель Aabb. Можете ли вы указать, что такое Aabb? Даже если это не ваша проблема, рассмотрите возможность использования приведения в новом стиле для лучшей проверки времени компиляции.

Aabb* aabb= static_cast<Aabb*> (result)
Добро пожаловать на сайт PullRequest, где вы можете задавать вопросы и получать ответы от других членов сообщества.
...