[WIP] scene

RenderKit · Nov 6, 2024 · e3bc653 · e3bc653
1 parent 8eb2895
commit e3bc653
Show file tree

Hide file tree

Showing 21 changed files with 44 additions and 2 deletions.
diff --git a/include/embree4/rtcore_scene.h b/include/embree4/rtcore_scene.h
@@ -92,6 +92,8 @@ RTC_API void rtcRetainScene(RTCScene scene);
 /* Releases the scene (decrements the reference count). */
 RTC_API void rtcReleaseScene(RTCScene scene);
 
+/* Returns the representation of the scene which is accessible on the device */
+RTC_API RTCScene rtcGetSceneDevicePointer(RTCScene scene);
 
 /* Attaches the geometry to a scene. */
 RTC_API unsigned int rtcAttachGeometry(RTCScene scene, RTCGeometry geometry);

diff --git a/kernels/common/rtcore.cpp b/kernels/common/rtcore.cpp
@@ -266,6 +266,19 @@ RTC_NAMESPACE_BEGIN;
     return (RTCDevice)nullptr;
   }
 
+  RTC_API RTCScene rtcGetSceneDevicePointer(RTCScene hscene)
+  {
+    Scene* scene = (Scene*) hscene;
+    RTC_CATCH_BEGIN;
+    RTC_TRACE(rtcGetSceneDevice);
+    RTC_VERIFY_HANDLE(hscene);
+    if (scene->getDevicePointer() == nullptr)
+      throw_RTCError(RTC_ERROR_INVALID_OPERATION,"scene device pointer is NULL");
+    return (RTCScene)scene->getDevicePointer();
+    RTC_CATCH_END2(scene);
+    return (RTCScene)nullptr;
+  }
+
   RTC_API void rtcSetSceneProgressMonitorFunction(RTCScene hscene, RTCProgressMonitorFunction progress, void* ptr) 
   {
     Scene* scene = (Scene*) hscene;

diff --git a/kernels/common/scene.cpp b/kernels/common/scene.cpp
@@ -42,6 +42,7 @@ namespace embree
 
   Scene::Scene (Device* device)
     : device(device),
+      scene_device(nullptr),
       flags_modified(true), enabled_geometry_types(0),
       scene_flags(RTC_SCENE_FLAG_NONE),
       quality_flags(RTC_BUILD_QUALITY_MEDIUM),

diff --git a/kernels/common/scene.h b/kernels/common/scene.h
@@ -37,7 +37,7 @@ namespace embree
   /*! Base class all scenes are derived from */
   class Scene : public AccelN
   {
-    ALIGNED_CLASS_USM_(std::alignment_of<Scene>::value);
+    ALIGNED_CLASS_(std::alignment_of<Scene>::value);
 
   public:
     template<typename Ty, bool mblur = false>
@@ -317,11 +317,17 @@ namespace embree
 
 #if defined(EMBREE_SYCL_SUPPORT)
     void syncWithDevice(sycl::queue* queue);
+
+  public:
+    Scene* getDevicePointer() {
+      return scene_device;
+    }
 #endif
 
 
   public:
     Device* device;
+    Scene* scene_device;
 
   public:
     IDPool<unsigned,0xFFFFFFFE> id_pool;

diff --git a/kernels/common/scene_instance.cpp b/kernels/common/scene_instance.cpp
@@ -172,6 +172,7 @@ namespace embree
 
     // override local2world value with device ptr in geometries_data_host
     Instance* instance = (Instance*)(data_host + offsetInstance);
+    instance->object = ((Scene*)(instance->object))->getDevicePointer();
     instance->local2world = (AffineSpace3ff*)(data_device + offsetInstance + sizeof(Instance));
   }
 

diff --git a/kernels/sycl/scene_sycl.cpp b/kernels/sycl/scene_sycl.cpp
@@ -180,6 +180,9 @@ void Scene::syncWithDevice(sycl::queue* queue_in)
     device->free(offsets);
 
   } // run
+
+  scene_device = (Scene*) device->malloc(sizeof(Scene), 16, EmbreeMemoryType::DEVICE);
+  queue.memcpy(scene_device, (void*)this, sizeof(Scene));
 
   if (!queue_in)
     queue.wait_and_throw();

diff --git a/tutorials/curve_geometry/curve_geometry_device.cpp b/tutorials/curve_geometry/curve_geometry_device.cpp
@@ -253,6 +253,7 @@ extern "C" void renderFrameStandard (int* pixels,
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/debug_device_memory/debug_device_memory_device.cpp b/tutorials/debug_device_memory/debug_device_memory_device.cpp
@@ -411,6 +411,7 @@ extern "C" void renderFrameStandard (int* pixels,
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/dynamic_scene/dynamic_scene_device.cpp b/tutorials/dynamic_scene/dynamic_scene_device.cpp
@@ -279,6 +279,7 @@ extern "C" void renderFrameStandard (int* pixels,
   /* render all pixels */
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/forest/forest_device.cpp b/tutorials/forest/forest_device.cpp
@@ -430,6 +430,7 @@ extern "C" void renderFrameStandard (int* pixels,
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/grid_geometry/grid_geometry_device.cpp b/tutorials/grid_geometry/grid_geometry_device.cpp
@@ -670,6 +670,7 @@ extern "C" void renderFrameStandard (int* pixels,
   /* render image */
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/hair_geometry/hair_geometry_device.cpp b/tutorials/hair_geometry/hair_geometry_device.cpp
@@ -429,6 +429,7 @@ extern "C" void renderFrameStandard (int* pixels,
   /* render frame */
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.scene = rtcGetSceneDevicePointer(data.scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/instanced_geometry/instanced_geometry_device.cpp b/tutorials/instanced_geometry/instanced_geometry_device.cpp
@@ -282,6 +282,7 @@ extern "C" void renderFrameStandard (int* pixels,
   /* render all pixels */
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/intersection_filter/intersection_filter_device.cpp b/tutorials/intersection_filter/intersection_filter_device.cpp
@@ -402,6 +402,7 @@ extern "C" void renderFrameStandard (int* pixels,
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/minimal/minimal_sycl.cpp b/tutorials/minimal/minimal_sycl.cpp
@@ -203,10 +203,12 @@ RTCScene initializeScene(RTCDevice device, const sycl::queue& queue)
  * (dx, dy, dz).
  */
 
-void castRay(sycl::queue& queue, const RTCScene scene, 
+void castRay(sycl::queue& queue, const RTCScene scene_in, 
              float ox, float oy, float oz,
              float dx, float dy, float dz, Result* result)
 {
+  RTCScene scene = rtcGetSceneDevicePointer(scene_in);
+
   queue.submit([=](sycl::handler& cgh)
   {
     cgh.set_specialization_constant<feature_mask>(required_features);

diff --git a/tutorials/motion_blur_geometry/motion_blur_geometry_device.cpp b/tutorials/motion_blur_geometry/motion_blur_geometry_device.cpp
@@ -738,6 +738,7 @@ extern "C" void renderFrameStandard (int* pixels,
   /* render next frame */
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/next_hit/next_hit_device.cpp b/tutorials/next_hit/next_hit_device.cpp
@@ -467,6 +467,7 @@ extern "C" void renderFrameStandard (int* pixels,
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.scene = rtcGetSceneDevicePointer(data.scene);
 
 #if defined(USE_SPECIALIZATION_CONSTANTS)
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh) {

diff --git a/tutorials/point_geometry/point_geometry_device.cpp b/tutorials/point_geometry/point_geometry_device.cpp
@@ -190,6 +190,7 @@ extern "C" void renderFrameStandard (int* pixels,
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/ray_mask/ray_mask_device.cpp b/tutorials/ray_mask/ray_mask_device.cpp
@@ -205,6 +205,7 @@ extern "C" void renderFrameStandard (int* pixels,
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/triangle_geometry/triangle_geometry_device.cpp b/tutorials/triangle_geometry/triangle_geometry_device.cpp
@@ -192,6 +192,7 @@ extern "C" void renderFrameStandard (int* pixels,
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {

diff --git a/tutorials/user_geometry/user_geometry_device.cpp b/tutorials/user_geometry/user_geometry_device.cpp
@@ -872,6 +872,7 @@ extern "C" void renderFrameStandard (int* pixels,
 {
 #if defined(EMBREE_SYCL_TUTORIAL) && !defined(EMBREE_SYCL_RT_SIMULATION)
   TutorialData ldata = data;
+  ldata.g_scene = rtcGetSceneDevicePointer(data.g_scene);
   sycl::event event = global_gpu_queue->submit([=](sycl::handler& cgh){
     const sycl::nd_range<2> nd_range = make_nd_range(height,width);
     cgh.parallel_for(nd_range,[=](sycl::nd_item<2> item) {