From 26c3b4f98564e4fbc8b64292cc28e41ce4de8076 Mon Sep 17 00:00:00 2001 From: kjayapra-amd <54370791+kjayapra-amd@users.noreply.github.com> Date: Wed, 16 Oct 2019 01:22:15 -0400 Subject: [PATCH 1/3] Use the correct return type in runTest in 11_texture_driver sample. (#1546) Fixes SWDEV-203394. Currently in runTest() returns true, even if the texture reference copy does not happen. Using the existing testResult Flag to return from runTest(). --- samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) mode change 100644 => 100755 samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp diff --git a/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp b/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp old mode 100644 new mode 100755 index 2cb9877cac..b3c1ef5d0c --- a/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp +++ b/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp @@ -30,7 +30,7 @@ THE SOFTWARE. #define fileName "tex2dKernel.code" texture tex; -bool testResult = false; +bool testResult = true; #define HIP_CHECK(cmd) \ { \ @@ -126,7 +126,7 @@ bool runTest(int argc, char** argv) { } hipFree(dData); hipFreeArray(array); - return true; + return testResult; } int main(int argc, char** argv) { From 033960c86e56a7bd048c7d21334e436127d199a5 Mon Sep 17 00:00:00 2001 From: satyanveshd <53337087+satyanveshd@users.noreply.github.com> Date: Thu, 24 Oct 2019 17:44:47 +0530 Subject: [PATCH 2/3] Fix occupany APIs (#1560) Addresses SWDEV-205006 --- samples/2_Cookbook/13_occupancy/occupancy.cpp | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/samples/2_Cookbook/13_occupancy/occupancy.cpp b/samples/2_Cookbook/13_occupancy/occupancy.cpp index a9f4e198b0..01fa7aafed 100644 --- a/samples/2_Cookbook/13_occupancy/occupancy.cpp +++ b/samples/2_Cookbook/13_occupancy/occupancy.cpp @@ -56,9 +56,9 @@ void launchKernel(float* C, float* A, float* B, bool manual){ const unsigned threadsperblock = 32; const unsigned blocks = (NUM/threadsperblock)+1; - uint32_t mingridSize = 0; - uint32_t gridSize = 0; - uint32_t blockSize = 0; + int mingridSize = 0; + int gridSize = 0; + int blockSize = 0; if (manual){ blockSize = threadsperblock; @@ -86,7 +86,7 @@ void launchKernel(float* C, float* A, float* B, bool manual){ printf("kernel Execution time = %6.3fms\n", eventMs); //Calculate Occupancy - uint32_t numBlock = 0; + int numBlock = 0; HIP_CHECK(hipOccupancyMaxActiveBlocksPerMultiprocessor(&numBlock, multiply, blockSize, 0)); if(devProp.maxThreadsPerMultiProcessor){ From bee4958708b4897387cc330c682b135cda970683 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Thu, 24 Oct 2019 05:15:51 -0700 Subject: [PATCH 3/3] Add HIP checks in texture driver sample (#1581) --- .../11_texture_driver/texture2dDrv.cpp | 26 +++++++++---------- 1 file changed, 13 insertions(+), 13 deletions(-) diff --git a/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp b/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp index b3c1ef5d0c..3be6d12b31 100755 --- a/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp +++ b/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp @@ -62,7 +62,7 @@ bool runTest(int argc, char** argv) { desc.NumChannels = 1; desc.Width = width; desc.Height = height; - hipArrayCreate(&array, &desc); + HIP_CHECK(hipArrayCreate(&array, &desc)); hip_Memcpy2D copyParam; memset(©Param, 0, sizeof(copyParam)); @@ -73,19 +73,19 @@ bool runTest(int argc, char** argv) { copyParam.srcPitch = width * sizeof(float); copyParam.WidthInBytes = copyParam.srcPitch; copyParam.Height = height; - hipMemcpyParam2D(©Param); + HIP_CHECK(hipMemcpyParam2D(©Param)); textureReference* texref; - hipModuleGetTexRef(&texref, Module, "tex"); - hipTexRefSetAddressMode(texref, 0, hipAddressModeWrap); - hipTexRefSetAddressMode(texref, 1, hipAddressModeWrap); - hipTexRefSetFilterMode(texref, hipFilterModePoint); - hipTexRefSetFlags(texref, 0); - hipTexRefSetFormat(texref, HIP_AD_FORMAT_FLOAT, 1); - hipTexRefSetArray(texref, array, HIP_TRSA_OVERRIDE_FORMAT); + HIP_CHECK(hipModuleGetTexRef(&texref, Module, "tex")); + HIP_CHECK(hipTexRefSetAddressMode(texref, 0, hipAddressModeWrap)); + HIP_CHECK(hipTexRefSetAddressMode(texref, 1, hipAddressModeWrap)); + HIP_CHECK(hipTexRefSetFilterMode(texref, hipFilterModePoint)); + HIP_CHECK(hipTexRefSetFlags(texref, 0)); + HIP_CHECK(hipTexRefSetFormat(texref, HIP_AD_FORMAT_FLOAT, 1)); + HIP_CHECK(hipTexRefSetArray(texref, array, HIP_TRSA_OVERRIDE_FORMAT)); float* dData = NULL; - hipMalloc((void**)&dData, size); + HIP_CHECK(hipMalloc((void**)&dData, size)); struct { void* _Ad; @@ -112,7 +112,7 @@ bool runTest(int argc, char** argv) { float* hOutputData = (float*)malloc(size); memset(hOutputData, 0, size); - hipMemcpy(hOutputData, dData, size, hipMemcpyDeviceToHost); + HIP_CHECK(hipMemcpy(hOutputData, dData, size, hipMemcpyDeviceToHost)); for (int i = 0; i < height; i++) { for (int j = 0; j < width; j++) { @@ -124,8 +124,8 @@ bool runTest(int argc, char** argv) { } } } - hipFree(dData); - hipFreeArray(array); + HIP_CHECK(hipFree(dData)); + HIP_CHECK(hipFreeArray(array)); return testResult; }