Changeset View
Changeset View
Standalone View
Standalone View
intern/cycles/kernel/device/optix/kernel.cu
| Show First 20 Lines • Show All 153 Lines • ▼ Show 20 Lines | #ifdef __BVH_LOCAL__ | ||||
| isect->type = kernel_tex_fetch(__objects, isect->object).primitive_type; | isect->type = kernel_tex_fetch(__objects, isect->object).primitive_type; | ||||
| const float2 barycentrics = optixGetTriangleBarycentrics(); | const float2 barycentrics = optixGetTriangleBarycentrics(); | ||||
| isect->u = 1.0f - barycentrics.y - barycentrics.x; | isect->u = 1.0f - barycentrics.y - barycentrics.x; | ||||
| isect->v = barycentrics.x; | isect->v = barycentrics.x; | ||||
| /* Record geometric normal. */ | /* Record geometric normal. */ | ||||
| const uint tri_vindex = kernel_tex_fetch(__tri_vindex, prim).w; | const uint tri_vindex = kernel_tex_fetch(__tri_vindex, prim).w; | ||||
| const float3 tri_a = float4_to_float3(kernel_tex_fetch(__tri_verts, tri_vindex + 0)); | const float3 tri_a = kernel_tex_fetch(__tri_verts, tri_vindex + 0); | ||||
| const float3 tri_b = float4_to_float3(kernel_tex_fetch(__tri_verts, tri_vindex + 1)); | const float3 tri_b = kernel_tex_fetch(__tri_verts, tri_vindex + 1); | ||||
| const float3 tri_c = float4_to_float3(kernel_tex_fetch(__tri_verts, tri_vindex + 2)); | const float3 tri_c = kernel_tex_fetch(__tri_verts, tri_vindex + 2); | ||||
| local_isect->Ng[hit] = normalize(cross(tri_b - tri_a, tri_c - tri_a)); | local_isect->Ng[hit] = normalize(cross(tri_b - tri_a, tri_c - tri_a)); | ||||
| /* Continue tracing (without this the trace call would return after the first hit). */ | /* Continue tracing (without this the trace call would return after the first hit). */ | ||||
| optixIgnoreIntersection(); | optixIgnoreIntersection(); | ||||
| #endif | #endif | ||||
| } | } | ||||
| extern "C" __global__ void __anyhit__kernel_optix_shadow_all_hit() | extern "C" __global__ void __anyhit__kernel_optix_shadow_all_hit() | ||||
| ▲ Show 20 Lines • Show All 234 Lines • Show Last 20 Lines | |||||