diff --git a/CMakeLists.txt b/CMakeLists.txt index 26fb50f0..f4835bc0 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -158,6 +158,9 @@ if (RGL_BUILD_PCL_EXTENSION) endif() if (RGL_BUILD_ROS2_EXTENSION) + if (WIN32) # Workaround for compilation on Windows. fastcdr must be found from the top-level of cmake. + find_package(fastcdr REQUIRED) + endif() add_subdirectory(extensions/ros2) endif() diff --git a/README.md b/README.md index 0b41eb55..75c6e302 100644 --- a/README.md +++ b/README.md @@ -45,14 +45,13 @@ And more: ## Runtime requirements -|Hardware|Requirement| -|:--|:--| -|GPU|CUDA-enabled| +| Hardware | Requirement | +|:---------|:-------------| +| GPU | CUDA-enabled | -|Software|Requirement| -|:--|:--| -|Nvidia Driver (Linux)|>=515.43.04| -|Nvidia Driver (Windows)|>=472.50| +| Software | Requirement | +|:--------------|:-------------------------------------------------------------------------------------------------| +| Nvidia Driver | - Ubuntu 22.04 `>=515.43.04`
- Ubuntu 24.04 `>=555.42.02`
- Windows 10/11 `>=472.50` | ## Usage @@ -62,8 +61,8 @@ An introduction to the RGL API along with an example can be found [here](docs/Us `RobotecGPULidar` library can be built with extensions enhancing RGL with additional functions: - `PCL` - adds nodes and functions for point cloud processing that uses [Point Cloud Library](https://pointclouds.org/). See [documentation](docs/PclExtension.md). -- `ROS2` - adds a node to publish point cloud messages to [ROS2](https://www.ros.org/). Check [ROS2 extension doc](docs/Ros2Extension.md) for more information, build instructions, and usage. -- `UDP` - adds a node to publish raw lidar packets, as emitted by physical lidar. Only available in the closed-source version. +- `ROS2` - adds a nodes for point cloud publishing to [ROS2](https://www.ros.org/). See [documentation](docs/Ros2Extension.md). +- `UDP` - adds a nodes for raw lidar packets publishing, as emitted by physical lidar. Only available in the closed-source version. ## Building in Docker (Linux) @@ -94,9 +93,11 @@ docker build \ --output=build . ``` -## Building on Ubuntu 22 +## Building on Ubuntu 22/24 -1. Install [CUDA Toolkit](https://developer.nvidia.com/cuda-downloads) **11.7+**. +1. Install [CUDA Toolkit](https://developer.nvidia.com/cuda-downloads) + - Ubuntu 22.04: **11.7+** + - Ubuntu 24.04: **12.6+** 2. Download [NVidia OptiX](https://developer.nvidia.com/designworks/optix/downloads/legacy) **7.2**. 1. You may be asked to create a Nvidia account to download 3. Export environment variable: diff --git a/docs/Ros2Extension.md b/docs/Ros2Extension.md index 64b38729..8c802c93 100644 --- a/docs/Ros2Extension.md +++ b/docs/Ros2Extension.md @@ -1,9 +1,10 @@ # RGL ROS2 extension -The extension introduces the node to publish [PointCloud2](https://docs.ros2.org/foxy/api/sensor_msgs/msg/PointCloud2.html) messages to ROS2. RGL creates a ROS2 node named `RobotecGPULidar` and registers publishers based on constructed RGL nodes. +The extension introduces nodes to publish [PointCloud2](https://docs.ros2.org/foxy/api/sensor_msgs/msg/PointCloud2.html) messages to ROS2. RGL creates a ROS2 node named `RobotecGPULidar` and registers publishers based on constructed RGL nodes. Supported ROS2 distributions: - Humble +- Jazzy Supported DDS implementations for ROS2 standalone build: - Eclipse Cyclone DDS @@ -14,24 +15,29 @@ Supported DDS implementations for ROS2 standalone build: RGL ROS2 extension can be built in two flavors: - **standalone** - ROS2 installation is not required on the target machine. RGL build will include all required ROS2 dependencies. -- **overlay** - Assumes the existence of supported ROS2 installation on the target machine. RGL will try to use the existing installation of ROS2 dependencies. +- **overlay** - Assumes the existence of supported ROS2 installation on the target machine. RGL will use ROS2 libraries already installed on the system. -Before building RGL PCL extension, it is necessary to get the required dependencies. +Before building RGL ROS2 extension, it is necessary to get the required dependencies. For some, the process has been automated - run `setup.py --install-ros2-deps` to get them. -### Ubuntu 22 +### Ubuntu 22/24 #### Prerequisites -- Requirements listed in the main [README](../README.md) (section `Building on Ubuntu`). -- ROS2 installed on the system and sourced. +- Requirements listed in the main [README](../README.md#building-on-ubuntu-2224). +- ROS2 installed on the system and sourced: + - ROS2 Humble for Ubuntu 22 + - ROS2 Jazzy for Ubuntu 24 +- Radar messages installed: + ```bash + apt install -y ros-${ROS_DISTRO}-radar-msgs + ``` - For standalone build: - ROS2 `cyclonedds` and `fastrtps` packages. - `patchelf` tool. ```bash apt install -y ros-${ROS_DISTRO}-cyclonedds ros-${ROS_DISTRO}-rmw-cyclonedds-cpp apt install -y ros-${ROS_DISTRO}-fastrtps ros-${ROS_DISTRO}-rmw-fastrtps-cpp - apt install -y ros-${ROS_DISTRO}-radar-msgs apt install patchelf ``` @@ -40,45 +46,38 @@ For some, the process has been automated - run `setup.py --install-ros2-deps` to #### Steps for standalone 1. Use script `setup.py` with option `--with-ros2-standalone` - - You can specify run-time search path(s) of RGL dependencies by adding the option `--lib-rpath `. This can be useful if you want RGL to search for ROS2 libraries in a different directory than RGL, e.g., if you work with a library providing its own ROS2 libraries, such as [ROS2ForUnity](https://github.com/RobotecAI/ros2-for-unity). -2. Copy all ROS2 libraries from `/ros2_standalone/` to `libRobotecGPULidar.so` location (or location defined with `--lib-rpath`). + - You can specify run-time search path(s) of RGL dependencies by adding the option `--lib-rpath `. This can be useful if you want RGL to search for ROS2 libraries in a different directory than RGL, e.g., if you work with a library that also provides its ROS2 libraries, such as [ROS2ForUnity](https://github.com/RobotecAI/ros2-for-unity). +2. Copy all ROS2 libraries from `/lib/ros2_standalone/` to `libRobotecGPULidar.so` location (or location defined with `--lib-rpath`). + +#### Tips for integrating RGL with [AWSIM](https://github.com/tier4/AWSIM/tree/main) + +AWSIM project includes two ROS-based plugins: [ROS2ForUnity](https://github.com/RobotecAI/ros2-for-unity) and [RGLUnityPlugin](https://github.com/tier4/AWSIM/tree/main/Assets/RGLUnityPlugin). Since AWSIM is configured to run in a standalone mode (where ROS does not need to be installed on the host machine), all necessary ROS libraries are included within the project. +Ros2ForUnity and RGLUnityPlugin share a common set of ROS libraries (e.g., `librcl.so`). +To avoid duplication - since Unity does not support having multiple instances of the same libraries - one of the solutions is to store all ROS libraries in a single plugin. +In this case, the ROS libraries are stored in the ROS2ForUnity plugin, and RGL is configured with a runtime search path to access them. -#### Tips for integrating RGL + ROS2 standalone with Unity and [ROS2ForUnity](https://github.com/RobotecAI/ros2-for-unity) plugin. 1. Build RGL with the command: ```bash ./setup.py --with-ros2-standalone --lib-rpath \$ORIGIN/../../../../Ros2ForUnity/Plugins/Linux/x86_64/ ``` *$ORIGIN represents the directory in which an object (library) originated. It is resolved at run-time.*\ *Your rpath may differ. It depends on your relative location between RGLUnityPlugin and Ros2ForUnity.* -2. Copy library `/libRobotecGPULidar.so` to the appropriate directory in your RGLUnityPlugin. -3. Copy all ROS2 libraries from `/ros2_standalone/` to `Ros2ForUnity\Plugins\Linux\x86_64` directory. Skip for duplicates. - -RGL's ROS2 standalone build is dependent on ROS2ForUnity's ROS2 standalone build. RobotecGPULidar library will find ROS2 because we have set the appropriate rpath for the ROS2 libraries in ROS2ForUnity. +2. Copy library `/lib/libRobotecGPULidar.so` to the appropriate directory in your RGLUnityPlugin. +3. Copy all ROS2 libraries from `/lib/ros2_standalone/` to `Ros2ForUnity\Plugins\Linux\x86_64` directory. **Skip for duplicates**. ### Windows #### Prerequisites -- Requirements listed in the main [README](../README.md) (section `Building on Windows`). -- ROS2 installed on the system and sourced. -- Fixed ROS2 logging macros in rclcpp package to make it compile with C++20. More about this bug: [github PR](https://github.com/ros2/rclcpp/pull/2063). +- Requirements listed in the main [README](../README.md#building-on-windows). +- ROS2 installed on the system and sourced. We recommend installation from a pre-built binary package: + - [ROS2 Humble](https://docs.ros.org/en/humble/Installation/Windows-Install-Binary.html) + - [ROS2 Jazzy](https://docs.ros.org/en/jazzy/Installation/Windows-Install-Binary.html) +- [ROS2 Humble only] Fixed ROS2 logging macros in rclcpp package to make it compile with C++20. More about this bug: [github PR](https://github.com/ros2/rclcpp/pull/2063). - Use `fix_ros2_humble.py` script to apply those changes: ```bash python ros2_standalone\fix_ros2_humble.py ``` -- For standalone build: - - If you have installed ROS2 from pre-built binaries, there is `rmw_cyclonedds_cpp` package missing. You need to build this one from the source: - 1. Make sure you have installed [ROS2 prerequisites](https://docs.ros.org/en/humble/Installation/Alternatives/Windows-Development-Setup.html#installing-prerequisites). - 2. A good practice is to disable Windows path limits, [see](https://learn.microsoft.com/en-us/windows/win32/fileio/maximum-file-path-limitation?tabs=registry). - 3. Run `x64 Native Tools Command Prompt for VS 2019`, setup development folder (e.g., `C:\rosrmw`), clone ROS2 repos, and build: - ```bash - md \rosrmw\src - cd \rosrmw - vcs import --input https://raw.githubusercontent.com/ros2/ros2/humble/ros2.repos src - colcon build --merge-install --packages-up-to rmw_cyclonedds_cpp - ``` - 4. Before building RGL, source `rosrmw` workspace: `call C:\rosrmw\install\setup.bat`. - #### Steps for overlay 1. Run `x64 Native Tools Command Prompt for VS 2019` and navigate to RGL repository. @@ -87,14 +86,17 @@ RGL's ROS2 standalone build is dependent on ROS2ForUnity's ROS2 standalone build #### Steps for standalone 1. Run `x64 Native Tools Command Prompt for VS 2019` and navigate to RGL repository. 2. Run `python setup.py --with-ros2-standalone` command to build RGL with ROS2 extension and install ROS2 libraries. -3. Copy all ROS2 libraries from `/ros2_standalone/` into `RobotecGPULidar.dll` location, or extend environment variable `Path` appropriately. +3. Copy all ROS2 libraries from `/lib/ros2_standalone/` into `RobotecGPULidar.dll` location, or extend environment variable `Path` appropriately. + +#### Tips for integrating RGL with [AWSIM](https://github.com/tier4/AWSIM/tree/main) + +See description from [Ubuntu section](#tips-for-integrating-rgl-with-awsim) for more details. -#### Tips for integrating RGL + ROS2 standalone with Unity and [ROS2ForUnity](https://github.com/RobotecAI/ros2-for-unity) plugin. 1. Build RGL with ROS2 standalone as described above. -2. Copy `/RobotecGPULidar.dll` and all depend libraries located in `` to the appropriate directory in your RGLUnityPlugin. -3. Copy all ROS2 libraries from `/ros2_standalone/` to `Ros2ForUnity\Plugins\Windows\x86_64` directory. Skip for duplicates. +2. Copy `/RobotecGPULidar.dll` and all dependent libraries located in `` to the appropriate directory in your RGLUnityPlugin. +3. Copy all ROS2 libraries from `/lib/ros2_standalone/` to `Ros2ForUnity\Plugins\Windows\x86_64` directory. Skip for duplicates. -RGL's ROS2 standalone build is dependent on ROS2ForUnity's ROS2 standalone build. RobotecGPULidar library will find ROS2 because ROS2ForUnity sets the environment variable `Path` for the ROS2 libraries. +*Note:* On Windows there is no `-rpath` linker flag to modify an executable’s search path. Instead, the `Path` environment variable can be set. ROS2ForUnity handles this automatically before the simulation starts. ## Usage diff --git a/extensions/ros2/CMakeLists.txt b/extensions/ros2/CMakeLists.txt index dbd21228..d07663ef 100644 --- a/extensions/ros2/CMakeLists.txt +++ b/extensions/ros2/CMakeLists.txt @@ -1,6 +1,9 @@ -if (NOT $ENV{ROS_DISTRO} STREQUAL "humble") - message(FATAL_ERROR "ROS $ENV{ROS_DISTRO} not supported. Only humble is available.") +set(SUPPORTED_ROS_DISTROS humble jazzy) + +list(FIND SUPPORTED_ROS_DISTROS $ENV{ROS_DISTRO} ROS_DISTRO_INDEX) +if (ROS_DISTRO_INDEX EQUAL -1) + message(FATAL_ERROR "ROS $ENV{ROS_DISTRO} not supported. Choose one of: '${SUPPORTED_ROS_DISTROS}'.") endif() find_package(rclcpp REQUIRED) diff --git a/extensions/ros2/install_deps.py b/extensions/ros2/install_deps.py index a4773790..ba415830 100755 --- a/extensions/ros2/install_deps.py +++ b/extensions/ros2/install_deps.py @@ -7,7 +7,7 @@ class Config: - SUPPORTED_ROS_DISTROS = ["humble"] + SUPPORTED_ROS_DISTROS = ["humble", "jazzy"] # Paths relative to project root RADAR_MSGS_DIR = os.path.join("external", "radar_msgs") @@ -39,7 +39,6 @@ def install_ros2_deps(cfg): run_subprocess_command("sudo apt-get install -y python3-colcon-common-extensions") # Clone radar msgs if not os.path.isdir(cfg.RADAR_MSGS_DIR): - run_subprocess_command("ls") run_subprocess_command( f"git clone --single-branch --depth 1 https://github.com/ros-perception/radar_msgs.git {cfg.RADAR_MSGS_DIR}") run_subprocess_command(f"cd {cfg.RADAR_MSGS_DIR} && git checkout {cfg.RADAR_MSGS_COMMIT} && cd ..") diff --git a/ros2_standalone/CMakeLists.txt b/ros2_standalone/CMakeLists.txt index d823c89c..d3b8b1ed 100644 --- a/ros2_standalone/CMakeLists.txt +++ b/ros2_standalone/CMakeLists.txt @@ -17,6 +17,7 @@ set(REQ_THIRD_PARTY_STANDALONE_LIBS "") set(REQ_STANDALONE_DLLS "") set(INSTALL_DESTINATION_DIR "lib/ros2_standalone") +set(ROS_DISTRO "$ENV{ROS_DISTRO}") # Extend REQ_THIRD_PARTY_STANDALONE_LIBS with _library_name third party dependencies macro(get_standalone_third_party_dependencies _library_name) @@ -104,6 +105,12 @@ macro(get_standalone_dependencies _library_name) endif() endif() + # Get rosidl_dynamic_typesupport_fastrtps + if("${_library_name}" STREQUAL "rosidl_dynamic_typesupport_fastrtps") + fetch_target_lib(rosidl_dynamic_typesupport_fastrtps::rosidl_dynamic_typesupport_fastrtps) + list(APPEND REQ_STANDALONE_LIBS ${rosidl_dynamic_typesupport_fastrtps_rosidl_dynamic_typesupport_fastrtps_LIB_PATH}) + endif() + # If library is msgs, fetch all targets to get libraries for dynamic type support # Those libraries are not listed in _LIBRARIES (which stands for libraries to link against to use ) if(${_library_name} MATCHES ".*\_msgs$") @@ -116,12 +123,20 @@ macro(get_standalone_dependencies _library_name) endif() # Get spdlog and dependency - if(UNIX AND "${_library_name}" STREQUAL "spdlog") + if("${_library_name}" STREQUAL "spdlog") include(${${_library_name}_CONFIG}) fetch_target_lib(spdlog::spdlog) list(APPEND REQ_STANDALONE_LIBS ${spdlog_spdlog_LIB_PATH}) - fetch_target_lib(fmt::fmt) - list(APPEND REQ_STANDALONE_LIBS ${fmt_fmt_LIB_PATH}) + if(UNIX) + fetch_target_lib(fmt::fmt) + list(APPEND REQ_STANDALONE_LIBS ${fmt_fmt_LIB_PATH}) + endif() + endif() + + # Get yaml + if(WIN32 AND "${_library_name}" STREQUAL "yaml") + fetch_target_lib(yaml) + list(APPEND REQ_STANDALONE_LIBS ${yaml_LIB_PATH}) endif() # We skip python libs @@ -223,14 +238,23 @@ set(ros2_standalone_libs rosidl_typesupport_cpp rosidl_typesupport_introspection_c rosidl_typesupport_introspection_cpp - spdlog tracetools - yaml tinyxml2 ) if(UNIX) - list(APPEND ros2_standalone_libs iceoryx_binding_c iceoryx_posh iceoryx_hoofs) + list(APPEND ros2_standalone_libs spdlog iceoryx_binding_c iceoryx_posh iceoryx_hoofs) +endif() + +if(ROS_DISTRO STREQUAL "humble") + list(APPEND ros2_standalone_libs yaml) +endif() + +if(ROS_DISTRO STREQUAL "jazzy") + list(APPEND ros2_standalone_libs rcl_logging_spdlog rcl_interfaces rosidl_dynamic_typesupport_fastrtps) + if(WIN32) + list(APPEND ros2_standalone_libs yaml spdlog) + endif() endif() foreach(ros2_standalone_lib ${ros2_standalone_libs}) diff --git a/src/gpu/nodeKernels.cu b/src/gpu/nodeKernels.cu index 984ef492..3917397c 100644 --- a/src/gpu/nodeKernels.cu +++ b/src/gpu/nodeKernels.cu @@ -12,16 +12,16 @@ // See the License for the specific language governing permissions and // limitations under the License. +#include +#include +#include + #include #include #include #include #include -#include -#include -#include - __global__ void kFormatSoaToAos(size_t pointCount, size_t pointSize, size_t fieldCount, const GPUFieldDesc* soaInData, char* aosOutData) { @@ -209,7 +209,8 @@ __device__ void saveReturnAsNonHit(const RaytraceRequestContext* ctx, int firstS __global__ void kRadarComputeEnergy(size_t count, float rayAzimuthStepRad, float rayElevationStepRad, float freq, Mat3x4f lookAtOriginTransform, const Field::type* rayPose, const Field::type* hitDist, const Field::type* hitNorm, - const Field::type* hitPos, Vector<3, thrust::complex>* outBUBRFactor) + const Field::type* hitPos, + Vector<3, cuda::std::complex>* outBUBRFactor) { LIMIT(count); @@ -219,7 +220,7 @@ __global__ void kRadarComputeEnergy(size_t count, float rayAzimuthStepRad, float constexpr float reflectionCoef = 1.0f; // TODO const float waveLen = c0 / freq; const float waveNum = 2.0f * static_cast(M_PI) / waveLen; - const thrust::complex i = {0, 1.0}; + const cuda::std::complex i = {0, 1.0}; const Vec3f dirX = {1, 0, 0}; const Vec3f dirY = {0, 1, 0}; const Vec3f dirZ = {0, 0, 1}; @@ -255,15 +256,15 @@ __global__ void kRadarComputeEnergy(size_t count, float rayAzimuthStepRad, float const Vec3f reflectedDir = (rayDir - hitNormalLocal * (2 * rayDir.dot(hitNormalLocal))).normalized(); const Vec3f reflectedPol = reflectPolarization(rayPol, hitNormalLocal, rayDir); - const Vector<3, thrust::complex> reflectedPolCplx = {reflectedPol.x(), reflectedPol.y(), reflectedPol.z()}; + const Vector<3, cuda::std::complex> reflectedPolCplx = {reflectedPol.x(), reflectedPol.y(), reflectedPol.z()}; const float kr = waveNum * hitDistance; if (log) printf("reflectedDir: (%.4f %.4f %.4f) reflectedPol: (%.4f %.4f %.4f)\n", reflectedDir.x(), reflectedDir.y(), reflectedDir.z(), reflectedPol.x(), reflectedPol.y(), reflectedPol.z()); - const Vector<3, thrust::complex> apE = reflectionCoef * exp(i * kr) * reflectedPolCplx; - const Vector<3, thrust::complex> apH = -apE.cross(reflectedDir); + const Vector<3, cuda::std::complex> apE = reflectionCoef * exp(i * kr) * reflectedPolCplx; + const Vector<3, cuda::std::complex> apH = -apE.cross(reflectedDir); if (log) printf("apE: [(%.2f + %.2fi) (%.2f + %.2fi) (%.2f + %.2fi)]\n", apE.x().real(), apE.x().imag(), apE.y().real(), @@ -272,9 +273,9 @@ __global__ void kRadarComputeEnergy(size_t count, float rayAzimuthStepRad, float printf("apH: [(%.2f + %.2fi) (%.2f + %.2fi) (%.2f + %.2fi)]\n", apH.x().real(), apH.x().imag(), apH.y().real(), apH.y().imag(), apH.z().real(), apH.z().imag()); - const Vector<3, thrust::complex> BU1 = apE.cross(-dirP); - const Vector<3, thrust::complex> BU2 = apH.cross(dirT); - const Vector<3, thrust::complex> refField1 = (-(BU1 + BU2)); + const Vector<3, cuda::std::complex> BU1 = apE.cross(-dirP); + const Vector<3, cuda::std::complex> BU2 = apH.cross(dirT); + const Vector<3, cuda::std::complex> refField1 = (-(BU1 + BU2)); if (log) printf("BU1: [(%.2f + %.2fi) (%.2f + %.2fi) (%.2f + %.2fi)]\n" @@ -284,9 +285,9 @@ __global__ void kRadarComputeEnergy(size_t count, float rayAzimuthStepRad, float BU2.x().imag(), BU2.y().real(), BU2.y().imag(), BU2.z().real(), BU2.z().imag(), refField1.x().real(), refField1.x().imag(), refField1.y().real(), refField1.y().imag(), refField1.z().real(), refField1.z().imag()); - const Vector<3, thrust::complex> BR1 = apE.cross(dirT); - const Vector<3, thrust::complex> BR2 = apH.cross(dirP); - const Vector<3, thrust::complex> refField2 = (-(BR1 + BR2)); + const Vector<3, cuda::std::complex> BR1 = apE.cross(dirT); + const Vector<3, cuda::std::complex> BR2 = apH.cross(dirP); + const Vector<3, cuda::std::complex> refField2 = (-(BR1 + BR2)); if (log) printf("BR1: [(%.2f + %.2fi) (%.2f + %.2fi) (%.2f + %.2fi)]\n" @@ -296,14 +297,15 @@ __global__ void kRadarComputeEnergy(size_t count, float rayAzimuthStepRad, float BR2.x().imag(), BR2.y().real(), BR2.y().imag(), BR2.z().real(), BR2.z().imag(), refField2.x().real(), refField2.x().imag(), refField2.y().real(), refField2.y().imag(), refField2.z().real(), refField2.z().imag()); - const thrust::complex BU = refField1.dot(reflectedDir); - const thrust::complex BR = refField2.dot(reflectedDir); - // const thrust::complex factor = thrust::complex(0.0, ((waveNum * rayArea) / (4.0f * static_cast(M_PI)))) * + const cuda::std::complex BU = refField1.dot(reflectedDir); + const cuda::std::complex BR = refField2.dot(reflectedDir); + // const cuda::std::complex factor = cuda::std::complex(0.0, ((waveNum * rayArea) / (4.0f * static_cast(M_PI)))) * // exp(-i * waveNum * hitDistance); - const thrust::complex factor = thrust::complex(0.0, ((waveNum * rayArea * reflectedDir.dot(hitNormalLocal)) / - (4.0f * static_cast(M_PI)))) * - exp(-i * waveNum * hitDistance); - // const thrust::complex factor = thrust::complex(0.0, ((waveNum * rayArea) / (4.0f * static_cast(M_PI)))) * + const cuda::std::complex factor = cuda::std::complex(0.0, + ((waveNum * rayArea * reflectedDir.dot(hitNormalLocal)) / + (4.0f * static_cast(M_PI)))) * + exp(-i * waveNum * hitDistance); + // const cuda::std::complex factor = cuda::std::complex(0.0, ((waveNum * rayArea) / (4.0f * static_cast(M_PI)))) * // exp(-i * vecK.dot(hitPosLocal)); const auto BUf = BU * factor; @@ -485,7 +487,7 @@ void gpuFilterGroundPoints(cudaStream_t stream, size_t pointCount, const Vec3f s void gpuRadarComputeEnergy(cudaStream_t stream, size_t count, float rayAzimuthStepRad, float rayElevationStepRad, float freq, Mat3x4f lookAtOriginTransform, const Field::type* rayPose, const Field::type* hitDist, const Field::type* hitNorm, - const Field::type* hitPos, Vector<3, thrust::complex>* outBUBRFactor) + const Field::type* hitPos, Vector<3, std::complex>* outBUBRFactor) { run(kRadarComputeEnergy, stream, count, rayAzimuthStepRad, rayElevationStepRad, freq, lookAtOriginTransform, rayPose, hitDist, hitNorm, hitPos, outBUBRFactor); diff --git a/src/gpu/nodeKernels.hpp b/src/gpu/nodeKernels.hpp index cfb45d8b..24fe1014 100644 --- a/src/gpu/nodeKernels.hpp +++ b/src/gpu/nodeKernels.hpp @@ -14,6 +14,7 @@ #pragma once +#include #include #include @@ -21,7 +22,6 @@ #include #include #include -#include #include #include @@ -52,6 +52,6 @@ void gpuFilterGroundPoints(cudaStream_t stream, size_t pointCount, const Vec3f s void gpuRadarComputeEnergy(cudaStream_t stream, size_t count, float rayAzimuthStepRad, float rayElevationStepRad, float freq, Mat3x4f lookAtOriginTransform, const Field::type* rayPose, const Field::type* hitDist, const Field::type* hitNorm, - const Field::type* hitPos, Vector<3, thrust::complex>* outBUBRFactor); + const Field::type* hitPos, Vector<3, std::complex>* outBUBRFactor); void gpuReduceDivergentBeams(cudaStream_t stream, size_t beamCount, int samplesPerBeam, rgl_return_mode_t returnMode, - const RaytraceRequestContext* ctx); \ No newline at end of file + const RaytraceRequestContext* ctx); diff --git a/src/graph/NodesCore.hpp b/src/graph/NodesCore.hpp index 09b3c495..fca18c9e 100644 --- a/src/graph/NodesCore.hpp +++ b/src/graph/NodesCore.hpp @@ -14,6 +14,7 @@ #pragma once +#include #include #include #include @@ -630,10 +631,10 @@ struct RadarPostprocessPointsNode : IPointsNodeSingleInput HostPinnedArray::type>::Ptr radialSpeedInputHost = HostPinnedArray::type>::create(); HostPinnedArray::type>::Ptr elevationInputHost = HostPinnedArray::type>::create(); - DeviceAsyncArray>>::Ptr outBUBRFactorDev = - DeviceAsyncArray>>::create(arrayMgr); - HostPinnedArray>>::Ptr outBUBRFactorHost = - HostPinnedArray>>::create(); + DeviceAsyncArray>>::Ptr outBUBRFactorDev = + DeviceAsyncArray>>::create(arrayMgr); + HostPinnedArray>>::Ptr outBUBRFactorHost = + HostPinnedArray>>::create(); HostPageableArray::type>::Ptr clusterRcsHost = HostPageableArray::type>::create(); HostPageableArray::type>::Ptr clusterPowerHost = HostPageableArray::type>::create(); diff --git a/src/graph/RadarPostprocessPointsNode.cpp b/src/graph/RadarPostprocessPointsNode.cpp index 252f1c49..3c0830af 100644 --- a/src/graph/RadarPostprocessPointsNode.cpp +++ b/src/graph/RadarPostprocessPointsNode.cpp @@ -14,6 +14,7 @@ #include #include +#include #include #include @@ -364,8 +365,10 @@ void RadarPostprocessPointsNode::RadarCluster::takeIndicesFrom(RadarCluster&& ot minMaxDistance[1] = std::max(minMaxDistance[1], other.minMaxDistance[1]); minMaxAzimuth[0] = std::min(minMaxAzimuth[0], other.minMaxAzimuth[0]); minMaxAzimuth[1] = std::max(minMaxAzimuth[1], other.minMaxAzimuth[1]); - minMaxRadialSpeed[0] = std::isnan(other.minMaxRadialSpeed[0]) ? minMaxRadialSpeed[0] : std::min(other.minMaxRadialSpeed[0], minMaxRadialSpeed[0]); - minMaxRadialSpeed[1] = std::isnan(other.minMaxRadialSpeed[1]) ? minMaxRadialSpeed[1] : std::min(other.minMaxRadialSpeed[1], minMaxRadialSpeed[1]); + minMaxRadialSpeed[0] = std::isnan(other.minMaxRadialSpeed[0]) ? minMaxRadialSpeed[0] : + std::min(other.minMaxRadialSpeed[0], minMaxRadialSpeed[0]); + minMaxRadialSpeed[1] = std::isnan(other.minMaxRadialSpeed[1]) ? minMaxRadialSpeed[1] : + std::min(other.minMaxRadialSpeed[1], minMaxRadialSpeed[1]); minMaxElevation[0] = std::min(minMaxElevation[0], other.minMaxElevation[0]); minMaxElevation[1] = std::max(minMaxElevation[1], other.minMaxElevation[1]);