Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Add foundational CUDA support and basic kernel #459

Draft
wants to merge 23 commits into
base: main
Choose a base branch
from
Draft
Changes from 1 commit
Commits
Show all changes
23 commits
Select commit Hold shift + click to select a range
08421f2
check in partially completed, need kernel to continue
corybarr Aug 22, 2023
a0bcc4e
fix merge conflict
corybarr Aug 22, 2023
1289ce1
add stub of CudaManager.h
corybarr Aug 23, 2023
2cc510b
Revert "Remove cuda from build. No longer needed for Kit 105"
lilleyse Aug 24, 2023
7dbbc2b
Fix all compilation errors
corybarr Aug 24, 2023
6477098
Merge branch 'main' of github.com:CesiumGS/cesium-omniverse into add-…
corybarr Aug 24, 2023
b06bbb1
update extern/CMakeLists.txt
corybarr Aug 24, 2023
d5546b2
add nvrtc and cuda to CMakeLists.txt
corybarr Aug 24, 2023
1c3c770
manually revert commit 34e6e3ff6917ef59d2786540fe4a5ac6f5e9f7f6
corybarr Aug 25, 2023
c844568
add in tileId
corybarr Aug 25, 2023
ab04c1f
enable basic print-to-screen kernel
corybarr Aug 25, 2023
18422c3
run placeholder kernel for every tile
corybarr Aug 25, 2023
0324bd4
merge with add-cuda-cmake
corybarr Aug 28, 2023
3e37801
remove cudart from CMakeLists.txt
corybarr Aug 28, 2023
643f76d
delete CudaRunners when tile is freed
corybarr Aug 28, 2023
f0ec135
clang formatting fixes
corybarr Aug 28, 2023
701b60b
cosmetic
corybarr Aug 28, 2023
f8a62f5
clean up CudaManager.h
corybarr Aug 28, 2023
8313f49
clean up CudaManager
corybarr Aug 28, 2023
39a702d
fix variable shadowing
corybarr Aug 28, 2023
49b897e
remove redundant lines in CMakeLists.txt
Aug 30, 2023
866ee66
Merge branch 'add-cuda-cmake' into add-cuda
corybarr Aug 31, 2023
f08be11
Revert "Temporarily remove nvrtc and nvrtc-builtins from install"
lilleyse Oct 3, 2023
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
Prev Previous commit
Next Next commit
delete CudaRunners when tile is freed
corybarr committed Aug 28, 2023

Verified

This commit was created on GitHub.com and signed with GitHub’s verified signature.
commit 643f76da95618eeddb082c2ab28d4379288aabb5
2 changes: 1 addition & 1 deletion src/core/include/cesium/omniverse/CudaKernels.h
Original file line number Diff line number Diff line change
@@ -32,7 +32,7 @@ namespace cesium::omniverse::cudaKernels {
size_t i = blockIdx.x * blockDim.x + threadIdx.x;
if (count <= i) return;

printf("Placeholder: create voxel at index %llu\n", i);
// printf("Placeholder: create voxel at index %llu\n", i);
}
)";
}
4 changes: 2 additions & 2 deletions src/core/include/cesium/omniverse/CudaManager.h
Original file line number Diff line number Diff line change
@@ -31,7 +31,7 @@ namespace cesium::omniverse {

enum CudaUpdateType {
ONCE,
ON_UPDATE
ON_UPDATE_FRAME
};

struct CudaKernelArgs {
@@ -47,7 +47,7 @@ namespace cesium::omniverse {

void onUpdateFrame();
void addRunner(CudaRunner& cudaRunner);
// void removeRunner(std::string tileId, CudaUpdateType updateType);
void removeRunner(int64_t tileId);
[[nodiscard]] const char* getKernelCode(CudaKernelType kernelType) const;
[[nodiscard]] const char* getFunctionName(CudaKernelType kernelType) const;

Original file line number Diff line number Diff line change
@@ -27,6 +27,7 @@ struct FabricMesh {
struct TileRenderResources {
glm::dmat4 tileTransform;
std::vector<FabricMesh> fabricMeshes;
int64_t tileId;
};

class FabricPrepareRenderResources final : public Cesium3DTilesSelection::IPrepareRendererResources {
9 changes: 9 additions & 0 deletions src/core/src/CudaManager.cpp
Original file line number Diff line number Diff line change
@@ -38,6 +38,9 @@ namespace cesium::omniverse {
void CudaManager::runAllRunners() {
for (auto& [updateType, runners] : _runnersByUpdateType) {
if (updateType == CudaUpdateType::ONCE) {
throw std::runtime_error("Single-run kernels are not yet supported.");
}
else if (updateType == CudaUpdateType::ON_UPDATE_FRAME) {
auto onceRunners = runners;
for (auto& [tileId, runner] : runners) {
runRunner(runner);
@@ -63,6 +66,12 @@ namespace cesium::omniverse {

}

void CudaManager::removeRunner(int64_t tileId) {
for (auto& [updateType, runners] : _runnersByUpdateType) {
runners.erase(tileId);
}
}

void** CudaManager::packArgs(CudaKernelArgs cudaKernelArgs, CudaKernelType cudaKernelType) {
void **args = new void*[10](); // dynamically allocate an array of 10 void pointers and zero-initialize them

2 changes: 1 addition & 1 deletion src/core/src/FabricGeometry.cpp
Original file line number Diff line number Diff line change
@@ -319,7 +319,7 @@ void FabricGeometry::setGeometry(
CudaKernelArgs kernelArgs;
kernelArgs.args["points"] = pointsFabric;
auto elementCount = pointsFabric.size();
CudaRunner runner{CudaKernelType::CREATE_VOXELS, CudaUpdateType::ONCE, tileId, kernelArgs, static_cast<int>(elementCount)}; //TODO: tile ID
CudaRunner runner{CudaKernelType::CREATE_VOXELS, CudaUpdateType::ON_UPDATE_FRAME, tileId, kernelArgs, static_cast<int>(elementCount)}; //TODO: tile ID
CudaManager::getInstance().addRunner(runner);

size_t vertIndex = 0;
4 changes: 4 additions & 0 deletions src/core/src/FabricPrepareRenderResources.cpp
Original file line number Diff line number Diff line change
@@ -1,6 +1,7 @@
#include "cesium/omniverse/FabricPrepareRenderResources.h"

#include "cesium/omniverse/Context.h"
#include "cesium/omniverse/CudaManager.h"
#include "cesium/omniverse/FabricGeometry.h"
#include "cesium/omniverse/FabricMaterial.h"
#include "cesium/omniverse/FabricResourceManager.h"
@@ -330,6 +331,7 @@ void* FabricPrepareRenderResources::prepareInMainThread(Cesium3DTilesSelection::
return new TileRenderResources{
tileTransform,
std::move(fabricMeshes),
(meshes.size() > 0) ? meshes[0].tileId : 0
};
}

@@ -346,6 +348,8 @@ void FabricPrepareRenderResources::free(
if (pMainThreadResult) {
const auto pTileRenderResources = static_cast<TileRenderResources*>(pMainThreadResult);
freeFabricMeshes(pTileRenderResources->fabricMeshes);
auto tileId = pTileRenderResources->tileId;
CudaManager::getInstance().removeRunner(static_cast<int64_t>(tileId));
delete pTileRenderResources;
}
}