From 654e9d92008d757b08b701471919ddcad89c6834 Mon Sep 17 00:00:00 2001 From: SrinivasaRao Date: Thu, 22 Jun 2023 17:32:36 +0530 Subject: [PATCH] SWDEV-403479 - [catch2][dtest] Adding Buffered compiler option to Existing Printf Tests Change-Id: I4fc422bd7d5ee1896743978fe078ac3f8d57e319 --- catch/include/hip_test_defgroups.hh | 7 ++ catch/unit/printf/CMakeLists.txt | 14 ++- catch/unit/printf/printfFlagsNonHost.cc | 61 +++++++++ catch/unit/printf/printfFlagsNonHost_exe.cc | 39 ++++++ catch/unit/printf/printfSpecifiersNonHost.cc | 116 ++++++++++++++++++ .../printf/printfSpecifiersNonHost_exe.cc | 66 ++++++++++ 6 files changed, 302 insertions(+), 1 deletion(-) create mode 100644 catch/unit/printf/printfFlagsNonHost.cc create mode 100644 catch/unit/printf/printfFlagsNonHost_exe.cc create mode 100644 catch/unit/printf/printfSpecifiersNonHost.cc create mode 100644 catch/unit/printf/printfSpecifiersNonHost_exe.cc diff --git a/catch/include/hip_test_defgroups.hh b/catch/include/hip_test_defgroups.hh index 5d4a31918..1cbab2f7f 100644 --- a/catch/include/hip_test_defgroups.hh +++ b/catch/include/hip_test_defgroups.hh @@ -142,3 +142,10 @@ THE SOFTWARE. * This section describes tests for the Vector type functions and operators. * @} */ + +/** + * @defgroup PrintfTest Printf API Management + * @{ + * This section describes the various Printf use case Scenarios. + * @} + */ diff --git a/catch/unit/printf/CMakeLists.txt b/catch/unit/printf/CMakeLists.txt index 74d4421a7..c3efbb9c9 100644 --- a/catch/unit/printf/CMakeLists.txt +++ b/catch/unit/printf/CMakeLists.txt @@ -2,8 +2,15 @@ set(TEST_SRC printfFlags.cc printfSpecifiers.cc + printfFlagsNonHost.cc + printfSpecifiersNonHost.cc ) - +if(UNIX) + if(HIP_PLATFORM MATCHES "amd") + set_source_files_properties(printfFlagsNonHost.cc PROPERTIES COMPILE_OPTIONS "-mprintf-kind=buffered") + set_source_files_properties(printfSpecifiersNonHost.cc PROPERTIES COMPILE_OPTIONS "-mprintf-kind=buffered") + endif() +endif() if(HIP_PLATFORM MATCHES "amd") hip_add_exe_to_target(NAME printfTests @@ -20,6 +27,11 @@ endif() # Standalone exes add_executable(printfFlags_exe EXCLUDE_FROM_ALL printfFlags_exe.cc) add_executable(printfSpecifiers_exe EXCLUDE_FROM_ALL printfSpecifiers_exe.cc) +add_executable(printfFlagsNonHost_exe EXCLUDE_FROM_ALL printfFlagsNonHost_exe.cc) +add_executable(printfSpecifiersNonHost_exe EXCLUDE_FROM_ALL printfSpecifiersNonHost_exe.cc) add_dependencies(build_tests printfFlags_exe) add_dependencies(build_tests printfSpecifiers_exe) +add_dependencies(build_tests printfFlagsNonHost_exe) +add_dependencies(build_tests printfSpecifiersNonHost_exe) + diff --git a/catch/unit/printf/printfFlagsNonHost.cc b/catch/unit/printf/printfFlagsNonHost.cc new file mode 100644 index 000000000..cea2ea607 --- /dev/null +++ b/catch/unit/printf/printfFlagsNonHost.cc @@ -0,0 +1,61 @@ +/*Copyright (c) 2023 Advanced Micro Devices, Inc. All rights reserved. +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ + +#include +#include +#include + +/** +* @addtogroup printf +* @{ +* @ingroup PrintfTest +* `int printf()` - +* Method to print the content on output device. +*/ +/** +* Test Description +* ------------------------ +* - Test case to verify the printf return value from other process for the compiler option -mprintf-kind=buffered +* - Fetch the printf content from a process. Compare it with reference string. +* Test source +* ------------------------ +* - catch/unit/printf/printfFlagsNonHost.cc +* Test requirements +* ------------------------ +* - HIP_VERSION >= 5.7 +*/ + +TEST_CASE("Unit_Buffered_Printf_Flags") { + std::string reference(R"here(00000042 +-0000042 +00000042 +0123.456 ++0000042 +-42 ++0000042 +xyzzy +-42 +00000042 + 00000042 +)here"); + + hip::SpawnProc proc("printfFlagsNonHost_exe", true); + REQUIRE(proc.run() == 0); + REQUIRE(proc.getOutput() == reference); +} + diff --git a/catch/unit/printf/printfFlagsNonHost_exe.cc b/catch/unit/printf/printfFlagsNonHost_exe.cc new file mode 100644 index 000000000..23f4e87d3 --- /dev/null +++ b/catch/unit/printf/printfFlagsNonHost_exe.cc @@ -0,0 +1,39 @@ +/* +Copyright (c) 2023 Advanced Micro Devices, Inc. All rights reserved. +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ + +#include + +__global__ void test_kernel() { + printf("%08d\n", 42); + printf("%08i\n", -42); + printf("%08u\n", 42); + printf("%08g\n", 123.456); + printf("%0+8d\n", 42); + printf("%+d\n", -42); + printf("%+08d\n", 42); + printf("%-8s\n", "xyzzy"); + printf("% i\n", -42); + printf("%-16.8d\n", 42); + printf("%16.8d\n", 42); +} + +int main() { + test_kernel<<<1, 1>>>(); + static_cast(hipDeviceSynchronize()); +} diff --git a/catch/unit/printf/printfSpecifiersNonHost.cc b/catch/unit/printf/printfSpecifiersNonHost.cc new file mode 100644 index 000000000..7c6559641 --- /dev/null +++ b/catch/unit/printf/printfSpecifiersNonHost.cc @@ -0,0 +1,116 @@ +/* +Copyright (c) 2023 Advanced Micro Devices, Inc. All rights reserved. + +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: + +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. + +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ + +#include +#include +#include + +/** +* @addtogroup printf +* @{ +* @ingroup PrintfTest +* `int printf()` - +* Method to print the content on output device. +*/ +/** +* Test Description +* ------------------------ +* - Test case to verify the different format specifiers. Test case should compile with the compiler option -mprintf-kind=buffered +* - Fetch the printf content from a process which will verify format specifier. Compare it with reference string. +* Test source +* ------------------------ +* - catch/unit/printf/printfSpecifiersNonHost.cc +* Test requirements +* ------------------------ +* - HIP_VERSION >= 5.7 +*/ + +TEST_CASE("Unit_Buffered_Printf_Specifier") { +#ifdef __HIP_PLATFORM_NVIDIA__ + std::string reference(R"here(xyzzy +% +hello % world +%s +%s0xf01dab1eca55e77e +%cxyzzy +sep +-42 +42 +123.456000 +-123.456000 +-1.234560e+02 +1.234560E+02 +123.456 +-123.456 +x +(null) +(nil) +3.14159000 hello 0xf01dab1eca55e77e +)here"); +#elif !defined(_WIN32) + std::string reference(R"here(xyzzy +% +hello % world +%s +%s0xf01dab1eca55e77e +%cxyzzy +sep +-42 +42 +123.456000 +-123.456000 +-1.234560e+02 +1.234560E+02 +123.456 +-123.456 +x + +(nil) +3.14159000 hello 0xf01dab1eca55e77e +)here"); +#else + std::string reference(R"here(xyzzy +% +hello % world +%s +%sF01DAB1ECA55E77E +%cxyzzy +sep +-42 +42 +123.456000 +-123.456000 +-1.234560e+02 +1.234560E+02 +123.456 +-123.456 +x + +0000000000000000 +3.14159000 hello F01DAB1ECA55E77E +)here"); +#endif + + hip::SpawnProc proc("printfSpecifiersNonHost_exe", true); + REQUIRE(0 == proc.run()); + REQUIRE(proc.getOutput() == reference); +} diff --git a/catch/unit/printf/printfSpecifiersNonHost_exe.cc b/catch/unit/printf/printfSpecifiersNonHost_exe.cc new file mode 100644 index 000000000..5c6277790 --- /dev/null +++ b/catch/unit/printf/printfSpecifiersNonHost_exe.cc @@ -0,0 +1,66 @@ +/* +Copyright (c) 2023 Advanced Micro Devices, Inc. All rights reserved. + +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: + +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. + +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ + +#include + +__global__ void test_kernel() { + const char* N = nullptr; + const char* s = "hello world"; + printf("xyzzy\n"); + printf("%%\n"); + printf("hello %% world\n"); + printf("%%s\n"); + // Two special tests to make sure that the compiler pass correctly + // skips over a '%%' without affecting the logic for locating + // string arguments. + printf("%%s%p\n", (void*)0xf01dab1eca55e77e); + printf("%%c%s\n", "xyzzy"); + printf("%c%c%c\n", 's', 'e', 'p'); + printf("%d\n", -42); + printf("%u\n", 42); + printf("%f\n", 123.456); +#ifdef __HIP_PLATFORM_AMD__ + printf("%F\n", -123.456); +#else + printf("%f\n", -123.456); +#endif + printf("%e\n", -123.456); + printf("%E\n", 123.456); + printf("%g\n", 123.456); + printf("%G\n", -123.456); + printf("%c\n", 'x'); + printf("%s\n", N); + printf("%p\n", (void *)N); +#ifdef __HIP_PLATFORM_AMD__ + printf("%.*f %*.*s %p\n", 8, 3.14159, 8, 5, s, (void*)0xf01dab1eca55e77e); +#else + // In Cuda, printf doesn't support %.*, %*.* + printf("%.8f %8.5s %p\n", 3.14159, s, (void*)0xf01dab1eca55e77e); +#endif +} + +int main() { + test_kernel<<<1, 1>>>(); + static_cast(hipDeviceSynchronize()); + return 0; +} +