Skip to content

Commit b2d72a2

Browse files
kiritigowdadependabot[bot]spolifroni-amdarvindcheruSundarRajan28
authored
ROCm 7.1 cherry-pick - Fixes and updates required for 7.1 (#1564)
* Docs - Bump rocm-docs-core[api_reference] from 1.22.0 to 1.23.0 in /docs/sphinx (#1550) Bumps [rocm-docs-core[api_reference]](https://github.com/ROCm/rocm-docs-core) from 1.22.0 to 1.23.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.22.0...v1.23.0) --- updated-dependencies: - dependency-name: rocm-docs-core[api_reference] dependency-version: 1.23.0 dependency-type: direct:production update-type: version-update:semver-minor ... Signed-off-by: dependabot[bot] <[email protected]> Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com> * Docs - update os support language (#1551) * updated wording around supported linux environments * RedHat -> RHEL * SLES 15-SP5 isn't supported; removed * updated SLES to SP6/7 * HIP - HIP_CHECK for hipLaunchKernelGGL for gated launch (#1552) * HIP Launch - Check failures * HIP Launch - Error checks * Error report - Updates * HIP - Capture Launch Errors * HIP Check - cleanup * Docs - removing SP6 (#1553) * Remove Meta Package dependency (#1554) Co-authored-by: Kiriti Gowda <[email protected]> * vx_rpp - Update blur (#1557) * Update box filter kernel API * Add bordertype variable * Docs - Bump rocm-docs-core[api_reference] from 1.23.0 to 1.24.1 in /docs/sphinx (#1558) Bumps [rocm-docs-core[api_reference]](https://github.com/ROCm/rocm-docs-core) from 1.23.0 to 1.24.1. - [Release notes](https://github.com/ROCm/rocm-docs-core/releases) - [Changelog](https://github.com/ROCm/rocm-docs-core/blob/v1.24.1/CHANGELOG.md) - [Commits](ROCm/rocm-docs-core@v1.23.0...v1.24.1) --- updated-dependencies: - dependency-name: rocm-docs-core[api_reference] dependency-version: 1.24.1 dependency-type: direct:production update-type: version-update:semver-minor ... Signed-off-by: dependabot[bot] <[email protected]> Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com> * Docs - Bump rocm-docs-core[api_reference] from 1.24.1 to 1.25.0 in /docs/sphinx (#1559) Bumps [rocm-docs-core[api_reference]](https://github.com/ROCm/rocm-docs-core) from 1.24.1 to 1.25.0. - [Release notes](https://github.com/ROCm/rocm-docs-core/releases) - [Changelog](https://github.com/ROCm/rocm-docs-core/blob/v1.25.0/CHANGELOG.md) - [Commits](ROCm/rocm-docs-core@v1.24.1...v1.25.0) --- updated-dependencies: - dependency-name: rocm-docs-core[api_reference] dependency-version: 1.25.0 dependency-type: direct:production update-type: version-update:semver-minor ... Signed-off-by: dependabot[bot] <[email protected]> Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com> * Docs - Bump rocm-docs-core[api_reference] from 1.25.0 to 1.26.0 in /docs/sphinx (#1560) Bumps [rocm-docs-core[api_reference]](https://github.com/ROCm/rocm-docs-core) from 1.25.0 to 1.26.0. - [Release notes](https://github.com/ROCm/rocm-docs-core/releases) - [Changelog](https://github.com/ROCm/rocm-docs-core/blob/v1.26.0/CHANGELOG.md) - [Commits](ROCm/rocm-docs-core@v1.25.0...v1.26.0) --- updated-dependencies: - dependency-name: rocm-docs-core[api_reference] dependency-version: 1.26.0 dependency-type: direct:production update-type: version-update:semver-minor ... Signed-off-by: dependabot[bot] <[email protected]> Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com> * Readme Updates - HIP Package changes (#1562) * ROCm 7.1 - Changelog updates (#1563) --------- Signed-off-by: dependabot[bot] <[email protected]> Co-authored-by: dependabot[bot] <49699333+dependabot[bot]@users.noreply.github.com> Co-authored-by: spolifroni-amd <[email protected]> Co-authored-by: arvindcheru <[email protected]> Co-authored-by: Sundar Rajan Vaithiyanathan <[email protected]>
1 parent 4abea28 commit b2d72a2

File tree

16 files changed

+400
-198
lines changed

16 files changed

+400
-198
lines changed

‎CHANGELOG.md‎

Lines changed: 12 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -4,17 +4,26 @@
44

55
The full documentation for MIVisionX is available at [https://rocm.docs.amd.com/projects/MIVisionX/en/latest/doxygen/html/index.html](https://rocm.docs.amd.com/projects/MIVisionX/en/latest/doxygen/html/index.html)
66

7-
## (Unreleased) MIVisionX 3.4.0
7+
## MIVisionX 3.4.0 for ROCm 7.1.0
8+
9+
### Added
10+
* VX_RPP - Update blur
11+
* HIP - HIP_CHECK for hipLaunchKernelGGL for gated launch
812

913
### Changed
1014
* AMD Custom V1.1.0 - OpenMP updates
15+
* HALF - Fix half.hpp path updates
1116

12-
### Known issues
17+
### Resolved issues
18+
* AMD Custom - dependency linking errors resolved
19+
* VX_RPP - Fix memory leak
20+
* Packaging - Remove Meta Package dependency for HIP
1321

22+
### Known issues
1423
* Installation on CentOS/RedHat/SLES requires the manual installation of the `FFMPEG` & `OpenCV` dev packages.
1524

1625
### Upcoming changes
17-
26+
* VX_AMD_MEDIA - rocDecode support for hardware decode
1827

1928
## MIVisionX 3.3.0 for ROCm 7.0.0
2029

‎CMakeLists.txt‎

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -236,12 +236,12 @@ if(DEFINED ENV{CPACK_RPM_PACKAGE_RELEASE})
236236
endif()
237237

238238
# Set the dependent packages
239-
set(MIVISIONX_RUNTIME_PACKAGE_LIST "rocm-hip-runtime, openmp-extras-runtime, rpp, rocblas, miopen-hip, migraphx")
239+
set(MIVISIONX_RUNTIME_PACKAGE_LIST "hip-runtime-amd, openmp-extras-runtime, rpp, rocblas, miopen-hip, migraphx")
240240

241241
# Set the dev dependent packages
242-
set(MIVISIONX_DEBIAN_DEV_PACKAGE_LIST "half, rocm-hip-runtime-dev, openmp-extras-dev, rpp-dev, rocblas-dev, miopen-hip-dev, migraphx-dev, pkg-config, libavcodec-dev, libavformat-dev, libavutil-dev, libswscale-dev, libopencv-dev")
242+
set(MIVISIONX_DEBIAN_DEV_PACKAGE_LIST "half, hip-dev, openmp-extras-dev, rpp-dev, rocblas-dev, miopen-hip-dev, migraphx-dev, pkg-config, libavcodec-dev, libavformat-dev, libavutil-dev, libswscale-dev, libopencv-dev")
243243
# TBD - Some RPM packages need Fusion Packages - libavcodec-devel, libavformat-devel, libavutil-devel, libswscale-devel, libopencv
244-
set(MIVISIONX_RPM_DEV_PACKAGE_LIST "half, rocm-hip-runtime-devel, openmp-extras-devel, rpp-devel, rocblas-devel, miopen-hip-devel, migraphx-devel, pkg-config")
244+
set(MIVISIONX_RPM_DEV_PACKAGE_LIST "half, hip-devel, openmp-extras-devel, rpp-devel, rocblas-devel, miopen-hip-devel, migraphx-devel, pkg-config")
245245

246246
# Add OS specific dependencies
247247
if(EXISTS "/etc/os-release")

‎README.md‎

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -84,7 +84,7 @@ MIVisionX toolkit provides tools for accomplishing your tasks throughout the who
8484
#### Linux
8585
* Ubuntu - `22.04` / `24.04`
8686
* RedHat - `8` / `9`
87-
* SLES - `15-SP5`
87+
* SLES - `15-SP7`
8888

8989
#### Windows
9090
* Windows `10` / `11`
@@ -104,7 +104,7 @@ MIVisionX toolkit provides tools for accomplishing your tasks throughout the who
104104
```
105105
* HIP
106106
```shell
107-
sudo apt install rocm-hip-runtime-dev
107+
sudo apt install hip-dev
108108
```
109109
* OpenMP
110110
```shell
@@ -364,7 +364,7 @@ Review all notable [changes](CHANGELOG.md#changelog) with the latest release
364364
* Linux distribution
365365
+ Ubuntu - `22.04` / `24.04`
366366
+ RHEL - `8` / `9`
367-
+ SLES - `15-SP5`
367+
+ SLES - `15 SP7`
368368
* ROCm: `7.0.0`
369369
* RPP - `2.0.0`
370370
* miopen-hip - `3.4.0`

‎amd_openvx/openvx/hipvx/arithmetic_kernels.cpp‎

Lines changed: 74 additions & 37 deletions
Large diffs are not rendered by default.

‎amd_openvx/openvx/hipvx/color_kernels.cpp‎

Lines changed: 103 additions & 53 deletions
Large diffs are not rendered by default.

‎amd_openvx/openvx/hipvx/filter_kernels.cpp‎

Lines changed: 24 additions & 12 deletions
Original file line numberDiff line numberDiff line change
@@ -206,7 +206,8 @@ int HipExec_Box_U8_U8_3x3(hipStream_t stream, vx_uint32 dstWidth, vx_uint32 dstH
206206
hipLaunchKernelGGL(Hip_Box_U8_U8_3x3, dim3(ceil((float)globalThreads_x/localThreads_x), ceil((float)globalThreads_y/localThreads_y)),
207207
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
208208
(const uchar *)pHipSrcImage, srcImageStrideInBytes);
209-
209+
HIP_CHECK(hipGetLastError()); // Check for launch error
210+
210211
return VX_SUCCESS;
211212
}
212213

@@ -351,7 +352,8 @@ int HipExec_Dilate_U8_U8_3x3(hipStream_t stream, vx_uint32 dstWidth, vx_uint32 d
351352
hipLaunchKernelGGL(Hip_Dilate_U8_U8_3x3, dim3(ceil((float)globalThreads_x/localThreads_x), ceil((float)globalThreads_y/localThreads_y)),
352353
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
353354
(const uchar *)pHipSrcImage, srcImageStrideInBytes);
354-
355+
HIP_CHECK(hipGetLastError()); // Check for launch error
356+
355357
return VX_SUCCESS;
356358
}
357359

@@ -496,7 +498,8 @@ int HipExec_Erode_U8_U8_3x3(hipStream_t stream, vx_uint32 dstWidth, vx_uint32 ds
496498
hipLaunchKernelGGL(Hip_Erode_U8_U8_3x3, dim3(ceil((float)globalThreads_x/localThreads_x), ceil((float)globalThreads_y/localThreads_y)),
497499
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
498500
(const uchar *)pHipSrcImage, srcImageStrideInBytes);
499-
501+
HIP_CHECK(hipGetLastError()); // Check for launch error
502+
500503
return VX_SUCCESS;
501504
}
502505

@@ -759,7 +762,8 @@ int HipExec_Median_U8_U8_3x3(hipStream_t stream, vx_uint32 dstWidth, vx_uint32 d
759762
hipLaunchKernelGGL(Hip_Median_U8_U8_3x3, dim3(ceil((float)globalThreads_x/localThreads_x), ceil((float)globalThreads_y/localThreads_y)),
760763
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
761764
(const uchar *)pHipSrcImage, srcImageStrideInBytes);
762-
765+
HIP_CHECK(hipGetLastError()); // Check for launch error
766+
763767
return VX_SUCCESS;
764768
}
765769

@@ -946,7 +950,8 @@ int HipExec_Gaussian_U8_U8_3x3(hipStream_t stream, vx_uint32 dstWidth, vx_uint32
946950
hipLaunchKernelGGL(Hip_Gaussian_U8_U8_3x3, dim3(ceil((float)globalThreads_x/localThreads_x), ceil((float)globalThreads_y/localThreads_y)),
947951
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
948952
(const uchar *)pHipSrcImage, srcImageStrideInBytes);
949-
953+
HIP_CHECK(hipGetLastError()); // Check for launch error
954+
950955
return VX_SUCCESS;
951956
}
952957

@@ -3673,7 +3678,8 @@ int HipExec_Convolve_U8_U8(hipStream_t stream, vx_uint32 dstWidth, vx_uint32 dst
36733678
} else {
36743679
return VX_ERROR_NOT_IMPLEMENTED;
36753680
}
3676-
3681+
HIP_CHECK(hipGetLastError()); // Check for launch error
3682+
36773683
return VX_SUCCESS;
36783684
}
36793685

@@ -6412,7 +6418,8 @@ int HipExec_Convolve_S16_U8(hipStream_t stream, vx_uint32 dstWidth, vx_uint32 ds
64126418
} else {
64136419
return VX_ERROR_NOT_IMPLEMENTED;
64146420
}
6415-
6421+
HIP_CHECK(hipGetLastError()); // Check for launch error
6422+
64166423
return VX_SUCCESS;
64176424
}
64186425

@@ -6572,7 +6579,8 @@ int HipExec_Sobel_S16_U8_3x3_GX(hipStream_t stream, vx_uint32 dstWidth, vx_uint3
65726579
hipLaunchKernelGGL(Hip_Sobel_S16_U8_3x3_GX, dim3(ceil((float)globalThreads_x/localThreads_x), ceil((float)globalThreads_y/localThreads_y)),
65736580
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
65746581
(const uchar *)pHipSrcImage, srcImageStrideInBytes);
6575-
6582+
HIP_CHECK(hipGetLastError()); // Check for launch error
6583+
65766584
return VX_SUCCESS;
65776585
}
65786586

@@ -6716,7 +6724,8 @@ int HipExec_Sobel_S16_U8_3x3_GY(hipStream_t stream, vx_uint32 dstWidth, vx_uint3
67166724
hipLaunchKernelGGL(Hip_Sobel_S16_U8_3x3_GY, dim3(ceil((float)globalThreads_x/localThreads_x), ceil((float)globalThreads_y/localThreads_y)),
67176725
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
67186726
(const uchar *)pHipSrcImage, srcImageStrideInBytes);
6719-
6727+
HIP_CHECK(hipGetLastError()); // Check for launch error
6728+
67206729
return VX_SUCCESS;
67216730
}
67226731

@@ -6935,7 +6944,8 @@ int HipExec_Sobel_S16S16_U8_3x3_GXY(hipStream_t stream, vx_uint32 dstWidth, vx_u
69356944
hipLaunchKernelGGL(Hip_Sobel_S16S16_U8_3x3_GXY, dim3(ceil((float)globalThreads_x/localThreads_x), ceil((float)globalThreads_y/localThreads_y)),
69366945
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage1, dstImage1StrideInBytes, (uchar *)pHipDstImage2, dstImage2StrideInBytes,
69376946
(const uchar *)pHipSrcImage, srcImageStrideInBytes);
6938-
6947+
HIP_CHECK(hipGetLastError()); // Check for launch error
6948+
69396949
return VX_SUCCESS;
69406950
}
69416951

@@ -7049,7 +7059,8 @@ int HipExec_ScaleGaussianHalf_U8_U8_3x3(hipStream_t stream, vx_uint32 dstWidth,
70497059
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
70507060
srcWidth, srcHeight, (const uchar *)pHipSrcImage, srcImageStrideInBytes,
70517061
dstWidthComp);
7052-
7062+
HIP_CHECK(hipGetLastError()); // Check for launch error
7063+
70537064
return VX_SUCCESS;
70547065
}
70557066

@@ -7303,6 +7314,7 @@ int HipExec_ScaleGaussianHalf_U8_U8_5x5(hipStream_t stream, vx_uint32 dstWidth,
73037314
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
73047315
srcWidth, srcHeight, (const uchar *)pHipSrcImage, srcImageStrideInBytes, srcImageBufferSize,
73057316
dstWidthComp);
7306-
7317+
HIP_CHECK(hipGetLastError()); // Check for launch error
7318+
73077319
return VX_SUCCESS;
73087320
}

‎amd_openvx/openvx/hipvx/geometric_kernels.cpp‎

Lines changed: 34 additions & 17 deletions
Original file line numberDiff line numberDiff line change
@@ -88,7 +88,8 @@ int HipExec_ScaleImage_U8_U8_Nearest(hipStream_t stream, vx_uint32 dstWidth, vx_
8888
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
8989
(const uchar *)pHipSrcImage, srcImageStrideInBytes,
9090
xscale, yscale, xoffset, yoffset);
91-
91+
HIP_CHECK(hipGetLastError()); // Check for launch error
92+
9293
return VX_SUCCESS;
9394
}
9495

@@ -174,7 +175,8 @@ int HipExec_ScaleImage_U8_U8_Bilinear(hipStream_t stream, vx_uint32 dstWidth, vx
174175
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
175176
(const uchar *)pHipSrcImage, srcImageStrideInBytes,
176177
xscale, yscale, xoffset, yoffset);
177-
178+
HIP_CHECK(hipGetLastError()); // Check for launch error
179+
178180
return VX_SUCCESS;
179181
}
180182

@@ -315,7 +317,8 @@ int HipExec_ScaleImage_U8_U8_Bilinear_Replicate(hipStream_t stream, vx_uint32 ds
315317
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
316318
(const uchar *)pHipSrcImage, srcImageStrideInBytes, srcWidth, srcHeight,
317319
xscale, yscale, xoffset, yoffset);
318-
320+
HIP_CHECK(hipGetLastError()); // Check for launch error
321+
319322
return VX_SUCCESS;
320323
}
321324

@@ -444,7 +447,8 @@ int HipExec_ScaleImage_U8_U8_Bilinear_Constant(hipStream_t stream, vx_uint32 dst
444447
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
445448
(const uchar *)pHipSrcImage, srcImageStrideInBytes, srcWidth, srcHeight,
446449
xscale, yscale, xoffset, yoffset, borderValue);
447-
450+
HIP_CHECK(hipGetLastError()); // Check for launch error
451+
448452
return VX_SUCCESS;
449453
}
450454

@@ -755,7 +759,8 @@ int HipExec_ScaleImage_U8_U8_Area(hipStream_t stream, vx_uint32 dstWidth, vx_uin
755759
(const uchar *)pHipSrcImage, srcImageStrideInBytes,
756760
Nx, Ny, iSxSy);
757761
}
758-
762+
HIP_CHECK(hipGetLastError()); // Check for launch error
763+
759764
return VX_SUCCESS;
760765
}
761766

@@ -843,7 +848,8 @@ int HipExec_WarpAffine_U8_U8_Nearest(hipStream_t stream, vx_uint32 dstWidth, vx_
843848
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
844849
(const uchar *)pHipSrcImage, srcImageStrideInBytes, srcImageBufferSize,
845850
(d_affine_matrix_t *) affineMatrix);
846-
851+
HIP_CHECK(hipGetLastError()); // Check for launch error
852+
847853
return VX_SUCCESS;
848854
}
849855

@@ -991,7 +997,8 @@ int HipExec_WarpAffine_U8_U8_Nearest_Constant(hipStream_t stream, vx_uint32 dstW
991997
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
992998
(const uchar *)pHipSrcImage, srcImageStrideInBytes,
993999
(d_affine_matrix_t *) affineMatrix, (uint) borderValue, rect_valid);
994-
1000+
HIP_CHECK(hipGetLastError()); // Check for launch error
1001+
9951002
return VX_SUCCESS;
9961003
}
9971004

@@ -1063,7 +1070,8 @@ int HipExec_WarpAffine_U8_U8_Bilinear(hipStream_t stream, vx_uint32 dstWidth, vx
10631070
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
10641071
(const uchar *)pHipSrcImage, srcImageStrideInBytes, srcImageBufferSize,
10651072
(d_affine_matrix_t *) affineMatrix);
1066-
1073+
HIP_CHECK(hipGetLastError()); // Check for launch error
1074+
10671075
return VX_SUCCESS;
10681076
}
10691077

@@ -1135,7 +1143,8 @@ int HipExec_WarpAffine_U8_U8_Bilinear_Constant(hipStream_t stream, vx_uint32 dst
11351143
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
11361144
(const uchar *)pHipSrcImage, srcImageStrideInBytes,
11371145
(d_affine_matrix_t *) affineMatrix, (uint) borderValue);
1138-
1146+
HIP_CHECK(hipGetLastError()); // Check for launch error
1147+
11391148
return VX_SUCCESS;
11401149
}
11411150

@@ -1240,7 +1249,8 @@ int HipExec_WarpPerspective_U8_U8_Nearest(hipStream_t stream, vx_uint32 dstWidth
12401249
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
12411250
(const uchar *)pHipSrcImage, srcImageStrideInBytes, srcImageBufferSize,
12421251
(d_perspective_matrix_t *) perspectiveMatrix);
1243-
1252+
HIP_CHECK(hipGetLastError()); // Check for launch error
1253+
12441254
return VX_SUCCESS;
12451255
}
12461256

@@ -1401,7 +1411,8 @@ int HipExec_WarpPerspective_U8_U8_Nearest_Constant(hipStream_t stream, vx_uint32
14011411
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
14021412
srcWidth, srcHeight, (const uchar *)pHipSrcImage, srcImageStrideInBytes,
14031413
(d_perspective_matrix_t *) perspectiveMatrix, (uint) borderValue);
1404-
1414+
HIP_CHECK(hipGetLastError()); // Check for launch error
1415+
14051416
return VX_SUCCESS;
14061417
}
14071418

@@ -1491,7 +1502,8 @@ int HipExec_WarpPerspective_U8_U8_Bilinear(hipStream_t stream, vx_uint32 dstWidt
14911502
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
14921503
(const uchar *)pHipSrcImage, srcImageStrideInBytes, srcImageBufferSize,
14931504
(d_perspective_matrix_t *) perspectiveMatrix);
1494-
1505+
HIP_CHECK(hipGetLastError()); // Check for launch error
1506+
14951507
return VX_SUCCESS;
14961508
}
14971509

@@ -1581,7 +1593,8 @@ int HipExec_WarpPerspective_U8_U8_Bilinear_Constant(hipStream_t stream, vx_uint3
15811593
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage , dstImageStrideInBytes,
15821594
srcWidth, srcHeight, (const uchar *)pHipSrcImage, srcImageStrideInBytes,
15831595
(d_perspective_matrix_t *) perspectiveMatrixLoc, (uint) borderValue);
1584-
1596+
HIP_CHECK(hipGetLastError()); // Check for launch error
1597+
15851598
return VX_SUCCESS;
15861599
}
15871600

@@ -1688,7 +1701,8 @@ int HipExec_Remap_U8_U8_Nearest(hipStream_t stream, vx_uint32 dstWidth, vx_uint3
16881701
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
16891702
(const uchar *)pHipSrcImage, srcImageStrideInBytes, srcImageBufferSize,
16901703
(uchar *) remap, remapStrideInBytes);
1691-
1704+
HIP_CHECK(hipGetLastError()); // Check for launch error
1705+
16921706
return VX_SUCCESS;
16931707
}
16941708

@@ -1833,7 +1847,8 @@ int HipExec_Remap_U8_U8_Nearest_Constant(hipStream_t stream, vx_uint32 dstWidth,
18331847
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
18341848
srcWidth, srcHeight, (const uchar *)pHipSrcImage, srcImageStrideInBytes, srcImageBufferSize,
18351849
(uchar *) remap, remapStrideInBytes, (uint) borderValue);
1836-
1850+
HIP_CHECK(hipGetLastError()); // Check for launch error
1851+
18371852
return VX_SUCCESS;
18381853
}
18391854

@@ -1893,7 +1908,8 @@ int HipExec_Remap_U8_U8_Bilinear(hipStream_t stream, vx_uint32 dstWidth, vx_uint
18931908
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
18941909
(const uchar *)pHipSrcImage, srcImageStrideInBytes, srcImageBufferSize,
18951910
(uchar *) remap, remapStrideInBytes);
1896-
1911+
HIP_CHECK(hipGetLastError()); // Check for launch error
1912+
18971913
return VX_SUCCESS;
18981914
}
18991915

@@ -1953,6 +1969,7 @@ int HipExec_Remap_U8_U8_Bilinear_Constant(hipStream_t stream, vx_uint32 dstWidth
19531969
dim3(localThreads_x, localThreads_y), 0, stream, dstWidth, dstHeight, (uchar *)pHipDstImage, dstImageStrideInBytes,
19541970
srcWidth, srcHeight, (const uchar *)pHipSrcImage, srcImageStrideInBytes,
19551971
(uchar *) remap, remapStrideInBytes, (uint) borderValue);
1956-
1972+
HIP_CHECK(hipGetLastError()); // Check for launch error
1973+
19571974
return VX_SUCCESS;
19581975
}

‎amd_openvx/openvx/hipvx/hip_common_funcs.h‎

Lines changed: 9 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -24,8 +24,17 @@ THE SOFTWARE.
2424
#ifndef MIVISIONX_HIP_COMMON_FUNCS_H
2525
#define MIVISIONX_HIP_COMMON_FUNCS_H
2626

27+
#include <iostream>
2728
#include "hip/hip_runtime.h"
2829

30+
#define HIP_CHECK(command) { \
31+
hipError_t status = command; \
32+
if (status != hipSuccess) { \
33+
std::cerr << "AMD OpenVX: HIP Error Reported -- " << hipGetErrorString(status) << std::endl; \
34+
return status; \
35+
} \
36+
}
37+
2938
#define MASK_EARLY_EXIT 4369
3039
#define HIPSELECT(a, b, c) (c ? b : a)
3140

0 commit comments

Comments
 (0)