Function trav_loop

kernels/sycl/rthwif_embree.cpp:598–664

template<typename Ray>
__forceinline void trav_loop(intel_ray_query_t& query, Ray& ray, Scene* scene, sycl::private_ptr<RayQueryContext> context, const RTCFeatureFlags feature_mask)
{
  Scenes scenes(scene);
  while (!intel_is_traversal_done(query))
  {
    intel_candidate_type_t candidate = intel_get_hit_candidate(query, intel_hit_type_potential_hit);

    const unsigned int bvh_level = intel_get_hit_bvh_level( query, intel_hit_type_potential_hit );
    const float3 org = intel_get_ray_origin   ( query, bvh_level );
    const float3 dir = intel_get_ray_direction( query, bvh_level );
    const float t = intel_get_hit_distance(query, intel_hit_type_potential_hit);
    const float2 uv = intel_get_hit_barycentrics (query, intel_hit_type_potential_hit);
    const unsigned int geomID = intel_get_hit_geometry_id(query, intel_hit_type_potential_hit);
    const unsigned int primID = intel_get_hit_primitive_id(query, intel_hit_type_potential_hit);

    ray.org = Vec3ff(org.x(), org.y(), org.z(), ray.tnear());
    ray.dir = Vec3ff(dir.x(), dir.y(), dir.z(), ray.time ());
    ray.tfar = intel_get_hit_distance(query, intel_hit_type_committed_hit);

#if RTC_MAX_INSTANCE_LEVEL_COUNT > 1
    context->user->instStackSize = bvh_level;
    Scene* scene = scenes.get(bvh_level);
#else
    const unsigned int instID = intel_get_hit_instance_id(query, intel_hit_type_potential_hit);

    /* assume software instancing mode by default (required for rtcForwardRay) */
    Scene* scene = scenes.get(bvh_level);

    /* if we are in hardware instancing mode and we need to read the scene from the instance */
    if (bvh_level > 0 && instID != RTC_INVALID_GEOMETRY_ID) {
      Instance* inst = scenes.get(0)->get<Instance>(instID);
      scene = (Scene*) inst->object;
      context->user->instID[0] = instID;
    }
    else if (bvh_level == 0)
      context->user->instID[0] = RTC_INVALID_GEOMETRY_ID;

#endif
    context->scene = scene;
    Geometry* geom = scene->get(geomID);

    /* perform ray masking */
#if defined(EMBREE_RAY_MASK)
    if (ray.mask & geom->mask)
#endif
    {
      if (candidate == intel_candidate_type_procedural)
      {
        if (intersect_primitive(query,ray,scenes,geom,context,geomID,primID,feature_mask))
          if (commit_potential_hit (query, ray))
            break; // shadow rays break at first hit

      }
      else // if (candidate == TRIANGLE)
      {
        ray.tfar = t;
        Vec3f Ng = intel_get_hit_triangle_normal(query, intel_hit_type_potential_hit);
        Hit hit(context->user,geomID,primID,Vec2f(uv.x(),uv.y()),Ng);