From 1a92b4cab7afcbcb455bc0d7eec0e66f21cda423 Mon Sep 17 00:00:00 2001 From: Ashraf Kattoura Date: Sun, 11 Jan 2026 18:38:37 +0200 Subject: [PATCH 1/4] realsense linux receiver demos --- docs/user_guide/examples.md | 51 +++ examples/linux_d555_dual_stream.py | 322 +++++++++++++++++ examples/linux_d555_peoplenet.py | 284 +++++++++++++++ examples/linux_d555_player.py | 239 ++++++++++++ python/hololink/hololink.cpp | 2 +- python/hololink/operators/CMakeLists.txt | 1 + python/hololink/operators/__init__.py | 1 + python/hololink/operators/base_receiver_op.py | 1 + .../operators/image_decoder/CMakeLists.txt | 8 + .../operators/image_decoder/image_decoder.cpp | 102 ++++++ .../operators/linux_receiver_operator.py | 2 +- python/hololink/sensors/__init__.py | 1 + python/hololink/sensors/d555/__init__.py | 21 ++ python/hololink/sensors/d555/d555.py | 196 ++++++++++ python/hololink/sensors/d555/d555_mode.py | 118 ++++++ python/hololink/sensors/d555/model.py | 10 + python/setup.py | 1 + src/hololink/core/data_channel.cpp | 4 +- src/hololink/core/data_channel.hpp | 2 +- src/hololink/operators/CMakeLists.txt | 1 + .../operators/image_decoder/CMakeLists.txt | 32 ++ .../operators/image_decoder/image_decoder.cpp | 340 ++++++++++++++++++ .../operators/image_decoder/image_decoder.hpp | 98 +++++ 23 files changed, 1832 insertions(+), 5 deletions(-) create mode 100644 examples/linux_d555_dual_stream.py create mode 100644 examples/linux_d555_peoplenet.py create mode 100644 examples/linux_d555_player.py create mode 100644 python/hololink/operators/image_decoder/CMakeLists.txt create mode 100644 python/hololink/operators/image_decoder/image_decoder.cpp create mode 100644 python/hololink/sensors/d555/__init__.py create mode 100644 python/hololink/sensors/d555/d555.py create mode 100644 python/hololink/sensors/d555/d555_mode.py create mode 100644 python/hololink/sensors/d555/model.py create mode 100644 src/hololink/operators/image_decoder/CMakeLists.txt create mode 100644 src/hololink/operators/image_decoder/image_decoder.cpp create mode 100644 src/hololink/operators/image_decoder/image_decoder.hpp diff --git a/docs/user_guide/examples.md b/docs/user_guide/examples.md index 1afcd42c..3c19f37f 100644 --- a/docs/user_guide/examples.md +++ b/docs/user_guide/examples.md @@ -91,6 +91,57 @@ lastly, running SIPL accelerated network python example on AGX Thor: $ python3 ./examples/sipl_player.py --json-config ./examples/sipl_config/vb1940_single.json ``` +## RealSense D555 player example + +This example is similar to the IMX274 player example above, using Realsense D555 +camera instead of IMX274. To run the high-speed video player with Realsense D555, in +the demo container with a ConnectX accelerated network controller, + +```sh +$ python3 examples/linux_d555_player.py +$ python3 examples/linux_d555_dual_stream.py +$ python3 examples/linux_d555_peoplenet.py +``` + +### Known Limitations + +1. **Resolution Changes**: Switching resolutions (not FPS) requires a camera reboot. The first resolution set after reboot will be used. + +2. **IP Address Discovery**: The camera IP may not be the default `192.168.0.2`. To discover the actual IP address, run the `hololink-enumerate` tool from within the Docker container: + +```sh +$ hololink-enumerate +``` + +The output should display information similar to: + +``` +mac_id=98:4F:EE:1A:F4:A9 hsb_ip_version=0x2501 fpga_crc=0xffff ip_address=192.168.11.55 fpga_uuid=889b7ce3-65a5-4247-8b05-4ff1904c3359 serial_number=70255f4343534c interface=eno1 board=hololink-lite +mac_id=98:4F:EE:1A:F4:A9 hsb_ip_version=0x2501 fpga_crc=0xffff ip_address=192.168.11.55 fpga_uuid=889b7ce3-65a5-4247-8b05-4ff1904c3359 serial_number=70255f4343534c interface=eno1 board=hololink-lite +mac_id=98:4F:EE:1A:F4:A9 hsb_ip_version=0x2501 fpga_crc=0xffff ip_address=192.168.11.55 fpga_uuid=889b7ce3-65a5-4247-8b05-4ff1904c3359 serial_number=70255f4343534c interface=eno1 board=hololink-lite +^C +``` + +Note the `ip_address` and `interface` fields from the output. + +### Host Network Configuration + +To configure the host subnet to match the camera's IP address and enable communication: + +```sh +EN0=eno1 # Replace with your interface name from hololink-enumerate output +sudo nmcli con add con-name hololink-$EN0 ifname $EN0 type ethernet ip4 192.168.11.101/24 +sudo nmcli connection up hololink-$EN0 +``` + +Replace `192.168.11.101/24` with an IP address in the same subnet as your camera's IP address (e.g., if camera is `192.168.11.55`, use `192.168.11.101/24`). + +After configuration, verify connectivity by pinging the camera: + +```sh +$ ping 192.168.11.55 # Replace with your camera's IP address +``` + ## Running the TAO PeopleNet example The tao-peoplenet example demonstrates running inference on a live video feed. diff --git a/examples/linux_d555_dual_stream.py b/examples/linux_d555_dual_stream.py new file mode 100644 index 00000000..a5a75769 --- /dev/null +++ b/examples/linux_d555_dual_stream.py @@ -0,0 +1,322 @@ +# SPDX-FileCopystream2Text: Copystream2 (c) 2023-2024 NVIDIA CORPORATION & AFFILIATES. All stream2s reserved. +# SPDX-License-Identifier: Apache-2.0 +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +# See README.md for detailed information. + +import argparse +import ctypes +import logging +logging.basicConfig(level=logging.INFO, format='%(asctime)s - %(levelname)s - %(message)s') +import os +import sys +sys.path.append(os.path.abspath(os.path.join(os.path.dirname(__file__), '..'))) + +import holoscan +from cuda import cuda + +import hololink as hololink_module + +class HoloscanApplication(holoscan.core.Application): + def __init__( + self, + headless, + fullscreen, + cuda_context, + cuda_device_ordinal, + hololink_channel_stream_depth, + camera_stream_depth, + hololink_channel_stream_rgb, + camera_stream_rgb, + frame_limit, + window_height, + window_width, + window_title, + ): + logging.info("__init__") + super().__init__() + self._headless = headless + self._fullscreen = fullscreen + self._cuda_context = cuda_context + self._cuda_device_ordinal = cuda_device_ordinal + self._hololink_channel_stream_depth = hololink_channel_stream_depth + self._camera_stream_depth = camera_stream_depth + self._hololink_channel_stream_rgb = hololink_channel_stream_rgb + self._camera_stream_rgb = camera_stream_rgb + self._frame_limit = frame_limit + self._window_height = window_height + self._window_width = window_width + self._window_title = window_title + self.is_metadata_enabled = True + self.metadata_policy = holoscan.core.MetadataPolicy.REJECT + + def compose(self): + logging.info("compose") + if self._frame_limit: + self._count_stream1 = holoscan.conditions.CountCondition( + self, + name="count_stream1", + count=self._frame_limit, + ) + condition_stream1 = self._count_stream1 + self._count_stream2 = holoscan.conditions.CountCondition( + self, + name="count_stream2", + count=self._frame_limit, + ) + condition_stream2 = self._count_stream2 + else: + self._ok_stream1 = holoscan.conditions.BooleanCondition( + self, name="ok_stream1", enable_tick=True + ) + condition_stream1 = self._ok_stream1 + self._ok_stream2 = holoscan.conditions.BooleanCondition( + self, name="ok_stream2", enable_tick=True + ) + condition_stream2 = self._ok_stream2 + + image_decoder_allocator_pool = holoscan.resources.UnboundedAllocator(self) + + + image_decoder_stream1 = hololink_module.operators.ImageDecoderOp( + self, + name="image_decoder_stream1", + out_tensor_name="right", + allocator=image_decoder_allocator_pool, + cuda_device_ordinal=self._cuda_device_ordinal, + ) + self._camera_stream_depth.configure_converter(image_decoder_stream1) + + image_decoder_stream2 = hololink_module.operators.ImageDecoderOp( + self, + name="image_decoder_stream2", + out_tensor_name="left", + allocator=image_decoder_allocator_pool, + cuda_device_ordinal=self._cuda_device_ordinal, + ) + self._camera_stream_rgb.configure_converter(image_decoder_stream2) + + frame_size = image_decoder_stream1.get_csi_length() + frame_context = self._cuda_context + + receiver_operator_stream1 = hololink_module.operators.LinuxReceiverOperator( + self, + condition_stream1, + name="receiver_stream1", + frame_size=frame_size, + frame_context=frame_context, + udp_port=54739 + hololink_module.sensors.d555.d555_mode.RealSense_StreamId.DEPTH.value, + hololink_channel=self._hololink_channel_stream_depth, + device=self._camera_stream_depth, + ) + + receiver_operator_stream2 = hololink_module.operators.LinuxReceiverOperator( + self, + condition_stream2, + name="receiver_stream2", + frame_size=frame_size, + frame_context=frame_context, + udp_port=54739 + hololink_module.sensors.d555.d555_mode.RealSense_StreamId.RGB.value, + hololink_channel=self._hololink_channel_stream_rgb, + device=self._camera_stream_rgb, + ) + + left_spec = holoscan.operators.HolovizOp.InputSpec( + "left", holoscan.operators.HolovizOp.InputType.COLOR + ) + left_spec_view = holoscan.operators.HolovizOp.InputSpec.View() + left_spec_view.offset_x = 0 + left_spec_view.offset_y = 0 + left_spec_view.width = 0.5 + left_spec_view.height = 1 + left_spec.views = [left_spec_view] + + right_spec = holoscan.operators.HolovizOp.InputSpec( + "right", holoscan.operators.HolovizOp.InputType.COLOR + ) + right_spec_view = holoscan.operators.HolovizOp.InputSpec.View() + right_spec_view.offset_x = 0.5 + right_spec_view.offset_y = 0 + right_spec_view.width = 0.5 + right_spec_view.height = 1 + right_spec.views = [right_spec_view] + + visualizer = holoscan.operators.HolovizOp( + self, + name="holoviz", + headless=self._headless, + framebuffer_srgb=False, + tensors=[left_spec, right_spec], + height=self._window_height, + width=self._window_width, + window_title=self._window_title, + ) + + self.add_flow(receiver_operator_stream1, image_decoder_stream1, {("output", "input")}) + self.add_flow(receiver_operator_stream2, image_decoder_stream2, {("output", "input")}) + self.add_flow(image_decoder_stream1, visualizer, {("output", "receivers")}) + self.add_flow(image_decoder_stream2, visualizer, {("output", "receivers")}) + + +def main(): + parser = argparse.ArgumentParser() + modes_depth = hololink_module.sensors.d555.d555_mode.RealSense_Depth_Mode + mode_depth_choices = [mode.value for mode in modes_depth] + mode_help = " ".join([f"{mode.value}:{mode.name}" for mode in modes_depth]) + parser.add_argument( + "--camera-mode-depth", + type=int, + choices=mode_depth_choices, + default=mode_depth_choices[4], # 1280x720 60fps depth + help=mode_help, + ) + modes_rgb = hololink_module.sensors.d555.d555_mode.RealSense_RGB_Mode + mode_rgb_choices = [mode.value for mode in modes_rgb] + mode_help = " ".join([f"{mode.value}:{mode.name}" for mode in modes_rgb]) + parser.add_argument( + "--camera-mode-rgb", + type=int, + choices=mode_rgb_choices, + default=mode_rgb_choices[6], # 1280x720 60fps rgb + help=mode_help, + ) + parser.add_argument("--headless", action="store_true", help="Run in headless mode") + parser.add_argument( + "--fullscreen", action="store_true", help="Run in fullscreen mode" + ) + parser.add_argument( + "--frame-limit", + type=int, + default=None, + help="Exit after receiving this many frames", + ) + default_configuration = os.path.join( + os.path.dirname(__file__), "example_configuration.yaml" + ) + parser.add_argument( + "--configuration", + default=default_configuration, + help="Configuration file", + ) + parser.add_argument( + "--hololink", + default="192.168.0.2", + help="IP address of Hololink board", + ) + parser.add_argument( + "--log-level", + type=int, + default=20, + help="Logging level to display", + ) + parser.add_argument( + "--window-height", + type=int, + default=720, # arbitrary default + help="Set the height of the displayed window", + ) + parser.add_argument( + "--window-width", + type=int, + default=1280 * 2, # arbitrary default + help="Set the width of the displayed window", + ) + parser.add_argument( + "--title", + help="Set the window title", + ) + args = parser.parse_args() + hololink_module.logging_level(args.log_level) + logging.getLogger().setLevel(args.log_level) + logging.info("Initializing.") + # Get a handle to the GPU + (cu_result,) = cuda.cuInit(0) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + cu_device_ordinal = 0 + cu_result, cu_device = cuda.cuDeviceGet(cu_device_ordinal) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + cu_result, cu_context = cuda.cuDevicePrimaryCtxRetain(cu_device) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + + # Get a handle to data sources. First, find an enumeration packet + # from the IP address we want to use. + channel_metadata = hololink_module.Enumerator.find_channel(channel_ip=args.hololink) + overrides = hololink_module.Metadata({ + "vsync_enable": 0, # or 1 + "block_enable": 0, # or 1 + }) + channel_metadata.update(overrides) + logging.info(f"{channel_metadata=}") + # Now make separate connection metadata for stream1 and stream2; and set them to + # use sensor 0 and 1 respectively. This will borrow the data plane + # configuration we found on that interface. + channel_metadata_stream1 = hololink_module.Metadata(channel_metadata) + hololink_module.DataChannel.use_sensor(channel_metadata_stream1, 0) + channel_metadata_stream2 = hololink_module.Metadata(channel_metadata) + hololink_module.DataChannel.use_sensor(channel_metadata_stream2, 1) + # + hololink_channel_stream_depth = hololink_module.DataChannel(channel_metadata_stream1) + hololink_channel_stream_rgb = hololink_module.DataChannel(channel_metadata_stream2) + # Get a handle to the camera + camera_stream_depth = hololink_module.sensors.d555.d555.RealsenseCamD555(hololink_channel_stream_depth, hololink_module.sensors.d555.d555_mode.RealSense_StreamId.DEPTH) + camera_stream_rgb = hololink_module.sensors.d555.d555.RealsenseCamD555(hololink_channel_stream_rgb, hololink_module.sensors.d555.d555_mode.RealSense_StreamId.RGB) + + logging.info("camera mode stream Depth: %s", args.camera_mode_depth) + logging.info("camera mode stream RGB: %s", args.camera_mode_rgb) + + camera_mode_depth = hololink_module.sensors.d555.d555_mode.RealSense_Depth_Mode( + args.camera_mode_depth + ) + + camera_mode_rgb = hololink_module.sensors.d555.d555_mode.RealSense_RGB_Mode( + args.camera_mode_rgb + ) + + window_title = f"Holoviz - {args.hololink}" + if args.title is not None: + window_title = args.title + + # Set up the application + application = HoloscanApplication( + args.headless, + args.fullscreen, + cu_context, + cu_device_ordinal, + hololink_channel_stream_depth, + camera_stream_depth, + hololink_channel_stream_rgb, + camera_stream_rgb, + args.frame_limit, + args.window_height, + args.window_width, + window_title, + ) + application.config(args.configuration) + # # Run it. + hololink = hololink_channel_stream_depth.hololink() + assert hololink is hololink_channel_stream_rgb.hololink() + hololink.start() + + camera_stream_depth.configure(camera_mode_depth) + camera_stream_rgb.configure(camera_mode_rgb) + os.environ["GXF_MEMORY_DEBUG"] = "1" + application.run() + hololink.stop() + + (cu_result,) = cuda.cuDevicePrimaryCtxRelease(cu_device) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + + +if __name__ == "__main__": + main() diff --git a/examples/linux_d555_peoplenet.py b/examples/linux_d555_peoplenet.py new file mode 100644 index 00000000..fb6366ad --- /dev/null +++ b/examples/linux_d555_peoplenet.py @@ -0,0 +1,284 @@ +# SPDX-FileCopyrightText: Copyright (c) 2023-2024 NVIDIA CORPORATION & AFFILIATES. All rights reserved. +# SPDX-License-Identifier: Apache-2.0 +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +# See README.md for detailed information. + +import argparse +import ctypes +import logging +logging.basicConfig(level=logging.INFO, format='%(asctime)s - %(levelname)s - %(message)s') +import os +import sys +sys.path.append(os.path.abspath(os.path.join(os.path.dirname(__file__), '..'))) + +import holoscan +from cuda import cuda + +import hololink as hololink_module + +from tao_peoplenet import FormatInferenceInputOp, PostprocessorOp + + +class HoloscanApplication(holoscan.core.Application): + def __init__( + self, + headless, + fullscreen, + cuda_context, + cuda_device_ordinal, + hololink_channel, + camera, + camera_mode, + frame_limit, + engine, + stream_id + ): + logging.info("__init__") + super().__init__() + self._headless = headless + self._fullscreen = fullscreen + self._cuda_context = cuda_context + self._cuda_device_ordinal = cuda_device_ordinal + self._hololink_channel = hololink_channel + self._camera = camera + self._camera_mode = camera_mode + self._frame_limit = frame_limit + self._engine = engine + self._stream_id = stream_id + self.is_metadata_enabled = True + self.metadata_policy = holoscan.core.MetadataPolicy.REJECT + + def compose(self): + logging.info("compose") + if self._frame_limit: + self._count = holoscan.conditions.CountCondition( + self, + name="count", + count=self._frame_limit, + ) + condition = self._count + else: + self._ok = holoscan.conditions.BooleanCondition( + self, name="ok", enable_tick=True + ) + condition = self._ok + self._camera.set_mode(self._camera_mode) + + image_decoder_allocator_pool = holoscan.resources.BlockMemoryPool( + self, + name="pool", + # storage_type of 1 is device memory + storage_type=1, + block_size=self._camera._width + * ctypes.sizeof(ctypes.c_uint16) * 2 + * self._camera._height, + num_blocks=2, + ) + + image_decoder = hololink_module.operators.ImageDecoderOp( + self, + name="image_decoder", + out_tensor_name="", + allocator=image_decoder_allocator_pool, + cuda_device_ordinal=self._cuda_device_ordinal, + ) + self._camera.configure_converter(image_decoder) + + frame_size = image_decoder.get_csi_length() + frame_context = self._cuda_context + receiver_operator = hololink_module.operators.LinuxReceiverOperator( + self, + condition, + name="receiver", + frame_size=frame_size, + udp_port=54739 + self._stream_id, + frame_context=frame_context, + hololink_channel=self._hololink_channel, + device=self._camera, + ) + + visualizer = holoscan.operators.HolovizOp( + self, + name="holoviz", + fullscreen=self._fullscreen, + headless=self._headless, + framebuffer_srgb=False, + **self.kwargs("holoviz"), + ) + + image_shift = hololink_module.operators.ImageShiftToUint8Operator( + self, name="image_shift", shift=0 + ) + + pool = holoscan.resources.UnboundedAllocator(self) + preprocessor_args = self.kwargs("preprocessor") + preprocessor = holoscan.operators.FormatConverterOp( + self, + name="preprocessor", + pool=pool, + **preprocessor_args, + ) + format_input = FormatInferenceInputOp( + self, + name="transpose", + pool=pool, + ) + inference = holoscan.operators.InferenceOp( + self, + name="inference", + allocator=pool, + model_path_map={ + "face_detect": self._engine, + }, + **self.kwargs("inference"), + ) + postprocessor_args = self.kwargs("postprocessor") + postprocessor_args["image_width"] = preprocessor_args["resize_width"] + postprocessor_args["image_height"] = preprocessor_args["resize_height"] + postprocessor = PostprocessorOp( + self, + name="postprocessor", + allocator=pool, + **postprocessor_args, + ) + + self.add_flow(receiver_operator, image_decoder, {("output", "input")}) + + self.add_flow(image_decoder, image_shift, {("output", "input")}) + self.add_flow(image_shift, preprocessor, {("output", "")}) + self.add_flow(image_shift, visualizer, {("output", "receivers")}) + self.add_flow(preprocessor, format_input) + self.add_flow(format_input, inference, {("", "receivers")}) + self.add_flow(inference, postprocessor, {("transmitter", "in")}) + self.add_flow(postprocessor, visualizer, {("out", "receivers")}) + + +def main(): + parser = argparse.ArgumentParser() + modes = hololink_module.sensors.d555.d555_mode.RealSense_RGB_Mode + mode_choices = [mode.value for mode in modes] + mode_help = " ".join([f"{mode.value}:{mode.name}" for mode in modes]) + parser.add_argument( + "--camera-mode", + type=int, + choices=mode_choices, + default=mode_choices[6], + help=mode_help, + ) + parser.add_argument("--headless", action="store_true", help="Run in headless mode") + parser.add_argument( + "--fullscreen", action="store_true", help="Run in fullscreen mode" + ) + parser.add_argument( + "--frame-limit", + type=int, + default=None, + help="Exit after receiving this many frames", + ) + default_configuration = os.path.join( + os.path.dirname(__file__), "tao_peoplenet.yaml" + ) + parser.add_argument( + "--configuration", default=default_configuration, help="Configuration file" + ) + parser.add_argument( + "--hololink", + default="192.168.0.2", + help="IP address of Hololink board", + ) + default_engine = os.path.join( + os.path.dirname(__file__), "resnet34_peoplenet_int8.onnx" + ) + parser.add_argument( + "--engine", + default=default_engine, + help="TRT engine model", + ) + parser.add_argument( + "--log-level", + type=int, + default=20, + help="Logging level to display", + ) + parser.add_argument( + "--expander-configuration", + type=int, + default=0, + choices=(0, 1), + help="I2C Expander configuration", + ) + parser.add_argument( + "--pattern", + type=int, + choices=range(12), + help="Configure to display a test pattern.", + ) + args = parser.parse_args() + hololink_module.logging_level(args.log_level) + logging.getLogger().setLevel(args.log_level) + logging.info("Initializing.") + # Get a handle to the GPU + (cu_result,) = cuda.cuInit(0) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + cu_device_ordinal = 0 + cu_result, cu_device = cuda.cuDeviceGet(cu_device_ordinal) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + cu_result, cu_context = cuda.cuDevicePrimaryCtxRetain(cu_device) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + # # Get a handle to the Hololink device + channel_metadata = hololink_module.Enumerator.find_channel(channel_ip=args.hololink) + overrides = hololink_module.Metadata({ + "vsync_enable": 0, # or 1 + "block_enable": 0, # or 1 + }) + channel_metadata.update(overrides) + hololink_channel = hololink_module.DataChannel(channel_metadata) + # # Get a handle to the camera + camera = hololink_module.sensors.d555.d555.RealsenseCamD555(hololink_channel, hololink_module.sensors.d555.d555_mode.RealSense_StreamId.RGB) + logging.info("camera mode: %s", args.camera_mode) + camera_mode = hololink_module.sensors.d555.d555_mode.RealSense_RGB_Mode( + args.camera_mode + ) + + # Set up the application + application = HoloscanApplication( + args.headless, + args.fullscreen, + cu_context, + cu_device_ordinal, + hololink_channel, + camera, + camera_mode, + args.frame_limit, + args.engine, + hololink_module.sensors.d555.d555_mode.RealSense_StreamId.RGB + ) + application.config(args.configuration) + # # Run it. + hololink = hololink_channel.hololink() + hololink.start() + camera.setup_clock() + camera.configure(camera_mode) + camera.set_digital_gain_reg(0x4) + os.environ["GXF_MEMORY_DEBUG"] = "1" + application.run() + hololink.stop() + + (cu_result,) = cuda.cuDevicePrimaryCtxRelease(cu_device) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + + +if __name__ == "__main__": + main() diff --git a/examples/linux_d555_player.py b/examples/linux_d555_player.py new file mode 100644 index 00000000..14c95159 --- /dev/null +++ b/examples/linux_d555_player.py @@ -0,0 +1,239 @@ +# SPDX-FileCopyrightText: Copyright (c) 2023-2024 NVIDIA CORPORATION & AFFILIATES. All rights reserved. +# SPDX-License-Identifier: Apache-2.0 +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +# See README.md for detailed information. + +import argparse +import ctypes +import logging + +logging.basicConfig( + level=logging.INFO, format="%(asctime)s - %(levelname)s - %(message)s" +) +import os +import sys + +sys.path.append(os.path.abspath(os.path.join(os.path.dirname(__file__), ".."))) + +import holoscan +from cuda import cuda + +import hololink as hololink_module + + +class HoloscanApplication(holoscan.core.Application): + def __init__( + self, + headless, + fullscreen, + cuda_context, + cuda_device_ordinal, + hololink_channel, + camera, + camera_mode, + frame_limit, + stream_id, + ): + logging.info("__init__") + super().__init__() + self._headless = headless + self._fullscreen = fullscreen + self._cuda_context = cuda_context + self._cuda_device_ordinal = cuda_device_ordinal + self._hololink_channel = hololink_channel + self._camera = camera + self._camera_mode = camera_mode + self._frame_limit = frame_limit + self._stream_id = stream_id + + def compose(self): + logging.info("compose") + if self._frame_limit: + self._count = holoscan.conditions.CountCondition( + self, + name="count", + count=self._frame_limit, + ) + condition = self._count + else: + self._ok = holoscan.conditions.BooleanCondition( + self, name="ok", enable_tick=True + ) + condition = self._ok + self._camera.set_mode(self._camera_mode) + + # image_decoder_allocator_pool = holoscan.resources.BlockMemoryPool( + # self, + # name="pool", + # # storage_type of 1 is device memory + # storage_type=1, + # block_size=self._camera._width + # * ctypes.sizeof(ctypes.c_uint16) * 2 + # * self._camera._height, + # num_blocks=2, + # ) + image_decoder_allocator_pool = holoscan.resources.UnboundedAllocator(self) + + image_decoder = hololink_module.operators.ImageDecoderOp( + self, + name="image_decoder", + out_tensor_name="output", + allocator=image_decoder_allocator_pool, + cuda_device_ordinal=self._cuda_device_ordinal, + ) + self._camera.configure_converter(image_decoder) + + frame_size = image_decoder.get_csi_length() + frame_context = self._cuda_context + receiver_operator = hololink_module.operators.LinuxReceiverOperator( + self, + condition, + name="receiver", + frame_size=frame_size, + frame_context=frame_context, + udp_port=54739 + self._stream_id, + hololink_channel=self._hololink_channel, + device=self._camera, + ) + + visualizer = holoscan.operators.HolovizOp( + self, + name="holoviz", + fullscreen=self._fullscreen, + headless=self._headless, + framebuffer_srgb=False, + ) + + self.add_flow(receiver_operator, image_decoder, {("output", "input")}) + self.add_flow(image_decoder, visualizer, {("output", "receivers")}) + + +def main(): + parser = argparse.ArgumentParser() + modes = hololink_module.sensors.d555.d555_mode.RealSense_Mode + mode_choices = [mode.value for mode in modes] + mode_help = " ".join([f"{mode.value}:{mode.name}" for mode in modes]) + parser.add_argument( + "--camera-mode", + type=int, + choices=mode_choices, + default=mode_choices[4], + help=mode_help, + ) + parser.add_argument("--headless", action="store_true", help="Run in headless mode") + parser.add_argument( + "--fullscreen", action="store_true", help="Run in fullscreen mode" + ) + parser.add_argument( + "--frame-limit", + type=int, + default=None, + help="Exit after receiving this many frames", + ) + default_configuration = os.path.join( + os.path.dirname(__file__), "example_configuration.yaml" + ) + parser.add_argument( + "--configuration", + default=default_configuration, + help="Configuration file", + ) + parser.add_argument( + "--hololink", + default="192.168.0.2", + help="IP address of Hololink board", + ) + parser.add_argument( + "--log-level", + type=int, + default=20, + help="Logging level to display", + ) + parser.add_argument( + "--expander-configuration", + type=int, + default=0, + choices=(0, 1), + help="I2C Expander configuration", + ) + parser.add_argument( + "--pattern", + type=int, + choices=range(12), + help="Configure to display a test pattern.", + ) + args = parser.parse_args() + hololink_module.logging_level(args.log_level) + logging.getLogger().setLevel(args.log_level) + logging.info("Initializing.") + # Get a handle to the GPU + (cu_result,) = cuda.cuInit(0) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + cu_device_ordinal = 0 + cu_result, cu_device = cuda.cuDeviceGet(cu_device_ordinal) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + cu_result, cu_context = cuda.cuDevicePrimaryCtxRetain(cu_device) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + # # Get a handle to the Hololink device + channel_metadata = hololink_module.Enumerator.find_channel(channel_ip=args.hololink) + overrides = hololink_module.Metadata( + { + "vsync_enable": 0, # or 1 + "block_enable": 0, # or 1 + } + ) + channel_metadata.update(overrides) + hololink_channel = hololink_module.DataChannel(channel_metadata) + # # Get a handle to the camera + logging.info("camera mode: %s", args.camera_mode) + camera_mode = hololink_module.sensors.d555.d555_mode.RealSense_Mode( + args.camera_mode % hololink_module.sensors.d555.d555_mode.PROFILE_COUNT + ) + stream_id = ( + hololink_module.sensors.d555.d555_mode.RealSense_StreamId.DEPTH + if args.camera_mode < hololink_module.sensors.d555.d555_mode.PROFILE_COUNT + else hololink_module.sensors.d555.d555_mode.RealSense_StreamId.RGB + ) + camera = hololink_module.sensors.d555.d555.RealsenseCamD555( + hololink_channel, stream_id + ) + + # Set up the application + application = HoloscanApplication( + args.headless, + args.fullscreen, + cu_context, + cu_device_ordinal, + hololink_channel, + camera, + camera_mode, + args.frame_limit, + stream_id.value, + ) + application.config(args.configuration) + # Run it. + hololink = hololink_channel.hololink() + hololink.start() + camera.configure(camera_mode) + os.environ["GXF_MEMORY_DEBUG"] = "1" + application.run() + hololink.stop() + + (cu_result,) = cuda.cuDevicePrimaryCtxRelease(cu_device) + assert cu_result == cuda.CUresult.CUDA_SUCCESS + + +if __name__ == "__main__": + main() diff --git a/python/hololink/hololink.cpp b/python/hololink/hololink.cpp index 054f7698..0fa3badb 100644 --- a/python/hololink/hololink.cpp +++ b/python/hololink/hololink.cpp @@ -466,7 +466,7 @@ PYBIND11_MODULE(_hololink, m) .def("unconfigure", &DataChannel::unconfigure) .def_static("use_multicast", &DataChannel::use_multicast, "metadata"_a, "address"_a, "port"_a) .def_static("use_broadcast", &DataChannel::use_broadcast, "metadata"_a, "port"_a) - .def("configure_socket", &DataChannel::configure_socket, "socket_fd"_a) + .def("configure_socket", &DataChannel::configure_socket, py::arg("socket_fd"), py::arg("udp_port") = 0) .def_static("use_sensor", &DataChannel::use_sensor, "metadata"_a, "sensor_number"_a) .def("frame_end_sequencer", &DataChannel::frame_end_sequencer) .def_static("use_data_plane_configuration", &DataChannel::use_data_plane_configuration, "metadata"_a, "data_plane"_a) diff --git a/python/hololink/operators/CMakeLists.txt b/python/hololink/operators/CMakeLists.txt index eae8eaea..fa9822aa 100644 --- a/python/hololink/operators/CMakeLists.txt +++ b/python/hololink/operators/CMakeLists.txt @@ -38,3 +38,4 @@ if(HOLOLINK_BUILD_SIPL) add_subdirectory(sipl_capture) endif() add_subdirectory(udp_transmitter) +add_subdirectory(image_decoder) \ No newline at end of file diff --git a/python/hololink/operators/__init__.py b/python/hololink/operators/__init__.py index 11af3b81..5676e95c 100644 --- a/python/hololink/operators/__init__.py +++ b/python/hololink/operators/__init__.py @@ -27,6 +27,7 @@ "CsiToBayerOp": "csi_to_bayer", "FusaCoeCaptureOp": "fusa_coe_capture", "ImageProcessorOp": "image_processor", + "ImageDecoderOp": "image_decoder", "ImageShiftToUint8Operator": "image_shift_to_uint8_operator", "IQDecoderOp": "iq_dec", "IQEncoderOp": "iq_enc", diff --git a/python/hololink/operators/base_receiver_op.py b/python/hololink/operators/base_receiver_op.py index a7fc1549..b10732da 100644 --- a/python/hololink/operators/base_receiver_op.py +++ b/python/hololink/operators/base_receiver_op.py @@ -58,6 +58,7 @@ def __init__( def setup(self, spec): logging.info("setup") + spec.param("udp_port", 0) spec.output("output") def start(self): diff --git a/python/hololink/operators/image_decoder/CMakeLists.txt b/python/hololink/operators/image_decoder/CMakeLists.txt new file mode 100644 index 00000000..9f24622f --- /dev/null +++ b/python/hololink/operators/image_decoder/CMakeLists.txt @@ -0,0 +1,8 @@ +include(pybind11_add_hololink_module) + +pybind11_add_hololink_module( + CPP_CMAKE_TARGET image_decoder + CLASS_NAME "ImageDecoderOp" + IMPORT "import holoscan.core" + SOURCES image_decoder.cpp +) diff --git a/python/hololink/operators/image_decoder/image_decoder.cpp b/python/hololink/operators/image_decoder/image_decoder.cpp new file mode 100644 index 00000000..63275d0c --- /dev/null +++ b/python/hololink/operators/image_decoder/image_decoder.cpp @@ -0,0 +1,102 @@ +/* + * SPDX-FileCopyrightText: Copyright (c) 2024 NVIDIA CORPORATION & AFFILIATES. All rights reserved. + * SPDX-License-Identifier: Apache-2.0 + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include + +#include +#include // for unordered_map -> dict, etc. + +#include +#include +#include + +#include +#include +#include +#include + +using std::string_literals::operator""s; +using pybind11::literals::operator""_a; + +#define STRINGIFY(x) #x +#define MACRO_STRINGIFY(x) STRINGIFY(x) + +namespace py = pybind11; + +namespace hololink::operators { + +/* Trampoline classes for handling Python kwargs + * + * These add a constructor that takes a Fragment for which to initialize the operator. + * The explicit parameter list and default arguments take care of providing a Pythonic + * kwarg-based interface with appropriate default values matching the operator's + * default parameters in the C++ API `setup` method. + * + * The sequence of events in this constructor is based on Fragment::make_operator + */ +class PyImageDecoder : public ImageDecoder { +public: + /* Inherit the constructors */ + using ImageDecoder::ImageDecoder; + + // Define a constructor that fully initializes the object. + PyImageDecoder(holoscan::Fragment* fragment, + const std::shared_ptr& allocator, int cuda_device_ordinal, + const std::string& name = "image_decoder", + const std::string& out_tensor_name = "", + const bool align_depth_to_rgb = false) + : ImageDecoder(holoscan::ArgList { holoscan::Arg { "allocator", allocator }, + holoscan::Arg { "cuda_device_ordinal", cuda_device_ordinal }, + holoscan::Arg { "out_tensor_name", out_tensor_name }, holoscan::Arg { "align_depth_to_rgb", align_depth_to_rgb } }) + { + name_ = name; + fragment_ = fragment; + spec_ = std::make_shared(fragment); + setup(*spec_.get()); + } +}; + +PYBIND11_MODULE(_image_decoder, m) +{ +#ifdef VERSION_INFO + m.attr("__version__") = MACRO_STRINGIFY(VERSION_INFO); +#else + m.attr("__version__") = "dev"; +#endif + + auto op = py::class_>(m, "ImageDecoderOp") + .def(py::init&, + int, const std::string&, const std::string&, const bool>(), + "fragment"_a, "allocator"_a, "cuda_device_ordinal"_a = 0, + "name"_a = "image_decoder"s, "out_tensor_name"_a = ""s, + "align_depth_to_rgb"_a = false) + .def("setup", &ImageDecoder::setup, "spec"_a) + .def("configure", &ImageDecoder::configure, "width"_a, "height"_a, + "pixel_format"_a, "frame_start_size"_a, "frame_end_size"_a, + "line_start_size"_a, "line_end_size"_a, "margin_left"_a = 0, + "margin_top"_a = 0, "margin_right"_a = 0, "margin_bottom"_a = 0) + .def("get_csi_length", &ImageDecoder::get_csi_length); + + py::enum_(op, "PixelFormat") + .value("Z16", ImageDecoder::PixelFormat::Z16, R"pbdoc(Z16-bit)pbdoc") + .value("YUYV", ImageDecoder::PixelFormat::YUYV, R"pbdoc(YUYV-bit)pbdoc") + .export_values(); + +} // PYBIND11_MODULE + +} // namespace hololink::operators diff --git a/python/hololink/operators/linux_receiver_operator.py b/python/hololink/operators/linux_receiver_operator.py index 6023fe53..d80baa0f 100644 --- a/python/hololink/operators/linux_receiver_operator.py +++ b/python/hololink/operators/linux_receiver_operator.py @@ -69,7 +69,7 @@ def __init__( def _start_receiver(self): self._check_buffer_size(self._frame_size) - self._hololink_channel.configure_socket(self._data_socket.fileno()) + self._hololink_channel.configure_socket(self._data_socket.fileno(), self.udp_port) self._receiver = hololink_module.operators.LinuxReceiver( self._frame_memory, self._frame_size, diff --git a/python/hololink/sensors/__init__.py b/python/hololink/sensors/__init__.py index 2521eee5..781ff56a 100644 --- a/python/hololink/sensors/__init__.py +++ b/python/hololink/sensors/__init__.py @@ -26,6 +26,7 @@ "imx477", "imx715", "vb1940", + "d555" ] _OBJECTS = { diff --git a/python/hololink/sensors/d555/__init__.py b/python/hololink/sensors/d555/__init__.py new file mode 100644 index 00000000..86e53d06 --- /dev/null +++ b/python/hololink/sensors/d555/__init__.py @@ -0,0 +1,21 @@ +# SPDX-FileCopyrightText: Copyright (c) 2023 NVIDIA CORPORATION & AFFILIATES. All rights reserved. +# SPDX-License-Identifier: Apache-2.0 +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +from . import d555, d555_mode + +__all__ = [ + "d555", + "d555_mode", +] diff --git a/python/hololink/sensors/d555/d555.py b/python/hololink/sensors/d555/d555.py new file mode 100644 index 00000000..7f69da5c --- /dev/null +++ b/python/hololink/sensors/d555/d555.py @@ -0,0 +1,196 @@ +""" +SPDX-FileCopyrightText: Copyright (c) 2023-2024 NVIDIA CORPORATION & AFFILIATES. All rights reserved. +SPDX-License-Identifier: Apache-2.0 +""" + +import logging +import time + +import hololink as hololink_module +from .model import Endianness, DataWidth + +from . import d555_mode +from .d555_mode import RealSense_StreamCommand, RealSense_StreamId + +# Camera info +DRIVER_NAME = "REALSENSE-D555" +VERSION = 1 + +MUX_I2C_ADDR = 0x1A + + +class RealsenseCamD555: + def __init__( + self, + hololink_channel, + stream_id=RealSense_StreamId.DEPTH, + i2c_controller_address=hololink_module.I2C_CTRL, + ): + self._hololink = hololink_channel.hololink() + self._i2c_controller_address = i2c_controller_address + + # default values + self._running = False + self._pixel_format = hololink_module.operators.ImageDecoderOp.PixelFormat.Z16 + self._width = 640 + self._height = 360 + self._mode = 0 + self._stream_id = stream_id + self._stream_profile = 0 + + def setup_clock(self): + pass + + def set_mode(self, realsense_mode): + logging.info(f"[Realsense Camera] Mode set to: {realsense_mode}") + self._mode = realsense_mode + mode_index = realsense_mode.value + + if self._stream_id == RealSense_StreamId.DEPTH: + profiles = d555_mode.depth_stream_profiles + elif self._stream_id == RealSense_StreamId.RGB: + profiles = d555_mode.rgb_stream_profiles + else: + logging.error("Incorrect mode for Realsense D555 camera.") + self._mode = -1 + return + + self._stream_profile = mode_index + stream_info = profiles[mode_index] + self._height = stream_info.height + self._width = stream_info.width + self._pixel_format = stream_info.pixel_format + + def power_on(self): + """Enable power to RealSense""" + time.sleep(0.1) + + def configure(self, camera_mode): + """Configure the camera (if needed)""" + self.power_on() + logging.info("Sending RealSense config") + + # configure the camera based on the mode + self.configure_camera(camera_mode) + + def start_camera_stream(self): + """Start Streaming""" + logging.info("RealSenseCam: start_camera_stream()") + + data_high = (self._stream_id.value << 8) | RealSense_StreamCommand.SET_PROFILE.value + data_low = self._stream_profile + self.set_register(address=MUX_I2C_ADDR, register=data_low, value=data_high,\ + reg_size=DataWidth.BITS_16, val_size=DataWidth.BITS_16, endian=Endianness.LITTLE) + + data_high = (self._stream_id.value << 8) | RealSense_StreamCommand.START_STREAM.value + data_low = 0 + self.set_register(address=MUX_I2C_ADDR, register=data_low, value=data_high,\ + reg_size=DataWidth.BITS_16, val_size=DataWidth.BITS_16, endian=Endianness.LITTLE) + + return True + + def stop_camera_stream(self): + """Stop Streaming""" + logging.info("RealSenseCam: stop_camera_stream()") + data_high = (self._stream_id.value << 8) | RealSense_StreamCommand.STOP_STREAM.value + data_low = 0 + self.set_register(address=MUX_I2C_ADDR, register=data_low, value=data_high,\ + reg_size=DataWidth.BITS_16, val_size=DataWidth.BITS_16, endian=Endianness.LITTLE) + + + def start(self): + """Start Streaming""" + logging.info("RealSenseCam: start()") + + # + # Setting these register is time-consuming. + logging.info( + "RealSenseCam: Stopping any previous streaming before starting new one." + ) + + # If the camera stream is already running, stop it first + # self.stop_camera_stream() + + logging.info("RealSenseCam: Attempting to start streaming...") + + self.start_camera_stream() + self._running = True + + def stop(self): + logging.info("RealSenseCam: stop()") + self.stop_camera_stream() + self._running = False + + def set_register( + self, + address, + register, + value, + reg_size: DataWidth, + val_size: DataWidth, + endian: Endianness, + ): + logging.debug( + f"WRITE >> address=0x{int(address):X} set_register(register=0x{int(register):04X}, value=0x{int(value):04X})" + ) + + write_bytes = bytearray(4) + serializer = hololink_module.Serializer(write_bytes) + + if reg_size == DataWidth.BITS_32: + if endian == Endianness.BIG: + serializer.append_uint32_be(register) + else: + serializer.append_uint32_le(register) + elif reg_size == DataWidth.BITS_16: + if endian == Endianness.BIG: + serializer.append_uint16_be(register) + else: + serializer.append_uint16_le(register) + elif reg_size == DataWidth.BITS_8: + serializer.append_uint8(register) + + if val_size == DataWidth.BITS_32: + if endian == Endianness.BIG: + serializer.append_uint32_be(value) + else: + serializer.append_uint32_le(value) + elif val_size == DataWidth.BITS_16: + if endian == Endianness.BIG: + serializer.append_uint16_be(value) + else: + serializer.append_uint16_le(value) + elif val_size == DataWidth.BITS_8: + serializer.append_uint8(value) + + + reg_data_buffer_ = self._i2c_controller_address + 16 # Offset for register data buffer + value = int.from_bytes(write_bytes[: serializer.length()], byteorder='little') + self._hololink.write_uint32(reg_data_buffer_, value, timeout=None) + + + def configure_converter(self, converter): + logging.debug(f"[Realsense] Configuring converter") + + converter.configure( + self._width, + self._height, + self._pixel_format, + 0, + 0, + 0, + 0, + ) + + def configure_camera(self, realsense_mode): + """Configure the camera with the specified mode.""" + self.set_mode(realsense_mode) + + def set_digital_gain_reg(self, val): + logging.info(f"[Realsense] Digital gain set to: {val}") + + def pixel_format(self): + return self._pixel_format + + def test_pattern(self, enable=False): + logging.info("Test pattern control is not implemented for RealSense.") diff --git a/python/hololink/sensors/d555/d555_mode.py b/python/hololink/sensors/d555/d555_mode.py new file mode 100644 index 00000000..8779047b --- /dev/null +++ b/python/hololink/sensors/d555/d555_mode.py @@ -0,0 +1,118 @@ +from collections import namedtuple +from enum import Enum +from enum import IntEnum + + +import hololink as hololink_module + + +class RealSense_StreamId(IntEnum): + RGB = 0 + DEPTH = 2 + +class RealSense_StreamCommand(IntEnum): + START_STREAM = 1 + STOP_STREAM = 2 + SET_PROFILE = 3 + +# Define Enum +class RealSense_RGB_Mode(Enum): + RGB_896x504_30FPS = 0 # 17 in single stream + RGB_896x504_15FPS = 1 # 18 + RGB_896x504_5FPS = 2 # 19 + RGB_896x504_60FPS = 3 # 20 + RGB_1280x800_30FPS = 4 # 21 + RGB_1280x800_15FPS = 5 # 22 + RGB_1280x720_30FPS = 6 # 23 + RGB_1280x720_15FPS = 7 # 24 + RGB_1280x720_5FPS = 8 # 25 + RGB_640x360_60FPS = 9 # 26 + RGB_640x360_30FPS = 10 # 27 + RGB_640x360_15FPS = 11 # 28 + RGB_640x360_5FPS = 12 # 29 + RGB_448x252_60FPS = 13 # 30 + RGB_448x252_30FPS = 14 # 31 + RGB_448x252_15FPS = 15 # 32 + RGB_448x252_5FPS = 16 # 33 + +class RealSense_Depth_Mode(Enum): + DEPTH_896x504_30FPS = 0 + DEPTH_896x504_15FPS = 1 + DEPTH_896x504_5FPS = 2 + DEPTH_896x504_60FPS = 3 + DEPTH_1280x720_30FPS = 4 + DEPTH_1280x720_15FPS = 5 + DEPTH_1280x720_5FPS = 6 + DEPTH_640x360_60FPS = 7 + DEPTH_640x360_30FPS = 8 + DEPTH_640x360_15FPS = 9 + DEPTH_640x360_5FPS = 10 + DEPTH_448x252_60FPS = 11 + DEPTH_448x252_30FPS = 12 + DEPTH_448x252_15FPS = 13 + DEPTH_448x252_5FPS = 14 + DEPTH_1280x800_15FPS = 15 + DEPTH_256x144_90FPS = 16 + + +PROFILE_COUNT = 17 # Total number of profiles for RealSense D555 camera per stream + +# Merge safely at class creation time +def create_combined_enum(name, *enums): + combined = {} + i = 0 + for enum_cls in enums: + for member in enum_cls: + combined[member.name] = i + i += 1 + return Enum(name, combined) + +RealSense_Mode = create_combined_enum("RealSense_Mode", RealSense_Depth_Mode, RealSense_RGB_Mode) + +# Define namedtuple +stream_info = namedtuple( + "stream_profile", + ["width", "height", "framerate", "pixel_format"] +) + +# Mapping +depth_stream_profiles = [] + +depth_profiles = [ + (896, 504, 30), (896, 504, 15), (896, 504, 5), (896, 504, 60), + (1280, 720, 30), (1280, 720, 15), (1280, 720, 5), + (640, 360, 60), (640, 360, 30), (640, 360, 15), (640, 360, 5), + (448, 252, 60), (448, 252, 30), (448, 252, 15), (448, 252, 5), + (1280, 800, 15), (256, 144, 90), +] + +for i, (w, h, fps) in enumerate(depth_profiles): + depth_stream_profiles.insert( + i, + stream_info( + w, h, fps, + hololink_module.operators.ImageDecoderOp.PixelFormat.Z16 + ) + ) + +# RGB profiles has same depth profiles but with different pixel format + +rgb_stream_profiles = [] + +rgb_profiles = [ + (896, 504, 30), (896, 504, 15), (896, 504, 5), (896, 504, 60), + (1280, 800, 30), (1280, 800, 15), + (1280, 720, 30), (1280, 720, 15), (1280, 720, 5), + (640, 360, 60), (640, 360, 30), (640, 360, 15), (640, 360, 5), + (448, 252, 60), (448, 252, 30), (448, 252, 15), (448, 252, 5), +] + + +for i, (w, h, fps) in enumerate(rgb_profiles): + rgb_stream_profiles.insert( + i, + stream_info( + w, h, fps, + hololink_module.operators.ImageDecoderOp.PixelFormat.YUYV, + ) + ) diff --git a/python/hololink/sensors/d555/model.py b/python/hololink/sensors/d555/model.py new file mode 100644 index 00000000..a5cb284b --- /dev/null +++ b/python/hololink/sensors/d555/model.py @@ -0,0 +1,10 @@ +from enum import Enum + +class Endianness(Enum): + LITTLE = "little" + BIG = "big" + +class DataWidth(Enum): + BITS_8 = 1 + BITS_16 = 2 + BITS_32 = 4 diff --git a/python/setup.py b/python/setup.py index 1b432745..33dad93a 100644 --- a/python/setup.py +++ b/python/setup.py @@ -57,6 +57,7 @@ def initialize_options(self): "hololink/sensors/imx715", "hololink/sensors/vb1940", "hololink/sensors/ecam0m30tof", + "hololink/sensors/d555", "tools", ], ext_modules=[ diff --git a/src/hololink/core/data_channel.cpp b/src/hololink/core/data_channel.cpp index 177ada7e..7a2b0b6b 100644 --- a/src/hololink/core/data_channel.cpp +++ b/src/hololink/core/data_channel.cpp @@ -299,7 +299,7 @@ void DataChannel::unconfigure() packetizer_program_->disable(*hololink_, sif_address_); } -void DataChannel::configure_socket(int socket_fd) +void DataChannel::configure_socket(int socket_fd, uint16_t udp_port) { const std::string& peer_ip = this->peer_ip(); auto [local_ip, local_device, local_mac] = core::local_ip_and_mac(peer_ip); @@ -363,7 +363,7 @@ void DataChannel::configure_socket(int socket_fd) // Not multicast; use bind(local_ip,0) so that the kernel assigns us a UDP port. sockaddr_in address {}; address.sin_family = AF_INET; - address.sin_port = htons(0); + address.sin_port = htons(udp_port); if (inet_pton(AF_INET, local_ip.c_str(), &address.sin_addr) != 1) { throw std::runtime_error( fmt::format("Failed to convert address {}", local_ip)); diff --git a/src/hololink/core/data_channel.hpp b/src/hololink/core/data_channel.hpp index 5d337d33..29bd470e 100644 --- a/src/hololink/core/data_channel.hpp +++ b/src/hololink/core/data_channel.hpp @@ -134,7 +134,7 @@ class DataChannel { * Configure the receiver to handle this traffic; this * is useful when using, say, multicast. */ - void configure_socket(int socket_fd); + void configure_socket(int socket_fd, uint16_t udp_port = 0); /** * Configure the given metadata to exchange data from the diff --git a/src/hololink/operators/CMakeLists.txt b/src/hololink/operators/CMakeLists.txt index 2d272d3a..9f9264ff 100644 --- a/src/hololink/operators/CMakeLists.txt +++ b/src/hololink/operators/CMakeLists.txt @@ -70,3 +70,4 @@ if(HOLOLINK_BUILD_SIPL) endif() add_subdirectory(udp_transmitter) add_subdirectory(emulator) +add_subdirectory(image_decoder) diff --git a/src/hololink/operators/image_decoder/CMakeLists.txt b/src/hololink/operators/image_decoder/CMakeLists.txt new file mode 100644 index 00000000..4b733423 --- /dev/null +++ b/src/hololink/operators/image_decoder/CMakeLists.txt @@ -0,0 +1,32 @@ +# SPDX-License-Identifier: Apache-2.0 +add_library(image_decoder STATIC + image_decoder.cpp +) + +set_property(TARGET image_decoder PROPERTY POSITION_INDEPENDENT_CODE ON) + +add_library(hololink::operators::image_decoder ALIAS image_decoder) + +target_include_directories(image_decoder + INTERFACE + $ + $ +) + +target_link_libraries(image_decoder + PUBLIC + hololink::core + # realsense2 + hololink + PRIVATE + holoscan::core +) + +# Installation of the image_decoder operator +install(TARGETS image_decoder + DESTINATION ${CMAKE_INSTALL_LIBDIR} + COMPONENT hololink-operators) + +install(FILES ${CMAKE_CURRENT_SOURCE_DIR}/image_decoder.hpp + DESTINATION ${CMAKE_INSTALL_INCLUDEDIR}/operators/image_decoder + COMPONENT hololink-operators) \ No newline at end of file diff --git a/src/hololink/operators/image_decoder/image_decoder.cpp b/src/hololink/operators/image_decoder/image_decoder.cpp new file mode 100644 index 00000000..c1a4df97 --- /dev/null +++ b/src/hololink/operators/image_decoder/image_decoder.cpp @@ -0,0 +1,340 @@ +// SPDX-FileCopyrightText: Copyright (c) 2024 NVIDIA CORPORATION & AFFILIATES. All rights reserved. +// SPDX-License-Identifier: Apache-2.0 + +#include "image_decoder.hpp" + +#include +#include +#include +#include + +namespace { +const char* source = R"( +extern "C" { + +typedef unsigned char uint8_t; +typedef unsigned short uint16_t; +typedef unsigned int uint32_t; + +__global__ void frameReconstructionZ16(unsigned short* out, + const unsigned char* in, + int per_line_size, + int width, + int height) +{ + int idx_x = blockIdx.x * blockDim.x + threadIdx.x; + int idx_y = blockIdx.y * blockDim.y + threadIdx.y; + if ((idx_x >= width) || (idx_y >= height)) return; + int out_index = idx_y * width + idx_x; + int in_index = (per_line_size * idx_y) + idx_x * 2; + unsigned short val = static_cast(in[in_index]) | + (static_cast(in[in_index + 1]) << 8); + out[out_index] = val; +} + +__global__ void frameReconstructionYUYV(uint8_t* out_rgb, + const uint8_t* in_yuyv, + int per_line_size, + int width, + int height) { + int x = blockIdx.x * blockDim.x + threadIdx.x; + int y = blockIdx.y * blockDim.y + threadIdx.y; + if ((x >= width) || (y >= height)) return; + int pixel_pair_idx = x / 2; + int in_idx = y * per_line_size + pixel_pair_idx * 4; + uint8_t Y0 = in_yuyv[in_idx + 0]; + uint8_t U = in_yuyv[in_idx + 1]; + uint8_t Y1 = in_yuyv[in_idx + 2]; + uint8_t V = in_yuyv[in_idx + 3]; + int c = x % 2; + uint8_t Y = (c == 0) ? Y0 : Y1; + int C = Y - 16, D = U - 128, E = V - 128; + int R = (298 * C + 409 * E + 128) >> 8; + int G = (298 * C - 100 * D - 208 * E + 128) >> 8; + int B = (298 * C + 516 * D + 128) >> 8; + R = R < 0 ? 0 : (R > 255 ? 255 : R); + G = G < 0 ? 0 : (G > 255 ? 255 : G); + B = B < 0 ? 0 : (B > 255 ? 255 : B); + int out_idx = (y * width + x) * 3; + out_rgb[out_idx + 0] = R; + out_rgb[out_idx + 1] = G; + out_rgb[out_idx + 2] = B; + +} + +__global__ void compute_histogram(const uint16_t* depth, int* hist, int size) { + int idx = blockIdx.x * blockDim.x + threadIdx.x; + if (idx >= size) return; + uint16_t d = depth[idx]; + if (d > 0 && d < 65536) atomicAdd(&hist[d], 1); +} + +__global__ void prefix_sum_histogram(int* hist, int size) { + for (int i = 1; i < size; ++i) { + hist[i] += hist[i - 1]; + } +} + +__device__ inline float3 interpolate_colormap(float value, const float3* colormap, int colormap_size) { + float t = fminf(fmaxf(value, 0.f), 1.f) * (colormap_size - 1); + int idx = (int)t; + float frac = t - idx; + float3 lo = colormap[idx]; + float3 hi = colormap[min(idx + 1, colormap_size - 1)]; + return make_float3(lo.x * (1.f - frac) + hi.x * frac, + lo.y * (1.f - frac) + hi.y * frac, + lo.z * (1.f - frac) + hi.z * frac); +} + +__global__ void depthToRGB(uint8_t* out_rgb, + const uint16_t* depth, + const int* hist, + int width, + int height, + float depth_units, + float min_m, + float max_m, + bool equalize, + const float3* colormap, + int colormap_size) { + int x = blockIdx.x * blockDim.x + threadIdx.x; + int y = blockIdx.y * blockDim.y + threadIdx.y; + if (x >= width || y >= height) return; + + int idx = y * width + x; + uint16_t d = depth[idx]; + + if (d == 0) { + out_rgb[3 * idx + 0] = 0; + out_rgb[3 * idx + 1] = 0; + out_rgb[3 * idx + 2] = 0; + return; + } + + float norm; + if (equalize) { + int total_hist = hist[65535]; + norm = (total_hist > 0) ? (float)(hist[d]) / total_hist : 0.f; + } else { + float depth_m = d * depth_units; + norm = (depth_m - min_m) / (max_m - min_m); + norm = fminf(fmaxf(norm, 0.f), 1.f); + } + + float3 c = interpolate_colormap(norm, colormap, colormap_size); + out_rgb[3 * idx + 0] = (uint8_t)(c.x); + out_rgb[3 * idx + 1] = (uint8_t)(c.y); + out_rgb[3 * idx + 2] = (uint8_t)(c.z); +} + +} +)"; +} // namespace + +namespace hololink::operators { + +void ImageDecoder::setup(holoscan::OperatorSpec& spec) { + spec.input("input"); + spec.output("output"); + spec.param(allocator_, "allocator", "Allocator", "Memory allocator"); + spec.param(cuda_device_ordinal_, "cuda_device_ordinal", "CudaDeviceOrdinal", "CUDA device"); + spec.param(out_tensor_name_, "out_tensor_name", "OutputTensorName", "Name of output tensor"); + cuda_stream_handler_.define_params(spec); +} + +void ImageDecoder::start() { + if (pixel_format_ == PixelFormat::INVALID) throw std::runtime_error("Decoder not configured"); + CudaCheck(cuInit(0)); + CudaCheck(cuDeviceGet(&cuda_device_, cuda_device_ordinal_.get())); + CudaCheck(cuDevicePrimaryCtxRetain(&cuda_context_, cuda_device_)); + hololink::common::CudaContextScopedPush cur_cuda_context(cuda_context_); + cuda_function_launcher_.reset(new hololink::common::CudaFunctionLauncher( + source, {"frameReconstructionZ16", "frameReconstructionYUYV","depthToRGB", "compute_histogram", + "prefix_sum_histogram"})); + + // Allocate d_hist_ (256KB) + cudaMalloc(&d_hist_, sizeof(int) * 0x10000); + + // Allocate and upload colormap + std::vector colormap = { + {0.f, 0.f, 255.f}, // Blue + {0.f, 255.f, 255.f}, // Cyan + {255.f, 255.f, 0.f}, // Yellow + {255.f, 0.f, 0.f}, // Red + {50.f, 0.f, 0.f} // Dark red + }; + colormap_size_ = colormap.size(); + cudaMalloc(&d_colormap_, colormap_size_ * sizeof(float3)); + cudaMemcpy(d_colormap_, colormap.data(), colormap_size_ * sizeof(float3), cudaMemcpyHostToDevice); +} + +void ImageDecoder::stop() { + hololink::common::CudaContextScopedPush cur_cuda_context(cuda_context_); + cuda_function_launcher_.reset(); + + if (d_hist_) { + cudaFree(d_hist_); + d_hist_ = nullptr; + } + + if (d_colormap_) { + cudaFree(d_colormap_); + d_colormap_ = nullptr; + colormap_size_ = 0; + } + + CudaCheck(cuDevicePrimaryCtxRelease(cuda_device_)); + cuda_context_ = nullptr; +} + +void ImageDecoder::compute(holoscan::InputContext& input, holoscan::OutputContext& output, + holoscan::ExecutionContext& context) { + auto maybe_entity = input.receive("input"); + if (!maybe_entity) throw std::runtime_error("No input entity"); + auto& entity = static_cast(maybe_entity.value()); + gxf_result_t stream_handler_result = cuda_stream_handler_.from_message(context.context(), entity); + if (stream_handler_result != GXF_SUCCESS) throw std::runtime_error("Failed to get stream"); + auto input_tensor = entity.get().value(); + + if (input_tensor->storage_type() == nvidia::gxf::MemoryStorageType::kHost) { + if (!is_integrated_ && !host_memory_warning_) { + host_memory_warning_ = true; + HSB_LOG_WARN( + "The input tensor is stored in host memory, this will reduce performance of this " + "operator. For best performance store the input tensor in device memory."); + } + } else if (input_tensor->storage_type() != nvidia::gxf::MemoryStorageType::kDevice) { + throw std::runtime_error( + fmt::format("Unsupported storage type {}", (int)input_tensor->storage_type())); + } + + if (input_tensor->rank() != 1) throw std::runtime_error("Tensor must be 1D"); + + const int32_t size = input_tensor->shape().dimension(0); + auto allocator = nvidia::gxf::Handle::Create( + fragment()->executor().context(), allocator_->gxf_cid()); + const uint32_t per_line_size = line_start_size_ + bytes_per_line_ + line_end_size_; + + switch (pixel_format_) { + case PixelFormat::Z16: { + // 1. Allocate depth tensor (device) + nvidia::gxf::Shape depth_shape{int(height_), int(width_), 1}; + auto depth_message = CreateTensorMap(context.context(), allocator.value(), {{ + "depth", nvidia::gxf::MemoryStorageType::kDevice, depth_shape, + nvidia::gxf::PrimitiveType::kUnsigned16, 0, + nvidia::gxf::ComputeTrivialStrides(depth_shape, 2)}}, false); + auto depth_tensor = depth_message.value().get("depth"); + + // 2. Reconstruct depth frame from CSI + cuda_function_launcher_->launch("frameReconstructionZ16", {width_, height_, 1}, + cuda_stream_handler_.get_cuda_stream(context.context()), + depth_tensor.value()->pointer(), + input_tensor->pointer() + frame_start_size_ + line_start_size_, + per_line_size, width_, height_); + + // 3. Allocate RGB tensor (device) + nvidia::gxf::Shape rgb_shape{int(height_), int(width_), 3}; + auto out_message = CreateTensorMap(context.context(), allocator.value(), {{ + out_tensor_name_.get(), nvidia::gxf::MemoryStorageType::kDevice, rgb_shape, + nvidia::gxf::PrimitiveType::kUnsigned8, 0, + nvidia::gxf::ComputeTrivialStrides(rgb_shape, 1)}}, false); + auto rgb_tensor = out_message.value().get(out_tensor_name_.get().c_str()); + + // 4. GPU histogram + cudaMemsetAsync(d_hist_, 0, sizeof(int) * 0x10000, cuda_stream_handler_.get_cuda_stream(context.context())); + + cuda_function_launcher_->launch("compute_histogram", {width_, height_, 1}, + cuda_stream_handler_.get_cuda_stream(context.context()), + depth_tensor.value()->pointer(), d_hist_, width_ * height_); + + cuda_function_launcher_->launch("prefix_sum_histogram", {1, 1, 1}, + cuda_stream_handler_.get_cuda_stream(context.context()), + d_hist_, + 0x10000); + + // 5. Run GPU depth → RGB colorizer + cuda_function_launcher_->launch("depthToRGB", {width_, height_, 1}, + cuda_stream_handler_.get_cuda_stream(context.context()), + rgb_tensor.value()->pointer(), + depth_tensor.value()->pointer(), + d_hist_, + width_, height_, + 0.001f, + 0.3f, 4.0f, + false, + d_colormap_, + static_cast(colormap_size_)); + + + // 6. Emit output + stream_handler_result = cuda_stream_handler_.to_message(out_message); + if (stream_handler_result != GXF_SUCCESS) { + throw std::runtime_error("Failed to emit RGB image"); + } + // auto& out_entity = out_message.value(); + auto out_entity = holoscan::gxf::Entity(std::move(out_message.value())); + output.emit(out_entity); + return; + } + case PixelFormat::YUYV: { + nvidia::gxf::Shape shape{int(height_), int(width_), 3}; + auto out_message = CreateTensorMap(context.context(), allocator.value(), {{ + out_tensor_name_.get(), nvidia::gxf::MemoryStorageType::kDevice, shape, + nvidia::gxf::PrimitiveType::kUnsigned8, 0, + nvidia::gxf::ComputeTrivialStrides(shape, 1)}}, false); + auto tensor = out_message.value().get(out_tensor_name_.get().c_str()); + cuda_function_launcher_->launch("frameReconstructionYUYV", {width_, height_, 1}, + cuda_stream_handler_.get_cuda_stream(context.context()), + tensor.value()->pointer(), + input_tensor->pointer() + frame_start_size_ + line_start_size_, + per_line_size, width_, height_); + stream_handler_result = cuda_stream_handler_.to_message(out_message); + auto out_entity = holoscan::gxf::Entity(std::move(out_message.value())); + output.emit(out_entity); + return; + } + default: + throw std::runtime_error("Unsupported pixel format"); + } +} + +void ImageDecoder::configure(uint32_t width, uint32_t height, PixelFormat pixel_format, + uint32_t frame_start_size, uint32_t frame_end_size, + uint32_t line_start_size, uint32_t line_end_size, + uint32_t margin_left, uint32_t margin_top, + uint32_t margin_right, uint32_t margin_bottom) { + width_ = width; + height_ = height; + pixel_format_ = pixel_format; + frame_start_size_ = frame_start_size; + frame_end_size_ = frame_end_size; + line_start_size_ = line_start_size; + line_end_size_ = line_end_size; + switch (pixel_format_) { + case PixelFormat::Z16: + bytes_per_line_ = width * 2; + line_start_size_ += margin_left * 2; + line_end_size_ += margin_right * 2; + break; + case PixelFormat::YUYV: + bytes_per_line_ = width * 2; + line_start_size_ += margin_left * 2; + line_end_size_ += margin_right * 2; + break; + default: + throw std::runtime_error("Unsupported pixel format"); + } + const uint32_t line_size = line_start_size_ + bytes_per_line_ + line_end_size_; + frame_start_size_ += margin_top * line_size; + frame_end_size_ += margin_bottom * line_size; + csi_length_ = (frame_start_size_ + line_size * height_ + frame_end_size_ + 7) & ~7; +} + +size_t ImageDecoder::get_csi_length() { + if (pixel_format_ == PixelFormat::INVALID) { + throw std::runtime_error("ImageDecoder is not configured."); + } + return csi_length_; +} + +} // namespace hololink::operators \ No newline at end of file diff --git a/src/hololink/operators/image_decoder/image_decoder.hpp b/src/hololink/operators/image_decoder/image_decoder.hpp new file mode 100644 index 00000000..3239b4fd --- /dev/null +++ b/src/hololink/operators/image_decoder/image_decoder.hpp @@ -0,0 +1,98 @@ +/* + * SPDX-FileCopyrightText: Copyright (c) 2023 NVIDIA CORPORATION & AFFILIATES. All rights reserved. + * SPDX-License-Identifier: Apache-2.0 + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#ifndef SRC_OPERATORS_IMAGE_DECODER_IMAGE_DECODER +#define SRC_OPERATORS_IMAGE_DECODER_IMAGE_DECODER + +#include + +#include +#include +#include + +#include + +namespace hololink::common { + +class CudaFunctionLauncher; + +} // namespace hololink::common + +namespace hololink::operators { + +class ImageDecoder : public holoscan::Operator { +public: + HOLOSCAN_OPERATOR_FORWARD_ARGS(ImageDecoder); + + enum class PixelFormat { + INVALID = -1, + Z16 = 0, // Z16 is a 16-bit depth format, not a Bayer format + YUYV = 1 // YUYV is a 16-bit YUV format, not a Bayer format + }; + + struct float3 { + float x, y, z; + float3 operator*(float t) const { return {x * t, y * t, z * t}; } + float3 operator+(const float3& o) const { return {x + o.x, y + o.y, z + o.z}; } + }; + + void start() override; + void stop() override; + void setup(holoscan::OperatorSpec& spec) override; + void compute(holoscan::InputContext&, holoscan::OutputContext& op_output, + holoscan::ExecutionContext&) override; + + void configure(uint32_t width, uint32_t height, PixelFormat pixel_format, + uint32_t frame_start_size, uint32_t frame_end_size, uint32_t line_start_size, + uint32_t line_end_size, uint32_t margin_left = 0, uint32_t margin_top = 0, + uint32_t margin_right = 0, uint32_t margin_bottom = 0); + size_t get_csi_length(); + +private: + holoscan::Parameter> allocator_; + holoscan::Parameter cuda_device_ordinal_; + holoscan::Parameter out_tensor_name_; + + CUcontext cuda_context_ = nullptr; + CUdevice cuda_device_ = 0; + bool is_integrated_ = false; + bool host_memory_warning_ = false; + + holoscan::CudaStreamHandler cuda_stream_handler_; + + std::shared_ptr cuda_function_launcher_; + + uint32_t width_ = 0; + uint32_t height_ = 0; + PixelFormat pixel_format_ = PixelFormat::INVALID; + uint32_t frame_start_size_ = 0; + uint32_t frame_end_size_ = 0; + uint32_t line_start_size_ = 0; + uint32_t line_end_size_ = 0; + + uint32_t bytes_per_line_ = 0; + size_t csi_length_ = 0; + + int* d_hist_ = nullptr; + float3* d_colormap_ = nullptr; + size_t colormap_size_ = 0; + +}; + +} // namespace hololink::operators + +#endif /* SRC_OPERATORS_IMAGE_DECODER_IMAGE_DECODER */ From 1d6720703cb0d5b227e9d7e1bfd1fc0eab768e6a Mon Sep 17 00:00:00 2001 From: Ashraf Kattoura Date: Sun, 11 Jan 2026 19:27:05 +0200 Subject: [PATCH 2/4] omit align_rgb_to_depth arg --- .../hololink/operators/image_decoder/image_decoder.cpp | 10 ++++------ 1 file changed, 4 insertions(+), 6 deletions(-) diff --git a/python/hololink/operators/image_decoder/image_decoder.cpp b/python/hololink/operators/image_decoder/image_decoder.cpp index 63275d0c..7247dc8a 100644 --- a/python/hololink/operators/image_decoder/image_decoder.cpp +++ b/python/hololink/operators/image_decoder/image_decoder.cpp @@ -57,11 +57,10 @@ class PyImageDecoder : public ImageDecoder { PyImageDecoder(holoscan::Fragment* fragment, const std::shared_ptr& allocator, int cuda_device_ordinal, const std::string& name = "image_decoder", - const std::string& out_tensor_name = "", - const bool align_depth_to_rgb = false) + const std::string& out_tensor_name = "") : ImageDecoder(holoscan::ArgList { holoscan::Arg { "allocator", allocator }, holoscan::Arg { "cuda_device_ordinal", cuda_device_ordinal }, - holoscan::Arg { "out_tensor_name", out_tensor_name }, holoscan::Arg { "align_depth_to_rgb", align_depth_to_rgb } }) + holoscan::Arg { "out_tensor_name", out_tensor_name } }) { name_ = name; fragment_ = fragment; @@ -81,10 +80,9 @@ PYBIND11_MODULE(_image_decoder, m) auto op = py::class_>(m, "ImageDecoderOp") .def(py::init&, - int, const std::string&, const std::string&, const bool>(), + int, const std::string&, const std::string&>(), "fragment"_a, "allocator"_a, "cuda_device_ordinal"_a = 0, - "name"_a = "image_decoder"s, "out_tensor_name"_a = ""s, - "align_depth_to_rgb"_a = false) + "name"_a = "image_decoder"s, "out_tensor_name"_a = ""s) .def("setup", &ImageDecoder::setup, "spec"_a) .def("configure", &ImageDecoder::configure, "width"_a, "height"_a, "pixel_format"_a, "frame_start_size"_a, "frame_end_size"_a, From 2736dac12b3d39bc64451c87e9d8cd4c466fcf00 Mon Sep 17 00:00:00 2001 From: Ashraf Kattoura Date: Mon, 12 Jan 2026 14:16:19 +0200 Subject: [PATCH 3/4] Add error checking for CUDA runtime API calls --- python/hololink/sensors/d555/d555.py | 3 ++- .../operators/image_decoder/image_decoder.cpp | 12 +++++++++--- 2 files changed, 11 insertions(+), 4 deletions(-) diff --git a/python/hololink/sensors/d555/d555.py b/python/hololink/sensors/d555/d555.py index 7f69da5c..c099cc89 100644 --- a/python/hololink/sensors/d555/d555.py +++ b/python/hololink/sensors/d555/d555.py @@ -134,7 +134,8 @@ def set_register( f"WRITE >> address=0x{int(address):X} set_register(register=0x{int(register):04X}, value=0x{int(value):04X})" ) - write_bytes = bytearray(4) + buffer_size = reg_size.value + val_size.value + write_bytes = bytearray(buffer_size) serializer = hololink_module.Serializer(write_bytes) if reg_size == DataWidth.BITS_32: diff --git a/src/hololink/operators/image_decoder/image_decoder.cpp b/src/hololink/operators/image_decoder/image_decoder.cpp index c1a4df97..f9a55a61 100644 --- a/src/hololink/operators/image_decoder/image_decoder.cpp +++ b/src/hololink/operators/image_decoder/image_decoder.cpp @@ -153,7 +153,9 @@ void ImageDecoder::start() { "prefix_sum_histogram"})); // Allocate d_hist_ (256KB) - cudaMalloc(&d_hist_, sizeof(int) * 0x10000); + if (cudaMalloc(&d_hist_, sizeof(int) * 0x10000) != cudaSuccess) { + throw std::runtime_error("Failed to allocate histogram buffer"); + } // Allocate and upload colormap std::vector colormap = { @@ -164,8 +166,12 @@ void ImageDecoder::start() { {50.f, 0.f, 0.f} // Dark red }; colormap_size_ = colormap.size(); - cudaMalloc(&d_colormap_, colormap_size_ * sizeof(float3)); - cudaMemcpy(d_colormap_, colormap.data(), colormap_size_ * sizeof(float3), cudaMemcpyHostToDevice); + if (cudaMalloc(&d_colormap_, colormap_size_ * sizeof(float3)) != cudaSuccess) { + throw std::runtime_error("Failed to allocate colormap buffer"); + } + if (cudaMemcpy(d_colormap_, colormap.data(), colormap_size_ * sizeof(float3), cudaMemcpyHostToDevice) != cudaSuccess) { + throw std::runtime_error("Failed to copy colormap to device"); + } } void ImageDecoder::stop() { From b8bf703a4e45f5726880c3ca93908c5b958944c9 Mon Sep 17 00:00:00 2001 From: Ashraf Kattoura Date: Tue, 20 Jan 2026 19:35:02 +0200 Subject: [PATCH 4/4] Fix Major issues --- python/hololink/sensors/d555/d555.py | 18 ++++++++++-------- 1 file changed, 10 insertions(+), 8 deletions(-) diff --git a/python/hololink/sensors/d555/d555.py b/python/hololink/sensors/d555/d555.py index c099cc89..0d3ca69e 100644 --- a/python/hololink/sensors/d555/d555.py +++ b/python/hololink/sensors/d555/d555.py @@ -79,12 +79,14 @@ def start_camera_stream(self): data_high = (self._stream_id.value << 8) | RealSense_StreamCommand.SET_PROFILE.value data_low = self._stream_profile - self.set_register(address=MUX_I2C_ADDR, register=data_low, value=data_high,\ + self.set_register( + address=MUX_I2C_ADDR, register=data_low, value=data_high, reg_size=DataWidth.BITS_16, val_size=DataWidth.BITS_16, endian=Endianness.LITTLE) data_high = (self._stream_id.value << 8) | RealSense_StreamCommand.START_STREAM.value data_low = 0 - self.set_register(address=MUX_I2C_ADDR, register=data_low, value=data_high,\ + self.set_register( + address=MUX_I2C_ADDR, register=data_low, value=data_high, reg_size=DataWidth.BITS_16, val_size=DataWidth.BITS_16, endian=Endianness.LITTLE) return True @@ -94,7 +96,8 @@ def stop_camera_stream(self): logging.info("RealSenseCam: stop_camera_stream()") data_high = (self._stream_id.value << 8) | RealSense_StreamCommand.STOP_STREAM.value data_low = 0 - self.set_register(address=MUX_I2C_ADDR, register=data_low, value=data_high,\ + self.set_register( + address=MUX_I2C_ADDR, register=data_low, value=data_high, reg_size=DataWidth.BITS_16, val_size=DataWidth.BITS_16, endian=Endianness.LITTLE) @@ -164,14 +167,13 @@ def set_register( elif val_size == DataWidth.BITS_8: serializer.append_uint8(value) - - reg_data_buffer_ = self._i2c_controller_address + 16 # Offset for register data buffer - value = int.from_bytes(write_bytes[: serializer.length()], byteorder='little') - self._hololink.write_uint32(reg_data_buffer_, value, timeout=None) + reg_data_buffer = self._i2c_controller_address + 16 # Offset for register data buffer + packed_value = int.from_bytes(write_bytes[: serializer.length()], byteorder='little') + self._hololink.write_uint32(reg_data_buffer, packed_value, timeout=None) def configure_converter(self, converter): - logging.debug(f"[Realsense] Configuring converter") + logging.debug("[Realsense] Configuring converter") converter.configure( self._width,