Skip to content

RPP Tensor Support - Fisheye on HOST and HIP#346

Closed
sampath1117 wants to merge 25 commits intor-abishek:developfrom
sampath1117:sr/opt_fisheye
Closed

RPP Tensor Support - Fisheye on HOST and HIP#346
sampath1117 wants to merge 25 commits intor-abishek:developfrom
sampath1117:sr/opt_fisheye

Conversation

@sampath1117
Copy link
Copy Markdown

  • Adds tensor support for Fog function optimized on AVX2 and HIP
  • Adds test suite support

@sampath1117 sampath1117 changed the title RPP Tensor Support - Fisheye on HOST and HIP WIP - RPP Tensor Support - Fisheye on HOST and HIP Sep 25, 2024
@sampath1117 sampath1117 changed the base branch from develop to master September 26, 2024 11:15
@sampath1117 sampath1117 changed the base branch from master to develop September 26, 2024 11:15
@sampath1117 sampath1117 changed the title WIP - RPP Tensor Support - Fisheye on HOST and HIP RPP Tensor Support - Fisheye on HOST and HIP Sep 27, 2024
CHANGELOG.md Outdated

Full documentation for RPP is available at [https://rocm.docs.amd.com/projects/rpp/en/latest](https://rocm.docs.amd.com/projects/rpp/en/latest)

## RPP 1.18.0 (unreleased)
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Add the change under 1.9.10

CMakeLists.txt Outdated

# RPP Version
set(VERSION "1.9.10")
set(VERSION "1.18.0")
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Set version here to 1.9.10

#endif
// NOTE: IMPORTANT: Match the version with CMakelists.txt version
#define RPP_VERSION_MAJOR 1
#define RPP_VERSION_MINOR 9
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Here also retain version to 1.9.10

const __m256 avx_p1op255 = _mm256_set1_ps(1.0f / 255.0f);
const __m256 avx_p1op3 = _mm256_set1_ps(1.0f / 3.0f);
const __m256 avx_p2op3 = _mm256_set1_ps(2.0f / 3.0f);
const __m256 avx_pMinus1 = _mm256_set1_ps(-1);
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Can we make this -1.0f?


inline void compute_fisheye_src_loc_avx(__m256 &pDstY, __m256 &pDstX, __m256 &pSrcY, __m256 &pSrcX, __m256 &pHeight, __m256 &pWidth)
{
__m256 pNormX, pNormY, pDist;
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Add the AVX helper functions overall inside #if AVX2

}

// fisheye without fused output-layout toggle (NHWC -> NHWC)
if ((srcDescPtr->c == 3) && (srcDescPtr->layout == RpptLayout::NHWC) && (dstDescPtr->layout == RpptLayout::NHWC))
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

this must be else if

{
__m256i pRow;
rpp_simd_load(rpp_generic_nn_load_i8pln1_avx, srcPtrTempChn, srcLocArray, invalidLoad, pRow);
rpp_storeu_si64((__m128i *)(dstPtrTempChn), _mm256_castsi256_si128(pRow));
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Can probably use reinterpret_cast<__m128i *>(dstPtrTempChn) here

int id_z = hipBlockIdx_z * hipBlockDim_z + hipThreadIdx_z;
int4 srcRoi_i4 = *(int4 *)&roiTensorPtrSrc[id_z];
int width = (srcRoi_i4.z - srcRoi_i4.x) + 1;
int height = (srcRoi_i4.w - srcRoi_i4.y) + 1;
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Fix indentation

Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

in all the functions

Copy link
Copy Markdown
Collaborator

@Srihari-mcw Srihari-mcw left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Please address review comments. Thanks

HazarathKumarM and others added 4 commits January 6, 2025 07:03
…ocs/sphinx (r-abishek#515)

Bumps [rocm-docs-core[api_reference]](https://github.com/ROCm/rocm-docs-core) from 1.14.1 to 1.15.0.
- [Release notes](https://github.com/ROCm/rocm-docs-core/releases)
- [Changelog](https://github.com/ROCm/rocm-docs-core/blob/develop/CHANGELOG.md)
- [Commits](ROCm/rocm-docs-core@v1.14.1...v1.15.0)

---
updated-dependencies:
- dependency-name: rocm-docs-core[api_reference]
  dependency-type: direct:production
  update-type: version-update:semver-minor
...

Signed-off-by: dependabot[bot] <support@github.com>
Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com>
@Srihari-mcw
Copy link
Copy Markdown
Collaborator

Moved to different PR - Closed

@Srihari-mcw Srihari-mcw closed this Dec 8, 2025
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.

3 participants