Skip to content

Commit 9d57510

Browse files
mortaciousmortacious
authored
Merge pull request #1 from keckj/master
Add optixHello example
2 parents b905a52 + 94630a4 commit 9d57510

3 files changed

Lines changed: 169 additions & 0 deletions

File tree

‎examples/cuda/hello.cu‎

Lines changed: 45 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,45 @@
1+
//
2+
// Copyright (c) 2021, NVIDIA CORPORATION. All rights reserved.
3+
//
4+
// Redistribution and use in source and binary forms, with or without
5+
// modification, are permitted provided that the following conditions
6+
// are met:
7+
// * Redistributions of source code must retain the above copyright
8+
// notice, this list of conditions and the following disclaimer.
9+
// * Redistributions in binary form must reproduce the above copyright
10+
// notice, this list of conditions and the following disclaimer in the
11+
// documentation and/or other materials provided with the distribution.
12+
// * Neither the name of NVIDIA CORPORATION nor the names of its
13+
// contributors may be used to endorse or promote products derived
14+
// from this software without specific prior written permission.
15+
//
16+
// THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``AS IS'' AND ANY
17+
// EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
18+
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
19+
// PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
20+
// CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
21+
// EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
22+
// PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
23+
// PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY
24+
// OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
25+
// (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
26+
// OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
27+
//
28+
29+
#include <optix.h>
30+
31+
#include "hello.h"
32+
#include "helpers.h"
33+
34+
extern "C" {
35+
__constant__ Params params;
36+
}
37+
38+
extern "C"
39+
__global__ void __raygen__draw_solid_color()
40+
{
41+
uint3 launch_index = optixGetLaunchIndex();
42+
RayGenData* rtData = (RayGenData*)optixGetSbtDataPointer();
43+
params.image[launch_index.y * params.image_width + launch_index.x] =
44+
make_color( make_float3( rtData->r, rtData->g, rtData->b ) );
45+
}

‎examples/cuda/hello.h‎

Lines changed: 38 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,38 @@
1+
//
2+
// Copyright (c) 2021, NVIDIA CORPORATION. All rights reserved.
3+
//
4+
// Redistribution and use in source and binary forms, with or without
5+
// modification, are permitted provided that the following conditions
6+
// are met:
7+
// * Redistributions of source code must retain the above copyright
8+
// notice, this list of conditions and the following disclaimer.
9+
// * Redistributions in binary form must reproduce the above copyright
10+
// notice, this list of conditions and the following disclaimer in the
11+
// documentation and/or other materials provided with the distribution.
12+
// * Neither the name of NVIDIA CORPORATION nor the names of its
13+
// contributors may be used to endorse or promote products derived
14+
// from this software without specific prior written permission.
15+
//
16+
// THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``AS IS'' AND ANY
17+
// EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
18+
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
19+
// PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT OWNER OR
20+
// CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
21+
// EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
22+
// PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
23+
// PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY
24+
// OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
25+
// (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
26+
// OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
27+
//
28+
29+
struct Params
30+
{
31+
uchar4* image;
32+
unsigned int image_width;
33+
};
34+
35+
struct RayGenData
36+
{
37+
float r,g,b;
38+
};

‎examples/hello.py‎

Lines changed: 86 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,86 @@
1+
import optix as ox
2+
import cupy as cp
3+
import numpy as np
4+
from PIL import Image, ImageOps
5+
import logging
6+
import sys
7+
logging.basicConfig(stream=sys.stdout, level=logging.DEBUG)
8+
log = logging.getLogger()
9+
10+
def create_module(ctx, pipeline_opts):
11+
compile_opts = ox.ModuleCompileOptions(debug_level=ox.CompileDebugLevel.LINEINFO)
12+
module = ox.Module(ctx, 'cuda/hello.cu', compile_opts, pipeline_opts)
13+
return module
14+
15+
16+
def create_program_groups(ctx, module):
17+
raygen_grp = ox.ProgramGroup.create_raygen(ctx, module, "__raygen__draw_solid_color")
18+
miss_grp = ox.ProgramGroup.create_miss(ctx, None, None)
19+
return raygen_grp, miss_grp
20+
21+
22+
def create_pipeline(ctx, program_grps, pipeline_options):
23+
link_opts = ox.PipelineLinkOptions(max_trace_depth=0, debug_level=ox.CompileDebugLevel.FULL)
24+
25+
pipeline = ox.Pipeline(ctx, compile_options=pipeline_options, link_options=link_opts,
26+
program_groups=program_grps, max_traversable_graph_depth=1)
27+
28+
pipeline.compute_stack_sizes(0, # max_trace_depth
29+
0, # max_cc_depth
30+
0) # max_dc_depth
31+
return pipeline
32+
33+
34+
def create_sbt(program_grps):
35+
raygen_grp, miss_grp = program_grps
36+
37+
raygen_sbt = ox.SbtRecord(raygen_grp, names=('r','g','b'), formats=('f4',)*3)
38+
raygen_sbt['r'] = 0.462
39+
raygen_sbt['g'] = 0.725
40+
raygen_sbt['b'] = 0.0
41+
42+
miss_sbt = ox.SbtRecord(miss_grp)
43+
44+
sbt = ox.ShaderBindingTable(raygen_record=raygen_sbt, miss_records=miss_sbt)
45+
46+
return sbt
47+
48+
49+
def launch_pipeline(pipeline : ox.Pipeline, sbt):
50+
width = 512
51+
height = 384
52+
53+
output_image = cp.empty(shape=(height,width,4), dtype=np.uint8)
54+
55+
params = ox.LaunchParamsRecord(names=('image', 'image_width'),
56+
formats=('u8', 'u4'))
57+
params['image'] = output_image.data.ptr
58+
params['image_width'] = width
59+
60+
stream = cp.cuda.Stream()
61+
62+
pipeline.launch(sbt, dimensions=(width, height), params=params, stream=stream)
63+
64+
stream.synchronize()
65+
66+
return cp.asnumpy(output_image)
67+
68+
69+
if __name__ == "__main__":
70+
logger = ox.Logger(log)
71+
ctx = ox.DeviceContext(validation_mode=True, log_callback_function=logger, log_callback_level=4)
72+
ctx.cache_enabled = False
73+
74+
tgf = ox.TraversableGraphFlags
75+
pipeline_options = ox.PipelineCompileOptions(traversable_graph_flags=tgf.ALLOW_SINGLE_LEVEL_INSTANCING | tgf.ALLOW_SINGLE_GAS,
76+
num_payload_values=0,
77+
num_attribute_values=0,
78+
exception_flags=ox.ExceptionFlags.NONE,
79+
pipeline_launch_params_variable_name="params")
80+
81+
module = create_module(ctx, pipeline_options)
82+
program_grps = create_program_groups(ctx, module)
83+
pipeline = create_pipeline(ctx, program_grps, pipeline_options)
84+
sbt = create_sbt(program_grps)
85+
img = launch_pipeline(pipeline, sbt)
86+
Image.fromarray(img, 'RGBA').show()

0 commit comments

Comments
 (0)