• Home
  • Line#
  • Scopes#
  • Navigate#
  • Raw
  • Download
1//
2// Copyright (C) 2009-2021 Intel Corporation
3//
4// SPDX-License-Identifier: MIT
5//
6//
7
8#include "api_interface.h"
9#include "d3d12.h"
10#include "common.h"
11
12GRL_ANNOTATE_IGC_DO_NOT_SPILL
13__attribute__((reqd_work_group_size(1, 1, 1))) void kernel compacted_size(global char *bvh_mem,
14                                                                          global char *postbuild_info)
15{
16    BVHBase *base = (BVHBase *)bvh_mem;
17    PostbuildInfoCompactedSize *postbuildInfoCompacted = (PostbuildInfoCompactedSize *)postbuild_info;
18
19    postbuildInfoCompacted->CompactedSizeInBytes = compute_compacted_size(base);
20}
21
22GRL_ANNOTATE_IGC_DO_NOT_SPILL
23__attribute__((reqd_work_group_size(1, 1, 1))) void kernel current_size(global char *bvh_mem,
24                                                                        global char *postbuild_info)
25{
26
27    BVHBase *base = (BVHBase *)bvh_mem;
28    PostbuildInfoCurrentSize *postbuildInfoCurrent = (PostbuildInfoCurrentSize *)postbuild_info;
29
30    postbuildInfoCurrent->CurrentSizeInBytes = base->Meta.allocationSize;
31}
32
33GRL_ANNOTATE_IGC_DO_NOT_SPILL
34__attribute__((reqd_work_group_size(1, 1, 1))) void kernel serialized_size(global char *bvh_mem,
35                                                                           global char *postbuild_info)
36{
37
38    BVHBase *base = (BVHBase *)bvh_mem;
39    PostbuildInfoSerializationDesc *postbuildInfoSerialization = (PostbuildInfoSerializationDesc *)postbuild_info;
40
41    uint64_t headerSize = sizeof(SerializationHeader);
42    uint64_t numInstances = base->Meta.instanceCount;
43
44    postbuildInfoSerialization->SerializedSizeInBytes = sizeof(SerializationHeader) +
45                                                        numInstances * sizeof(gpuva_t) +
46                                                        compute_compacted_size(base);
47                                                        //base->Meta.allocationSize;
48    postbuildInfoSerialization->NumBottomLevelAccelerationStructurePointers = numInstances;
49}
50
51void countTrianglesAndProcedurals(GeoMetaData *geoMetaData,
52                                  uint64_t numGeos,
53                                  uint64_t *numTriangles,
54                                  uint64_t *numProcedurals)
55{
56    uint64_t numTrianglesLoc = 0;
57    uint64_t numProceduralsLoc = 0;
58
59    for (uint64_t geoIndex = get_local_id(0); geoIndex < numGeos; geoIndex += get_local_size(0))
60    {
61        if (geoMetaData[geoIndex].Type == GEOMETRY_TYPE_TRIANGLES)
62        {
63            *numTriangles += geoMetaData[geoIndex].PrimitiveCount;
64        }
65        else
66        {
67            *numProcedurals += geoMetaData[geoIndex].PrimitiveCount;
68        }
69    }
70}
71
72GRL_ANNOTATE_IGC_DO_NOT_SPILL
73__attribute__((reqd_work_group_size(1, 1, 1))) void kernel decoded_size(global char *bvh_mem,
74                                                                        global char *postbuild_info)
75{
76    BVHBase *base = (BVHBase *)bvh_mem;
77    PostbuildInfoToolsVisualizationDesc *postbuildInfoDecoded = (PostbuildInfoToolsVisualizationDesc *)postbuild_info;
78
79    uint64_t numTriangles = 0;
80    uint64_t numProcedurals = 0;
81    countTrianglesAndProcedurals((GeoMetaData *)((uint64_t)base + (uint64_t)base->Meta.geoDescsStart),
82                                 base->Meta.geoCount,
83                                 &numTriangles,
84                                 &numProcedurals);
85    uint64_t numInstances = base->Meta.instanceCount;
86    uint64_t numDescs = base->Meta.geoCount;
87    uint64_t headerSize = sizeof(DecodeHeader);
88    uint64_t descsSize = numDescs * sizeof(D3D12_RAYTRACING_GEOMETRY_DESC) +
89                         numInstances * sizeof(D3D12_RAYTRACING_INSTANCE_DESC);
90
91    // Each triangle is stored separately - 3 vertices (9 floats) per triangle
92    uint64_t triangleDataSize = 9 * sizeof(float);
93    uint64_t proceduralDataSize = sizeof(D3D12_RAYTRACING_AABB);
94    uint64_t geoDataSize = numTriangles * triangleDataSize + numProcedurals * proceduralDataSize;
95
96    postbuildInfoDecoded->DecodedSizeInBytes = headerSize + descsSize + geoDataSize;
97}
98