推出 HIP RT – HIP 中的光线追踪库

首次发布时间:
Takahiro Harada's avatar
原田孝裕 (Takahiro Harada)
Daniel Meister's avatar
Daniel Meister

今天,我们发布了 HIP RT – 一个新的 HIP 光线追踪库。HIP RT 使在 HIP 中编写光线追踪应用程序变得容易,该库和 API 的设计宗旨是易于使用并能轻松集成到任何现有的 HIP 应用程序中。

虽然有一些其他光线追踪 API 引入了许多新方面,但我们通过消除学习许多新内核类型的需求,以略有不同的方式设计了 HIP RT。HIP RT 引入了 hiprtGeometryhiprtScene 等新对象类型。一旦将几何信息传递给 HIP RT,该库就会构建数据结构,然后将其传递给 HIP 内核。在此阶段,可以使用设备端库 API 来执行相交测试。

当前一代显卡,例如基于 AMD RDNA™ 2 架构的 GPU,支持硬件光线追踪加速以进一步优化渲染时间。然而,到目前为止,支持 HIP 的应用程序还无法利用此硬件加速。HIP RT 的设计目的是允许开发人员充分利用 AMD RDNA 2 架构 GPU 中用于硬件光线追踪的光线加速器。

功能

在此第一个版本中,我们实现了一些光线追踪应用程序所需的が基本功能。

三角形网格是 HIP RT 的基本图元,我们从中构建 hiprtGeometry。通常,场景由许多网格组成。开发人员可以利用这些网格之上的 hiprtScene 支持来创建加速结构,当光线与许多网格相交时,可以使用该结构来高效地搜索网格。此过程允许开发人员找到光线的最近命中点或捕获所有命中点。后一种情况对于阴影和透明度很有用。

为了加速光线追踪操作,HIP RT 在库中构建了一个加速结构。没有一个结构适合所有情况,因此,我们提供了几种选项供您选择,以找到适合您需求的结构。例如,您可以选择一个高质量的构建,它需要比其他选项稍长的时间来生成,但结果是一个高质量的数据结构,可以加快单次光线追踪查询的速度。或者,高级光线追踪用户可以选择使用 HIP RT 来加载自建的包围盒层次结构(BVH),然后 HIP RT 可以用于在 GPU 上执行相交。如今,这是 HIP RT 的一个独特功能,其他光线追踪 API 均不支持。例如,HIP RT 可用于为场景中的静态几何体构建高质量的 BVH,然后在运行时加载它。

我们的加速构建器不仅接受来自三角形网格的三角形,还接受轴对齐包围盒(AABBs)。如果对象在动态变化,此过程很有用,因为更改逻辑在 GPU 上执行。一个例子是蒙皮。将一个包含 AABBs 的缓冲区传递时,数据永远不会离开 GPU。用户只需在 GPU 上执行几何更新逻辑,将更新后的 AABBs 写入缓冲区,然后将其传递给我们即可。

如果您的场景包含三角形以外的其他图元,HIP RT 也可以处理它们。您需要做的就是为图元提供自定义相交函数,在 HIP RT 中注册该函数,然后当光线与图元相交时将调用该函数。

您的虚拟相机的曝光时间是有限的吗?那么,您可能想尝试我们的运动模糊。您需要做的就是在曝光时间内设置一些变换。关键帧之间的时间间隔不需要是均匀的。您可以在曝光时间内传递关键帧的位置。

代码示例

既然您对 HIP RT 的功能有了初步了解,让我们来看看 API。

这里有一些基本示例,但更详细的信息可在教程中找到。

创建上下文

创建 HIP 上下文和设备后,将其传递给 hiprtCreateContext 以创建 HIP RT 上下文。您可以在 00_context_creation 中找到代码。

构建几何体

创建 hiprtContext 后,下一步是构建 hiprtGeometry,这是我们在设备端执行光线相交所需的数据结构。创建 hiprtGeometry 的步骤如下:

hiprtTriangleMeshPrimitive mesh;
//set up the mesh
//get the temporary buffer size
hiprtGetGeometryBuildTemporaryBufferSize( ... );
//create hiprtGeometry
hiprtGeometry geom;
hiprtCreateGeometry( ... );
//build hiprtGeometry
hiprtBuildGeometry( ... );

您首先创建一个 hiprtTriangleMeshPrimitive 对象,然后通过调用 hiprtCreateGeometry() 创建一个 hiprtGeomety,可以在 hiprtBuildGeometry() 中构建它。您可以在 01_geom_intersection 中找到示例代码。

上面的代码用于启动设备端的光线投射过程,所以现在让我们转向设备代码。我们的第一个内核可以非常简单:

#include <hiprt/hiprt_device.h>
extern "C"
__global__ void MeshIntersectionKernel(hiprtGeometry geom, unsigned char* gDst, int2 cRes)
{
const int gIdx = blockIdx.x * blockDim.x + threadIdx.x;
const int gIdy = blockIdx.y * blockDim.y + threadIdx.y;
hiprtRay ray;
ray.origin = { gIdx / (float)cRes.x, gIdy / (float)cRes.y, -1.0f};
ray.direction = { 0.0f, 0.0f, 1.0f};
ray.maxT = 1000.f;
hiprtGeomTraversalClosest tr(geom, ray);
hiprtHit hit = tr.getNextHit();
bool hasHit = tr.hasHit();
int dstIdx = gIdx + gIdy * cRes.x;
gDst[ dstIdx * 4 + 0 ] = hasHit ? ((float)gIdx / cRes.x) * 255 : 0;
gDst[ dstIdx * 4 + 1 ] = hasHit ? ((float)gIdy / cRes.y) * 255 : 0;
gDst[ dstIdx * 4 + 2 ] = 0;
gDst[ dstIdx * 4 + 3 ] = 255;
}

显然,要使用我们的内置函数,您需要包含 hiprt_device.h。然后您可以看到我们正在接收一个 hiprtGeometry 对象,该对象刚刚在主机端构建。要执行光线场景相交,请填充 hiprtRay,它只是光线原点和方向,然后创建 hiprtGeomTraversalClosest,并调用 getNextHit(),这将返回命中信息。就这么简单!

构建场景

如果场景中只有一个三角形网格,构建 hiprtGeometry 就足够了。然而,在大多数情况下,场景中有许多对象,我们也会考虑这些对象。因此,我们有 hiprtScene 对象,它将这些 hiprtGeometry 对象捆绑在一起,然后创建一个我们可以传递给设备端的单个对象。hiprtScene 的创建如下。

hiprtScene scene;
//create hiprtScene
hiprtCreateScene( ... );
//build hiprtScene
hiprtBuildScene( ... );

虽然输入与构建 hiprtGeometry 不同,但结构是相同的。现在我们需要通过将其存储到 hiprtSceneBuildInput 对象中来传递场景信息,该对象被传递给 hiprtBuildScene()。您可以在 02_scene_intersection 中找到示例代码。

在设备端,我们可以重用上面的程序。我们所需要做的就是获取 hiprtScene 并创建一个 hiprtSceneTraversalClosest 对象。然后您应该能够通过场景追踪光线并渲染出这样的图像(当然是我们最喜欢的一个!)。

高级示例

虽然我们这里只展示了最基本的内容,但在我们的 SDK 中可以找到一些有趣的教程,例如用于提高性能的共享堆栈使用、运动模糊和自定义相交函数。

立即下载 HIP RT!

在这篇博文中,我们介绍了 HIP RT,这是一个 HIP 中的光线追踪库,并快速浏览了 API。我们希望您觉得这个第一个版本对您的项目有用。您可以在未来期待 HIP RT 的改进,敬请关注。

访问我们新的 HIP RT 页面立即下载

Takahiro Harada's avatar

原田孝裕 (Takahiro Harada)

Takahiro Harada 是 AMD 的一名研究员,也是一款名为 Radeon ProRender 的 GPU 全局照明渲染器的架构师。
Daniel Meister's avatar

Daniel Meister

Daniel Meister 是 AMD 的一名研究员和软件工程师。他的研究兴趣包括实时光线追踪、加速数据结构、全局照明、GPGPU 和用于渲染的机器学习。

相关新闻和技术文章

© . This site is unofficial and not affiliated with AMD.