Erode - HOST and HIP update by r-abishek · Pull Request #666 · ROCm/rpp

r-abishek · 2026-01-27T04:53:38Z

Adds tensor style implementation for Erode HOST and updates to HIP implementation.
Adds relevant QA unit and performance tests.

removed commented code

…for single line condition

Erode - rebased with latest changes

Copilot

Pull request overview

This PR adds HOST backend support and improves HIP implementation for the Erode morphological operation in the tensor API. The changes enable erode to work with multiple data types (U8, I8, F16, F32) and layouts (NCHW, NHWC) on the HOST backend.

Changes:

Added HOST backend support for erode operation in the tensor API
Implemented SIMD-optimized helpers for erode minimum operations across different kernel sizes (3x3, 5x5, 7x7, 9x9)
Enhanced test coverage by integrating erode into the HOST test suite

Reviewed changes

Copilot reviewed 7 out of 17 changed files in this pull request and generated 1 comment.

Show a summary per file

File	Description
utilities/test_suite/common.py	Added HOST backend to erode's supported platforms
utilities/test_suite/HOST/runImageTests.py	Included erode in kernel size iteration for unit and performance tests
utilities/test_suite/HOST/Tensor_image_host.cpp	Added erode test case handler with HOST API invocation
src/modules/tensor/rppt_tensor_morphological_operations.cpp	Implemented HOST backend function for erode with datatype support and validation
src/include/tensor/host_tensor_executors.hpp	Declared template functions for erode char and float implementations
src/include/common/cpu/rpp_cpu_filter.hpp	Added SIMD helper functions and template utilities for morphological operations
api/rppt_tensor_morphological_operations.h	Added HOST API documentation and fixed parameter description in HIP API docs

💡 Add Copilot custom instructions for smarter, more guided reviews. Learn how to get started.

Copilot · 2026-01-27T04:54:08Z

src/include/common/cpu/rpp_cpu_filter.hpp

+    static inline __m256  pad_float() { return avx_p1; }
+};
+
+// MorphPad_Dilate is defined for future morphological dilate implementation.


Comment references future dilate implementation but the struct is defined in this PR. Update comment to clarify that the struct is included for extensibility rather than being incomplete.

Suggested change

// MorphPad_Dilate is defined for future morphological dilate implementation.

// MorphPad_Dilate defines the padding policy for morphological dilation and

// is included here for API completeness and extensibility, even if unused by

// current kernels.

codecov · 2026-01-29T09:00:23Z

Codecov Report

❌ Patch coverage is 96.66667% with 5 lines in your changes missing coverage. Please review.

Files with missing lines	Patch %	Lines
...es/tensor/rppt_tensor_morphological_operations.cpp	95.16%	3 Missing ⚠️
src/include/common/cpu/rpp_cpu_filter.hpp	97.73%	2 Missing ⚠️

Additional details and impacted files

@@             Coverage Diff             @@
##           develop     #666      +/-   ##
===========================================
- Coverage    92.65%   92.59%   -0.07%     
===========================================
  Files          194      195       +1     
  Lines        82918    85265    +2347     
===========================================
+ Hits         76827    78943    +2116     
- Misses        6091     6322     +231

Files with missing lines	Coverage Δ
src/modules/tensor/cpu/kernel/erode.cpp	`92.99% <ø> (ø)`
src/modules/tensor/hip/kernel/erode.cpp	`99.64% <ø> (ø)`
src/include/common/cpu/rpp_cpu_filter.hpp	`99.57% <97.73%> (-0.43%)`	⬇️
...es/tensor/rppt_tensor_morphological_operations.cpp	`95.73% <95.16%> (-0.35%)`	⬇️

... and 3 files with indirect coverage changes

🚀 New features to boost your workflow:

❄️ Test Analytics: Detect flaky tests, report on failures, and find test suite problems.

src/modules/tensor/rppt_tensor_morphological_operations.cpp

* added initial api support for erode * added support for U8 and I8 bitdepths for 3, 5, 7, 9 kernel sizes * added F16 and F32 bitdepth support * added generic kernel support * added golden outputs removed commented code * fix build errors * Fix build and test_suite errors * revert padding changes * updated erode HIP kernel with latest changes * Add F32 QA * minor formatting fixes * minor comment fix * resolve copilot comments * resolve review comments * resolved review comments * Add unpack templating changes and fix segmentation issue * Fix PKD to PKD kernel 9 for Pack and Unpack changes. * Add and template signext function * Fix min Comments * Fix one min Comments * Add unroll and rename of preLoadRows * Fix remane of Loader and MorphVecLoader * Add empty line before comment * Fix remove empty line, rename of kernelSze & padPolicy and remove {} for single line condition * resolved review comments * fix build warnings --------- Co-authored-by: sampath1117 <sampath.rachumallu@multicorewareinc.com> Co-authored-by: HazarathKumarM <hazarathkumar@multicorewareinc.com> Co-authored-by: Mukesh Jayakodi <mukesh.jayakodi@multicorewareinc.com> Co-authored-by: Kiriti Gowda <kiritigowda@gmail.com> Co-authored-by: HazarathKumarM <119284987+HazarathKumarM@users.noreply.github.com> Co-authored-by: Lakshmi Kumar <lakshmi.kumar@amd.com>

* add support for dilate in HOST backend * minor fix in changelog * added golden outputs remove commented code * resolve build errors * Add padding changes in HIP backend * fix sigsev issues * fix QA for 9x9 kernel * Add if condition for pack function and template for unpack and signext function * Fix the rename of preLoadRows and max Comments * Fix Fix remane of Loader and MorphVecLoader * Fix empty space, dilate_row_hip_compute function, removed if & else and aligned indent R. * Fix remove whitespace and restored all unnecessary changes. * Fix remove precision line and reverted back to static cast. * Fix remove empty line, rename of kernelSze & padPolicy and remove {} for single line condition * Fix Indentation of IF condition. * resolved review comments * resolve review comments * Test suite - Add QA pass/fail tests for F32 bit depth (#665) * Travis CI - key error fix * Fix Bug in ColorTwist (#6) (#8) (#9) * Added golden outputs and resolved HOST backend * Updated bin files for median filter and resize crop mirror * Updated bin files * Updated bin files for the next set of kernel F32 QA * Updated bin files for jpeg_compression_distortion * Fixed resize QA failures * Fix for Resize bilinear F32 QA HOST and HIP * Fix for lens correction QA f32 for HOST and HIP for 1e-4 precision * Fixed HIP rcm QA * updates for warp Affine F32 QA * Fix for RCM QA match for U8 and F32 updates AVX * Fix for lens correction AVX * Removed space * Fixed warp affine for every other varient with the updated changes * Add fixes to match precision in quantization * Fix Precision mismatches * Update default cutoff to 1e-5 and specialized cutoff to 1e-4 * F32 QA Fix * Made Quality percentage as arg from testsuite * Resolved copilot comments * Resolved the copilot comments * Resolved Codex comments * HOST and HIP - pinned buffers for respective API (#628) * Removed memcpy and used hipHostMalloc for allocation : blend * Removed memcpy and used hipHostMalloc for allocation : brightness * Removed memcpy and used hipHostMalloc for allocation : color cast * Removed memcpy and used hipHostMalloc for allocation : color twist * Removed memcpy and used hipHostMalloc for allocation : contrast * Removed memcpy and used hipHostMalloc for allocation : crop mirror normalize * Removed memcpy and used hipHostMalloc for allocation : Exposure * Removed memcpy and used hipHostMalloc for allocation : Gamma correction * Removed memcpy and used hipHostMalloc for allocation : gaussian filter * Removed memcpy and used hipHostMalloc for allocation : Noise * Removed memcpy and used hipHostMalloc for allocation : Non linear blend * Removed memcpy and used hipHostMalloc for allocation : Resize mirror normalize * Removed memcpy and used hipHostMalloc for allocation : Water * Added hipHostFree for all kernels in test suite * Added hipHostFree for all kernels in test suite * Removed memcpy and used hipHostMalloc for allocation : Flip, spatter, rcm, color temperature * Resolved copilot review comments * Updated version * Removed unused parameter * Updated version in cmakeList * removed the host to device mem copies for warp affine and rotate * Updated version * Removed comment * Updated Chnagelog file * Update patch version from 2.2.0 to 2.2.1 * Update CHANGELOG * Address copilot comments for HIP HOST consistent allocation * Documentation changes for updated memcpy changes * Update ricap outer API to use pinned memory and remove mem copy * Fix memory allocation and deallocation for permutationTensor * Update api/rppt_tensor_effects_augmentations.h Co-authored-by: Copilot <175728472+Copilot@users.noreply.github.com> * Fix spelling of noiseProbability and saltProbability * Fix deallocation --------- Co-authored-by: HazarathKumarM <hazarathkumar@multicorewareinc.com> Co-authored-by: Kiriti Gowda <kiritigowda@gmail.com> Co-authored-by: Srihari-mcw <srihari@multicorewareinc.com> Co-authored-by: hmaddise <HazarathKumar.Maddisetty@amd.com> Co-authored-by: Copilot <175728472+Copilot@users.noreply.github.com> * resolved review comments * minor comment change * Resolved copilot review comments * Update src/modules/tensor/cpu/kernel/resize.cpp Co-authored-by: Copilot <175728472+Copilot@users.noreply.github.com> * Update src/modules/tensor/cpu/kernel/resize.cpp Co-authored-by: Copilot <175728472+Copilot@users.noreply.github.com> * Update src/modules/tensor/hip/kernel/jpeg_compression_distortion.cpp Co-authored-by: Copilot <175728472+Copilot@users.noreply.github.com> * Updated test suite and resoled review comments * Updated HIP for F32 QA reduction function cases --------- Co-authored-by: Kiriti Gowda <kiriti.nageshgowda@amd.com> Co-authored-by: Lokesh Bonta <lokeswara@multicorewareinc.com> Co-authored-by: sampath117 <snehaa@multicorewareinc.com> Co-authored-by: HazarathKumarM <hazarathkumar@multicorewareinc.com> Co-authored-by: Srihari-mcw <srihari@multicorewareinc.com> Co-authored-by: ManasaDattaT <tammisetti.manasadatta@multicorewareinc.com> Co-authored-by: Kiriti Gowda <kiritigowda@gmail.com> Co-authored-by: hmaddise <HazarathKumar.Maddisetty@amd.com> Co-authored-by: Copilot <175728472+Copilot@users.noreply.github.com> Co-authored-by: Lakshmi Kumar <lakshmi.kumar@amd.com> * Erode - HOST and HIP update (#666) * added initial api support for erode * added support for U8 and I8 bitdepths for 3, 5, 7, 9 kernel sizes * added F16 and F32 bitdepth support * added generic kernel support * added golden outputs removed commented code * fix build errors * Fix build and test_suite errors * revert padding changes * updated erode HIP kernel with latest changes * Add F32 QA * minor formatting fixes * minor comment fix * resolve copilot comments * resolve review comments * resolved review comments * Add unpack templating changes and fix segmentation issue * Fix PKD to PKD kernel 9 for Pack and Unpack changes. * Add and template signext function * Fix min Comments * Fix one min Comments * Add unroll and rename of preLoadRows * Fix remane of Loader and MorphVecLoader * Add empty line before comment * Fix remove empty line, rename of kernelSze & padPolicy and remove {} for single line condition * resolved review comments * fix build warnings --------- Co-authored-by: sampath1117 <sampath.rachumallu@multicorewareinc.com> Co-authored-by: HazarathKumarM <hazarathkumar@multicorewareinc.com> Co-authored-by: Mukesh Jayakodi <mukesh.jayakodi@multicorewareinc.com> Co-authored-by: Kiriti Gowda <kiritigowda@gmail.com> Co-authored-by: HazarathKumarM <119284987+HazarathKumarM@users.noreply.github.com> Co-authored-by: Lakshmi Kumar <lakshmi.kumar@amd.com> * fix build errors --------- Co-authored-by: sampath1117 <sampath.rachumallu@multicorewareinc.com> Co-authored-by: HazarathKumarM <hazarathkumar@multicorewareinc.com> Co-authored-by: Mukesh Jayakodi <mukesh.jayakodi@multicorewareinc.com> Co-authored-by: Kiriti Gowda <kiritigowda@gmail.com> Co-authored-by: HazarathKumarM <119284987+HazarathKumarM@users.noreply.github.com> Co-authored-by: Kiriti Gowda <kiriti.nageshgowda@amd.com> Co-authored-by: Lokesh Bonta <lokeswara@multicorewareinc.com> Co-authored-by: sampath117 <snehaa@multicorewareinc.com> Co-authored-by: Srihari-mcw <srihari@multicorewareinc.com> Co-authored-by: ManasaDattaT <tammisetti.manasadatta@multicorewareinc.com> Co-authored-by: hmaddise <HazarathKumar.Maddisetty@amd.com> Co-authored-by: Copilot <175728472+Copilot@users.noreply.github.com> Co-authored-by: Lakshmi Kumar <lakshmi.kumar@amd.com>

sampath1117 and others added 27 commits September 5, 2024 11:14

added initial api support for erode

142b006

added support for U8 and I8 bitdepths for 3, 5, 7, 9 kernel sizes

8642d83

added F16 and F32 bitdepth support

874aad6

added generic kernel support

854e5db

added golden outputs

4c572c7

removed commented code

Merge remote-tracking branch 'develop' into sr/opt_erode_host

1e9f45a

fix build errors

f3b721b

Fix build and test_suite errors

333484f

revert padding changes

d831c5f

updated erode HIP kernel with latest changes

57152c4

Add F32 QA

ae95005

minor formatting fixes

ce7aa2f

minor comment fix

77f84cc

resolve copilot comments

9821642

resolve review comments

a36f0fe

resolved review comments

0105c2f

Add unpack templating changes and fix segmentation issue

99786e5

Fix PKD to PKD kernel 9 for Pack and Unpack changes.

1b6c4db

Add and template signext function

67fb3b1

Fix min Comments

1e99ef7

Fix one min Comments

d81db53

Add unroll and rename of preLoadRows

055bfd2

Fix remane of Loader and MorphVecLoader

1782665

Add empty line before comment

79653c0

Fix remove empty line, rename of kernelSze & padPolicy and remove {} …

e09f82c

…for single line condition

Merge pull request #549 from HazarathKumarM/erode_rebased

8c88da6

Erode - rebased with latest changes

Merge branch 'ROCm:develop' into ar/opt_erode

088ffa2

r-abishek requested a review from Copilot January 27, 2026 04:53

r-abishek added the enhancement New feature or request label Jan 27, 2026

Copilot AI reviewed Jan 27, 2026

View reviewed changes

SundarRajan98 added the ci:precheckin label Jan 29, 2026

kiritigowda self-assigned this Feb 3, 2026

kiritigowda requested a review from LakshmiKumar23 February 3, 2026 19:03

Merge branch 'develop' into ar/opt_erode

167a190

LakshmiKumar23 requested a review from AryanSalmanpour February 5, 2026 03:57

AryanSalmanpour requested changes Feb 10, 2026

View reviewed changes

src/modules/tensor/rppt_tensor_morphological_operations.cpp Show resolved Hide resolved

HazarathKumarM and others added 3 commits February 10, 2026 04:09

resolved review comments

c6d7135

Merge branch 'develop' into ar/opt_erode

2f3bcf6

fix build warnings

785d650

LakshmiKumar23 requested a review from AryanSalmanpour February 11, 2026 18:25

AryanSalmanpour approved these changes Feb 11, 2026

View reviewed changes

LakshmiKumar23 added 2 commits February 11, 2026 10:52

Merge branch 'develop' into ar/opt_erode

23263d3

Merge branch 'develop' into ar/opt_erode

731abb9

LakshmiKumar23 approved these changes Feb 11, 2026

View reviewed changes

LakshmiKumar23 merged commit 1101e7f into ROCm:develop Feb 11, 2026
1 check passed

SundarRajan98 mentioned this pull request Feb 13, 2026

Add Erode and Dilate Host support ROCm/MIVisionX#1617

Merged

1 task

Provide feedback

Saved searches

Use saved searches to filter your results more quickly

Erode - HOST and HIP update#666

Erode - HOST and HIP update#666
LakshmiKumar23 merged 33 commits intoROCm:developfrom
r-abishek:ar/opt_erode

r-abishek commented Jan 27, 2026

Uh oh!

Copilot AI left a comment

Uh oh!

Copilot AI Jan 27, 2026

Uh oh!

HazarathKumarM Feb 10, 2026

Uh oh!

codecov bot commented Jan 29, 2026 •

edited

Loading

Uh oh!

Uh oh!

Uh oh!

Reviewers

Assignees

Labels

Projects

Milestone

Development

Uh oh!

9 participants

-// MorphPad_Dilate is defined for future morphological dilate implementation.
+// MorphPad_Dilate defines the padding policy for morphological dilation and
+// is included here for API completeness and extensibility, even if unused by
+// current kernels.

Conversation

r-abishek commented Jan 27, 2026

Uh oh!

Copilot AI left a comment

Choose a reason for hiding this comment

Pull request overview

Reviewed changes

Uh oh!

Copilot AI Jan 27, 2026

Choose a reason for hiding this comment

Uh oh!

HazarathKumarM Feb 10, 2026

Choose a reason for hiding this comment

Uh oh!

codecov bot commented Jan 29, 2026 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Codecov Report

Uh oh!

Uh oh!

Uh oh!

Reviewers

Assignees

Labels

Projects

Milestone

Development

Uh oh!

9 participants

codecov bot commented Jan 29, 2026 •

edited

Loading