From f1e3cbdc6c1ebc8e9be557edefaa8d9b1f5beb8e Mon Sep 17 00:00:00 2001 From: Lin Jiang Date: Wed, 12 Oct 2022 19:32:38 +0800 Subject: [PATCH 1/7] [Lang] [bug] Change the format string of 64bit integer types from %lld/u to %ld/u --- taichi/ir/type_utils.cpp | 4 ++-- tests/python/test_print.py | 22 ++++++++++++++++++++++ 2 files changed, 24 insertions(+), 2 deletions(-) diff --git a/taichi/ir/type_utils.cpp b/taichi/ir/type_utils.cpp index b9ae8a0ab2a0f..9c9d858c174a3 100644 --- a/taichi/ir/type_utils.cpp +++ b/taichi/ir/type_utils.cpp @@ -116,9 +116,9 @@ std::string data_type_format(DataType dt) { } else if (dt->is_primitive(PrimitiveTypeID::i64)) { // Use %lld on Windows. // Discussion: https://github.com/taichi-dev/taichi/issues/2522 - return "%lld"; + return "%ld"; } else if (dt->is_primitive(PrimitiveTypeID::u64)) { - return "%llu"; + return "%lu"; } else if (dt->is_primitive(PrimitiveTypeID::f32)) { return "%f"; } else if (dt->is_primitive(PrimitiveTypeID::f64)) { diff --git a/tests/python/test_print.py b/tests/python/test_print.py index 49244dd5f14ff..6b24d06d6f63a 100644 --- a/tests/python/test_print.py +++ b/tests/python/test_print.py @@ -160,3 +160,25 @@ def func(i: ti.i32, f: ti.f32): func(123, 4.56) ti.sync() + + +@test_utils.test(arch=[ti.cpu, ti.cuda, ti.vulkan], exclude=[vk_on_mac], debug=True) +def test_print_u64(): + + @ti.kernel + def func(i: ti.u64): + print("i = ", i) + + func(2 ** 64 - 1) + ti.sync() + + +@test_utils.test(arch=[ti.cpu, ti.cuda, ti.vulkan], exclude=[vk_on_mac], debug=True) +def test_print_i64(): + + @ti.kernel + def func(i: ti.i64): + print("i = ", i) + + func(-2 ** 63) + ti.sync() From e486777010f782ea348e5b1173f0c632fa7dfc5e Mon Sep 17 00:00:00 2001 From: Lin Jiang Date: Wed, 12 Oct 2022 20:09:11 +0800 Subject: [PATCH 2/7] fix --- .github/workflows/scripts/win_build.ps1 | 4 ++-- taichi/codegen/spirv/spirv_codegen.cpp | 2 +- taichi/ir/type_utils.cpp | 12 ++++++------ taichi/ir/type_utils.h | 3 ++- tests/python/test_print.py | 4 ++-- 5 files changed, 13 insertions(+), 12 deletions(-) diff --git a/.github/workflows/scripts/win_build.ps1 b/.github/workflows/scripts/win_build.ps1 index 13da3ce821219..df83454c3f5fd 100644 --- a/.github/workflows/scripts/win_build.ps1 +++ b/.github/workflows/scripts/win_build.ps1 @@ -81,9 +81,9 @@ if ($installVulkan) { -OutFile VulkanSDK.exe $installer = Start-Process -FilePath VulkanSDK.exe -Wait -PassThru -ArgumentList @("/S") $installer.WaitForExit(); + $env:VULKAN_SDK = "C:\VulkanSDK\1.2.189.0" + $env:PATH += ";$env:VULKAN_SDK\Bin" } - $env:VULKAN_SDK = "C:\VulkanSDK\1.2.189.0" - $env:PATH += ";$env:VULKAN_SDK\Bin" $env:TAICHI_CMAKE_ARGS += " -DTI_WITH_VULKAN:BOOL=ON" } diff --git a/taichi/codegen/spirv/spirv_codegen.cpp b/taichi/codegen/spirv/spirv_codegen.cpp index bcd83fc8c4043..39e06fc9cb23a 100644 --- a/taichi/codegen/spirv/spirv_codegen.cpp +++ b/taichi/codegen/spirv/spirv_codegen.cpp @@ -161,7 +161,7 @@ class TaskCodegen : public IRVisitor { auto value = ir_->query_value(arg_stmt->raw_name()); vals.push_back(value); - formats += data_type_format(arg_stmt->ret_type); + formats += data_type_format(arg_stmt->ret_type, Arch::vulkan); } else { auto arg_str = std::get(content); formats += arg_str; diff --git a/taichi/ir/type_utils.cpp b/taichi/ir/type_utils.cpp index 9c9d858c174a3..90ec2e239c6be 100644 --- a/taichi/ir/type_utils.cpp +++ b/taichi/ir/type_utils.cpp @@ -89,16 +89,16 @@ std::string tensor_type_format_helper(const std::vector &shape, return fmt; } -std::string tensor_type_format(DataType t) { +std::string tensor_type_format(DataType t, Arch arch) { TI_ASSERT(t->is()); auto tensor_type = t->as(); auto shape = tensor_type->get_shape(); auto element_type = tensor_type->get_element_type(); - auto element_type_format = data_type_format(element_type); + auto element_type_format = data_type_format(element_type, arch); return tensor_type_format_helper(shape, element_type_format, 0); } -std::string data_type_format(DataType dt) { +std::string data_type_format(DataType dt, Arch arch) { if (dt->is_primitive(PrimitiveTypeID::i8)) { // i8/u8 is converted to i16/u16 before printing, because CUDA doesn't // support the "%hhd"/"%hhu" specifiers. @@ -116,9 +116,9 @@ std::string data_type_format(DataType dt) { } else if (dt->is_primitive(PrimitiveTypeID::i64)) { // Use %lld on Windows. // Discussion: https://github.com/taichi-dev/taichi/issues/2522 - return "%ld"; + return arch == Arch::vulkan ? "%ld" : "%lld"; } else if (dt->is_primitive(PrimitiveTypeID::u64)) { - return "%lu"; + return arch == Arch::vulkan ? "%lu" : "%llu"; } else if (dt->is_primitive(PrimitiveTypeID::f32)) { return "%f"; } else if (dt->is_primitive(PrimitiveTypeID::f64)) { @@ -131,7 +131,7 @@ std::string data_type_format(DataType dt) { // TaskCodeGenCUDA::visit(PrintStmt *stmt) for more details. return "%f"; } else if (dt->is()) { - return tensor_type_format(dt); + return tensor_type_format(dt, arch); } else { TI_NOT_IMPLEMENTED } diff --git a/taichi/ir/type_utils.h b/taichi/ir/type_utils.h index a674a9b136607..9dc46f091bf3f 100644 --- a/taichi/ir/type_utils.h +++ b/taichi/ir/type_utils.h @@ -2,6 +2,7 @@ #include "taichi/ir/type.h" #include "taichi/ir/type_factory.h" +#include "taichi/rhi/arch.h" namespace taichi::lang { @@ -11,7 +12,7 @@ TI_DLL_EXPORT std::string data_type_name(DataType t); TI_DLL_EXPORT int data_type_size(DataType t); -TI_DLL_EXPORT std::string data_type_format(DataType dt); +TI_DLL_EXPORT std::string data_type_format(DataType dt, Arch arch = Arch::x64); inline int data_type_bits(DataType t) { return data_type_size(t) * 8; diff --git a/tests/python/test_print.py b/tests/python/test_print.py index 6b24d06d6f63a..a463bdda19f25 100644 --- a/tests/python/test_print.py +++ b/tests/python/test_print.py @@ -167,7 +167,7 @@ def test_print_u64(): @ti.kernel def func(i: ti.u64): - print("i = ", i) + print("i =", i) func(2 ** 64 - 1) ti.sync() @@ -178,7 +178,7 @@ def test_print_i64(): @ti.kernel def func(i: ti.i64): - print("i = ", i) + print("i =", i) func(-2 ** 63) ti.sync() From e0881b07fd7d5c4da5fb9f4bb0459397c4a25ce3 Mon Sep 17 00:00:00 2001 From: Lin Jiang Date: Wed, 12 Oct 2022 20:12:41 +0800 Subject: [PATCH 3/7] fix --- .github/workflows/scripts/win_build.ps1 | 5 +++-- 1 file changed, 3 insertions(+), 2 deletions(-) diff --git a/.github/workflows/scripts/win_build.ps1 b/.github/workflows/scripts/win_build.ps1 index df83454c3f5fd..64c0661a57bb7 100644 --- a/.github/workflows/scripts/win_build.ps1 +++ b/.github/workflows/scripts/win_build.ps1 @@ -81,9 +81,10 @@ if ($installVulkan) { -OutFile VulkanSDK.exe $installer = Start-Process -FilePath VulkanSDK.exe -Wait -PassThru -ArgumentList @("/S") $installer.WaitForExit(); - $env:VULKAN_SDK = "C:\VulkanSDK\1.2.189.0" - $env:PATH += ";$env:VULKAN_SDK\Bin" + } + $env:VULKAN_SDK = "C:\VulkanSDK\1.2.189.0" + $env:PATH += ";$env:VULKAN_SDK\Bin" $env:TAICHI_CMAKE_ARGS += " -DTI_WITH_VULKAN:BOOL=ON" } From 3900e2752d89cc9cd07df6176786deb34aa2ab85 Mon Sep 17 00:00:00 2001 From: Lin Jiang Date: Wed, 12 Oct 2022 20:13:01 +0800 Subject: [PATCH 4/7] fix --- .github/workflows/scripts/win_build.ps1 | 1 - 1 file changed, 1 deletion(-) diff --git a/.github/workflows/scripts/win_build.ps1 b/.github/workflows/scripts/win_build.ps1 index 64c0661a57bb7..13da3ce821219 100644 --- a/.github/workflows/scripts/win_build.ps1 +++ b/.github/workflows/scripts/win_build.ps1 @@ -81,7 +81,6 @@ if ($installVulkan) { -OutFile VulkanSDK.exe $installer = Start-Process -FilePath VulkanSDK.exe -Wait -PassThru -ArgumentList @("/S") $installer.WaitForExit(); - } $env:VULKAN_SDK = "C:\VulkanSDK\1.2.189.0" $env:PATH += ";$env:VULKAN_SDK\Bin" From 52d9d3b84a1a3e6cb31bc0a42968d3ff083e153d Mon Sep 17 00:00:00 2001 From: "pre-commit-ci[bot]" <66853113+pre-commit-ci[bot]@users.noreply.github.com> Date: Wed, 12 Oct 2022 12:15:12 +0000 Subject: [PATCH 5/7] [pre-commit.ci] auto fixes from pre-commit.com hooks for more information, see https://pre-commit.ci --- tests/python/test_print.py | 14 ++++++++------ 1 file changed, 8 insertions(+), 6 deletions(-) diff --git a/tests/python/test_print.py b/tests/python/test_print.py index a463bdda19f25..dea7c0317c83a 100644 --- a/tests/python/test_print.py +++ b/tests/python/test_print.py @@ -162,23 +162,25 @@ def func(i: ti.i32, f: ti.f32): ti.sync() -@test_utils.test(arch=[ti.cpu, ti.cuda, ti.vulkan], exclude=[vk_on_mac], debug=True) +@test_utils.test(arch=[ti.cpu, ti.cuda, ti.vulkan], + exclude=[vk_on_mac], + debug=True) def test_print_u64(): - @ti.kernel def func(i: ti.u64): print("i =", i) - func(2 ** 64 - 1) + func(2**64 - 1) ti.sync() -@test_utils.test(arch=[ti.cpu, ti.cuda, ti.vulkan], exclude=[vk_on_mac], debug=True) +@test_utils.test(arch=[ti.cpu, ti.cuda, ti.vulkan], + exclude=[vk_on_mac], + debug=True) def test_print_i64(): - @ti.kernel def func(i: ti.i64): print("i =", i) - func(-2 ** 63) + func(-2**63) ti.sync() From 6de5c7f48ffc202262f92aa671c2e5ec7fa822ed Mon Sep 17 00:00:00 2001 From: Lin Jiang Date: Wed, 12 Oct 2022 20:46:29 +0800 Subject: [PATCH 6/7] fix --- taichi/ir/type_utils.cpp | 4 +++- tests/python/test_print.py | 2 +- 2 files changed, 4 insertions(+), 2 deletions(-) diff --git a/taichi/ir/type_utils.cpp b/taichi/ir/type_utils.cpp index 90ec2e239c6be..2fdb816d518d4 100644 --- a/taichi/ir/type_utils.cpp +++ b/taichi/ir/type_utils.cpp @@ -116,8 +116,10 @@ std::string data_type_format(DataType dt, Arch arch) { } else if (dt->is_primitive(PrimitiveTypeID::i64)) { // Use %lld on Windows. // Discussion: https://github.com/taichi-dev/taichi/issues/2522 - return arch == Arch::vulkan ? "%ld" : "%lld"; + // Vulkan does not support printing 64-bit signed integer + return "%lld"; } else if (dt->is_primitive(PrimitiveTypeID::u64)) { + // Vulkan requires %lu to print 64-bit unsigned integer return arch == Arch::vulkan ? "%lu" : "%llu"; } else if (dt->is_primitive(PrimitiveTypeID::f32)) { return "%f"; diff --git a/tests/python/test_print.py b/tests/python/test_print.py index dea7c0317c83a..89cea2da18841 100644 --- a/tests/python/test_print.py +++ b/tests/python/test_print.py @@ -182,5 +182,5 @@ def test_print_i64(): def func(i: ti.i64): print("i =", i) - func(-2**63) + func(-2**63 + 2 ** 31) ti.sync() From 01f507fbe47441c0911bcecdfffce2cbc313c066 Mon Sep 17 00:00:00 2001 From: "pre-commit-ci[bot]" <66853113+pre-commit-ci[bot]@users.noreply.github.com> Date: Wed, 12 Oct 2022 12:50:06 +0000 Subject: [PATCH 7/7] [pre-commit.ci] auto fixes from pre-commit.com hooks for more information, see https://pre-commit.ci --- tests/python/test_print.py | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/tests/python/test_print.py b/tests/python/test_print.py index 89cea2da18841..039cdd8481692 100644 --- a/tests/python/test_print.py +++ b/tests/python/test_print.py @@ -182,5 +182,5 @@ def test_print_i64(): def func(i: ti.i64): print("i =", i) - func(-2**63 + 2 ** 31) + func(-2**63 + 2**31) ti.sync()