From d82ea9045792a1f14a04a03c0b9292bb4c7235c5 Mon Sep 17 00:00:00 2001 From: PGZXB <420254146@qq.com> Date: Sat, 23 Apr 2022 10:23:44 +0800 Subject: [PATCH 01/12] [bug] Fix bug that building with TI_EXPORT_CORE:BOOL=ON failed (#4850) * Fix bug that building with TI_EXPORT_CORE:BOOL=ON failed * [pre-commit.ci] auto fixes from pre-commit.com hooks for more information, see https://pre-commit.ci Co-authored-by: pre-commit-ci[bot] <66853113+pre-commit-ci[bot]@users.noreply.github.com> --- taichi/analysis/offline_cache_util.cpp | 175 +++++++++++++++++++++++++ taichi/analysis/offline_cache_util.h | 16 +++ taichi/codegen/codegen_llvm.cpp | 1 + taichi/ir/expression_printer.h | 2 +- taichi/llvm/llvm_offline_cache.cpp | 159 ---------------------- taichi/llvm/llvm_offline_cache.h | 3 - 6 files changed, 193 insertions(+), 163 deletions(-) create mode 100644 taichi/analysis/offline_cache_util.cpp create mode 100644 taichi/analysis/offline_cache_util.h diff --git a/taichi/analysis/offline_cache_util.cpp b/taichi/analysis/offline_cache_util.cpp new file mode 100644 index 0000000000000..c73ef3440389c --- /dev/null +++ b/taichi/analysis/offline_cache_util.cpp @@ -0,0 +1,175 @@ +#include "offline_cache_util.h" + +#include "taichi/common/core.h" +#include "taichi/common/serialization.h" +#include "taichi/ir/snode.h" +#include "taichi/ir/transforms.h" +#include "taichi/program/compile_config.h" +#include "taichi/program/kernel.h" + +#include "picosha2.h" + +#include + +namespace taichi { +namespace lang { + +static std::vector get_offline_cache_key_of_compile_config( + CompileConfig *config) { + TI_ASSERT(config); + BinaryOutputSerializer serializer; + serializer.initialize(); + serializer(config->arch); + serializer(config->debug); + serializer(config->cfg_optimization); + serializer(config->check_out_of_bound); + serializer(config->opt_level); + serializer(config->external_optimization_level); + serializer(config->packed); + serializer(config->move_loop_invariant_outside_if); + serializer(config->demote_dense_struct_fors); + serializer(config->advanced_optimization); + serializer(config->constant_folding); + serializer(config->fast_math); + serializer(config->flatten_if); + serializer(config->make_thread_local); + serializer(config->make_block_local); + serializer(config->detect_read_only); + serializer(config->default_fp->to_string()); + serializer(config->default_ip.to_string()); + if (arch_is_cpu(config->arch)) { + serializer(config->default_cpu_block_dim); + serializer(config->cpu_max_num_threads); + } else if (arch_is_gpu(config->arch)) { + serializer(config->default_gpu_block_dim); + serializer(config->gpu_max_reg); + serializer(config->saturating_grid_dim); + serializer(config->cpu_max_num_threads); + } + serializer(config->ad_stack_size); + serializer(config->default_ad_stack_size); + serializer(config->random_seed); + if (config->arch == Arch::cc) { + serializer(config->cc_compile_cmd); + serializer(config->cc_link_cmd); + } else if (config->arch == Arch::opengl) { + serializer(config->allow_nv_shader_extension); + serializer(config->use_gles); + } + serializer(config->make_mesh_block_local); + serializer(config->optimize_mesh_reordered_mapping); + serializer(config->mesh_localize_to_end_mapping); + serializer(config->mesh_localize_from_end_mapping); + serializer(config->mesh_localize_all_attr_mappings); + serializer(config->demote_no_access_mesh_fors); + serializer(config->experimental_auto_mesh_local); + serializer(config->auto_mesh_local_default_occupacy); + serializer.finalize(); + + return serializer.data; +} + +static void get_offline_cache_key_of_snode_impl( + SNode *snode, + BinaryOutputSerializer &serializer) { + for (auto &c : snode->ch) { + get_offline_cache_key_of_snode_impl(c.get(), serializer); + } + for (int i = 0; i < taichi_max_num_indices; ++i) { + auto &extractor = snode->extractors[i]; + serializer(extractor.num_elements_from_root); + serializer(extractor.shape); + serializer(extractor.acc_shape); + serializer(extractor.num_bits); + serializer(extractor.acc_offset); + serializer(extractor.active); + } + serializer(snode->index_offsets); + serializer(snode->num_active_indices); + serializer(snode->physical_index_position); + serializer(snode->id); + serializer(snode->depth); + serializer(snode->name); + serializer(snode->num_cells_per_container); + serializer(snode->total_num_bits); + serializer(snode->total_bit_start); + serializer(snode->chunk_size); + serializer(snode->cell_size_bytes); + serializer(snode->offset_bytes_in_parent_cell); + if (snode->physical_type) { + serializer(snode->physical_type->to_string()); + } + serializer(snode->dt->to_string()); + serializer(snode->has_ambient); + if (!snode->ambient_val.dt->is_primitive(PrimitiveTypeID::unknown)) { + serializer(snode->ambient_val.stringify()); + } + if (snode->grad_info && !snode->grad_info->is_primal()) { + if (auto *grad_snode = snode->grad_info->grad_snode()) { + get_offline_cache_key_of_snode_impl(grad_snode, serializer); + } + } + if (snode->exp_snode) { + get_offline_cache_key_of_snode_impl(snode->exp_snode, serializer); + } + serializer(snode->bit_offset); + serializer(snode->placing_shared_exp); + serializer(snode->owns_shared_exponent); + for (auto s : snode->exponent_users) { + get_offline_cache_key_of_snode_impl(s, serializer); + } + if (snode->currently_placing_exp_snode) { + get_offline_cache_key_of_snode_impl(snode->currently_placing_exp_snode, + serializer); + } + if (snode->currently_placing_exp_snode_dtype) { + serializer(snode->currently_placing_exp_snode_dtype->to_string()); + } + serializer(snode->is_bit_level); + serializer(snode->is_path_all_dense); + serializer(snode->node_type_name); + serializer(snode->type); + serializer(snode->_morton); + serializer(snode->get_snode_tree_id()); +} + +std::string get_hashed_offline_cache_key_of_snode(SNode *snode) { + TI_ASSERT(snode); + + BinaryOutputSerializer serializer; + serializer.initialize(); + get_offline_cache_key_of_snode_impl(snode, serializer); + serializer.finalize(); + + picosha2::hash256_one_by_one hasher; + hasher.process(serializer.data.begin(), serializer.data.end()); + hasher.finish(); + + return picosha2::get_hash_hex_string(hasher); +} + +std::string get_hashed_offline_cache_key(CompileConfig *config, + Kernel *kernel) { + std::string kernel_ast_string; + if (kernel) { + irpass::gen_offline_cache_key(kernel->program, kernel->ir.get(), + &kernel_ast_string); + } + + std::vector compile_config_key; + if (config) { + compile_config_key = get_offline_cache_key_of_compile_config(config); + } + + picosha2::hash256_one_by_one hasher; + hasher.process(compile_config_key.begin(), compile_config_key.end()); + hasher.process(kernel_ast_string.begin(), kernel_ast_string.end()); + hasher.finish(); + + auto res = picosha2::get_hash_hex_string(hasher); + res.insert(res.begin(), kernel->grad ? 'g' : 'n'); + return res; +} + +} // namespace lang +} // namespace taichi diff --git a/taichi/analysis/offline_cache_util.h b/taichi/analysis/offline_cache_util.h new file mode 100644 index 0000000000000..03ca580d96ef5 --- /dev/null +++ b/taichi/analysis/offline_cache_util.h @@ -0,0 +1,16 @@ +#pragma once + +#include + +namespace taichi { +namespace lang { + +struct CompileConfig; +class Kernel; +class SNode; + +std::string get_hashed_offline_cache_key_of_snode(SNode *snode); +std::string get_hashed_offline_cache_key(CompileConfig *config, Kernel *kernel); + +} // namespace lang +} // namespace taichi diff --git a/taichi/codegen/codegen_llvm.cpp b/taichi/codegen/codegen_llvm.cpp index 2d251d6b084eb..694edca7c6401 100644 --- a/taichi/codegen/codegen_llvm.cpp +++ b/taichi/codegen/codegen_llvm.cpp @@ -1,4 +1,5 @@ #ifdef TI_WITH_LLVM +#include "taichi/analysis/offline_cache_util.h" #include "taichi/codegen/codegen_llvm.h" #include "taichi/llvm/llvm_offline_cache.h" #include "taichi/ir/statements.h" diff --git a/taichi/ir/expression_printer.h b/taichi/ir/expression_printer.h index 2141d82facdcd..b2215db252611 100644 --- a/taichi/ir/expression_printer.h +++ b/taichi/ir/expression_printer.h @@ -4,7 +4,7 @@ #include "taichi/ir/expression.h" #include "taichi/ir/frontend_ir.h" #include "taichi/program/program.h" -#include "taichi/llvm/llvm_offline_cache.h" +#include "taichi/analysis/offline_cache_util.h" namespace taichi { namespace lang { diff --git a/taichi/llvm/llvm_offline_cache.cpp b/taichi/llvm/llvm_offline_cache.cpp index 662e126030a82..c7e625b35a673 100644 --- a/taichi/llvm/llvm_offline_cache.cpp +++ b/taichi/llvm/llvm_offline_cache.cpp @@ -6,168 +6,9 @@ #include "llvm/IR/Module.h" #include "taichi/ir/transforms.h" -#include "picosha2.h" - namespace taichi { namespace lang { -static TI_FORCE_INLINE std::vector -get_offline_cache_key_of_compile_config(CompileConfig *config) { - TI_ASSERT(config); - BinaryOutputSerializer serializer; - serializer.initialize(); - serializer(config->arch); - serializer(config->debug); - serializer(config->cfg_optimization); - serializer(config->check_out_of_bound); - serializer(config->opt_level); - serializer(config->external_optimization_level); - serializer(config->packed); - serializer(config->move_loop_invariant_outside_if); - serializer(config->demote_dense_struct_fors); - serializer(config->advanced_optimization); - serializer(config->constant_folding); - serializer(config->fast_math); - serializer(config->flatten_if); - serializer(config->make_thread_local); - serializer(config->make_block_local); - serializer(config->detect_read_only); - serializer(config->default_fp->to_string()); - serializer(config->default_ip.to_string()); - if (arch_is_cpu(config->arch)) { - serializer(config->default_cpu_block_dim); - serializer(config->cpu_max_num_threads); - } else if (arch_is_gpu(config->arch)) { - serializer(config->default_gpu_block_dim); - serializer(config->gpu_max_reg); - serializer(config->saturating_grid_dim); - serializer(config->cpu_max_num_threads); - } - serializer(config->ad_stack_size); - serializer(config->default_ad_stack_size); - serializer(config->random_seed); - if (config->arch == Arch::cc) { - serializer(config->cc_compile_cmd); - serializer(config->cc_link_cmd); - } else if (config->arch == Arch::opengl) { - serializer(config->allow_nv_shader_extension); - serializer(config->use_gles); - } - serializer(config->make_mesh_block_local); - serializer(config->optimize_mesh_reordered_mapping); - serializer(config->mesh_localize_to_end_mapping); - serializer(config->mesh_localize_from_end_mapping); - serializer(config->mesh_localize_all_attr_mappings); - serializer(config->demote_no_access_mesh_fors); - serializer(config->experimental_auto_mesh_local); - serializer(config->auto_mesh_local_default_occupacy); - serializer.finalize(); - - return serializer.data; -} - -static void get_offline_cache_key_of_snode_impl( - SNode *snode, - BinaryOutputSerializer &serializer) { - for (auto &c : snode->ch) { - get_offline_cache_key_of_snode_impl(c.get(), serializer); - } - for (int i = 0; i < taichi_max_num_indices; ++i) { - auto &extractor = snode->extractors[i]; - serializer(extractor.num_elements_from_root); - serializer(extractor.shape); - serializer(extractor.acc_shape); - serializer(extractor.num_bits); - serializer(extractor.acc_offset); - serializer(extractor.active); - } - serializer(snode->index_offsets); - serializer(snode->num_active_indices); - serializer(snode->physical_index_position); - serializer(snode->id); - serializer(snode->depth); - serializer(snode->name); - serializer(snode->num_cells_per_container); - serializer(snode->total_num_bits); - serializer(snode->total_bit_start); - serializer(snode->chunk_size); - serializer(snode->cell_size_bytes); - serializer(snode->offset_bytes_in_parent_cell); - if (snode->physical_type) { - serializer(snode->physical_type->to_string()); - } - serializer(snode->dt->to_string()); - serializer(snode->has_ambient); - if (!snode->ambient_val.dt->is_primitive(PrimitiveTypeID::unknown)) { - serializer(snode->ambient_val.stringify()); - } - if (snode->grad_info && !snode->grad_info->is_primal()) { - if (auto *grad_snode = snode->grad_info->grad_snode()) { - get_offline_cache_key_of_snode_impl(grad_snode, serializer); - } - } - if (snode->exp_snode) { - get_offline_cache_key_of_snode_impl(snode->exp_snode, serializer); - } - serializer(snode->bit_offset); - serializer(snode->placing_shared_exp); - serializer(snode->owns_shared_exponent); - for (auto s : snode->exponent_users) { - get_offline_cache_key_of_snode_impl(s, serializer); - } - if (snode->currently_placing_exp_snode) { - get_offline_cache_key_of_snode_impl(snode->currently_placing_exp_snode, - serializer); - } - if (snode->currently_placing_exp_snode_dtype) { - serializer(snode->currently_placing_exp_snode_dtype->to_string()); - } - serializer(snode->is_bit_level); - serializer(snode->is_path_all_dense); - serializer(snode->node_type_name); - serializer(snode->type); - serializer(snode->_morton); - serializer(snode->get_snode_tree_id()); -} - -std::string get_hashed_offline_cache_key_of_snode(SNode *snode) { - TI_ASSERT(snode); - - BinaryOutputSerializer serializer; - serializer.initialize(); - get_offline_cache_key_of_snode_impl(snode, serializer); - serializer.finalize(); - - picosha2::hash256_one_by_one hasher; - hasher.process(serializer.data.begin(), serializer.data.end()); - hasher.finish(); - - return picosha2::get_hash_hex_string(hasher); -} - -std::string get_hashed_offline_cache_key(CompileConfig *config, - Kernel *kernel) { - std::string kernel_ast_string; - if (kernel) { - irpass::gen_offline_cache_key(kernel->program, kernel->ir.get(), - &kernel_ast_string); - } - - std::vector compile_config_key; - if (config) { - compile_config_key = get_offline_cache_key_of_compile_config(config); - } - - picosha2::hash256_one_by_one hasher; - hasher.process(compile_config_key.begin(), compile_config_key.end()); - hasher.process(kernel_ast_string.begin(), kernel_ast_string.end()); - hasher.finish(); - - auto res = picosha2::get_hash_hex_string(hasher); - res.insert(res.begin(), kernel->grad ? 'g' : 'n'); - return res; -} - bool LlvmOfflineCacheFileReader::get_kernel_cache( LlvmOfflineCache::KernelCacheData &res, const std::string &key, diff --git a/taichi/llvm/llvm_offline_cache.h b/taichi/llvm/llvm_offline_cache.h index 0177fffde028f..0f98fd60227ed 100644 --- a/taichi/llvm/llvm_offline_cache.h +++ b/taichi/llvm/llvm_offline_cache.h @@ -9,9 +9,6 @@ namespace taichi { namespace lang { -std::string get_hashed_offline_cache_key_of_snode(SNode *snode); -std::string get_hashed_offline_cache_key(CompileConfig *config, Kernel *kernel); - struct LlvmOfflineCache { struct OffloadedTaskCacheData { std::string name; From 9178a5be1b4d6566cb1124776a488c2498520c64 Mon Sep 17 00:00:00 2001 From: Vissidarte-Herman <93570324+Vissidarte-Herman@users.noreply.github.com> Date: Tue, 26 Apr 2022 14:50:44 +0800 Subject: [PATCH 02/12] [Doc] Updated URL (#4847) * Updated URL * Replaced docs.taichi.graphics with docs.taichi-lang.org --- .github/ISSUE_TEMPLATE/bug_report.md | 2 +- .github/ISSUE_TEMPLATE/config.yml | 2 +- .github/ISSUE_TEMPLATE/question.md | 2 +- .github/pull_request_template.md | 4 ++-- CMakeLists.txt | 2 +- CONTRIBUTING.md | 4 ++-- README.md | 14 +++++++------- benchmarks/microbenchmarks/_items.py | 2 +- docs/lang/articles/basic/overview.md | 2 +- .../articles/contribution/contributor_guide.md | 2 +- docs/lang/articles/contribution/dev_install.md | 2 +- docs/lang/articles/reference.md | 4 ++-- python/taichi/_lib/utils.py | 2 +- python/taichi/_logging.py | 4 ++-- python/taichi/_main.py | 2 +- python/taichi/lang/impl.py | 6 +++--- python/taichi/lang/kernel_impl.py | 4 ++-- python/taichi/lang/matrix.py | 4 ++-- python/taichi/lang/misc.py | 4 ++-- python/taichi/lang/snode.py | 2 +- python/taichi/profiler/kernel_metrics.py | 2 +- python/taichi/profiler/kernel_profiler.py | 4 ++-- python/taichi/types/annotations.py | 2 +- taichi/analysis/verify.cpp | 2 +- taichi/backends/cuda/cupti_toolkit.cpp | 4 ++-- taichi/gui/x11.cpp | 2 +- taichi/ir/snode.h | 2 +- taichi/system/traceback.cpp | 2 +- taichi/transforms/reverse_segments.cpp | 2 +- tests/python/test_fields_builder.py | 2 +- tests/python/test_simple_matrix_slice.py | 2 +- 31 files changed, 48 insertions(+), 48 deletions(-) diff --git a/.github/ISSUE_TEMPLATE/bug_report.md b/.github/ISSUE_TEMPLATE/bug_report.md index 0c144f2eb4082..ba245f01619f8 100644 --- a/.github/ISSUE_TEMPLATE/bug_report.md +++ b/.github/ISSUE_TEMPLATE/bug_report.md @@ -7,7 +7,7 @@ assignees: '' --- - + **Describe the bug** A clear and concise description of what the bug is, ideally within 20 words. diff --git a/.github/ISSUE_TEMPLATE/config.yml b/.github/ISSUE_TEMPLATE/config.yml index d4675bdea7b29..eaddd6150ca49 100644 --- a/.github/ISSUE_TEMPLATE/config.yml +++ b/.github/ISSUE_TEMPLATE/config.yml @@ -2,7 +2,7 @@ blank_issues_enabled: true contact_links: - name: Contributor Guideline - url: https://docs.taichi.graphics/lang/articles/contributor_guide + url: https://docs.taichi-lang.org/lang/articles/contributor_guide about: Please check this out if you'd like to contribute by opening a PR :) - name: Taichi Forum url: https://forum.taichi.graphics diff --git a/.github/ISSUE_TEMPLATE/question.md b/.github/ISSUE_TEMPLATE/question.md index accca01a54df9..afe2bc3760767 100644 --- a/.github/ISSUE_TEMPLATE/question.md +++ b/.github/ISSUE_TEMPLATE/question.md @@ -12,6 +12,6 @@ Before asking a question, please first consider: - Searching Google - Searching [existing issues](https://github.com/taichi-dev/taichi/issues) -- Searching [Taichi Doc](https://docs.taichi.graphics/) +- Searching [Taichi Doc](https://docs.taichi-lang.org/) - Searching [Taichi Forum](https://forum.taichi.graphics/) --> diff --git a/.github/pull_request_template.md b/.github/pull_request_template.md index 32afd19a32305..b7f0ebaabc84d 100644 --- a/.github/pull_request_template.md +++ b/.github/pull_request_template.md @@ -4,11 +4,11 @@ Related issue = # Thank you for your contribution! If it is your first time contributing to Taichi, please read our Contributor Guidelines: - https://docs.taichi.graphics/lang/articles/contribution/contributor_guide + https://docs.taichi-lang.org/lang/articles/contributor_guide - Please always prepend your PR title with tags such as [CUDA], [Lang], [Doc], [Example]. For a complete list of valid PR tags, please check out https://github.com/taichi-dev/taichi/blob/master/misc/prtags.json. - Use upper-case tags (e.g., [Metal]) for PRs that change public APIs. Otherwise, please use lower-case tags (e.g., [metal]). -- More details: https://docs.taichi.graphics/lang/articles/contribution/contributor_guide#pr-title-format-and-tags +- More details: https://docs.taichi-lang.org/lang/articles/contributor_guide#pr-title-format-and-tags - Please fill in the issue number that this PR relates to. - If your PR fixes the issue **completely**, use the `close` or `fixes` prefix so that GitHub automatically closes the issue when the PR is merged. For example, diff --git a/CMakeLists.txt b/CMakeLists.txt index 474f8ce90348b..17db545ac8194 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -7,7 +7,7 @@ cmake_minimum_required(VERSION 3.12) project(taichi) if (NOT DEFINED TI_VERSION_MAJOR) - message(WARNING "It seems that you are running cmake manually, which may cause issues. Please use setup.py to build taichi from source, see https://docs.taichi.graphics/lang/articles/dev_install for more details.") + message(WARNING "It seems that you are running cmake manually, which may cause issues. Please use setup.py to build taichi from source, see https://docs.taichi-lang.org/lang/articles/dev_install for more details.") set(TI_VERSION_MAJOR 0) set(TI_VERSION_MINOR 0) set(TI_VERSION_PATCH 0) diff --git a/CONTRIBUTING.md b/CONTRIBUTING.md index b5c31c47530c4..b0d3ad75cb08a 100644 --- a/CONTRIBUTING.md +++ b/CONTRIBUTING.md @@ -1,12 +1,12 @@ # Contributing Guide -Thank you for your interest in contributing to Taichi! Please check out the [Contribution Guidelines](https://docs.taichi.graphics/lang/articles/contributor_guide) for how to make a contribution. +Thank you for your interest in contributing to Taichi! Please check out the [Contribution Guidelines](https://docs.taichi-lang.org/lang/articles/contributor_guide) for how to make a contribution. All contributors are expected to follow the [code of conduct](https://github.com/taichi-dev/taichi/blob/master/CODE_OF_CONDUCT.md). ## Developer installation -Taichi is developed mainly in C++17 and Python3. Please check out the [Developer Installation](https://docs.taichi.graphics/lang/articles/dev_install) to build Taichi from source. Note that Taichi is LLVM-10.0.0 dependent and that we recommend installing [our pre-built LLVM libraries](https://docs.taichi.graphics/lang/articles/dev_install#install-llvm) for your platform. +Taichi is developed mainly in C++17 and Python3. Please check out the [Developer Installation](https://docs.taichi-lang.org/lang/articles/dev_install) to build Taichi from source. Note that Taichi is LLVM-10.0.0 dependent and that we recommend installing [our pre-built LLVM libraries](https://docs.taichi-lang.org/lang/articles/dev_install#install-llvm) for your platform. ## Contribution opportunities diff --git a/README.md b/README.md index e7ec0345bc704..e2b0bcb172a70 100644 --- a/README.md +++ b/README.md @@ -36,10 +36,10 @@ The language has broad applications spanning real-time physical simulation, numb ## Why Taichi? - Built around Python: Taichi shares almost the same syntax with Python, allowing you to write algorithms with minimal language barrier. It is also well integrated into the Python ecosystem, including NumPy and PyTorch. -- Flexibility: Taichi provides a set of generic data containers known as *SNode* (/ˈsnoʊd/), an effective mechanism for composing hierarchical, multi-dimensional fields. This can cover many use patterns in numerical simulation (e.g. [spatially sparse computing](https://docs.taichi.graphics/lang/articles/sparse)). +- Flexibility: Taichi provides a set of generic data containers known as *SNode* (/ˈsnoʊd/), an effective mechanism for composing hierarchical, multi-dimensional fields. This can cover many use patterns in numerical simulation (e.g. [spatially sparse computing](https://docs.taichi-lang.org/lang/articles/sparse)). - Performance: With the `@ti.kernel` decorator, Taichi's JIT compiler automatically compiles your Python functions into efficient GPU or CPU machine code for parallel execution. - Portability: Write your code once and run it everywhere. Currently, Taichi supports most mainstream GPU APIs, such as CUDA and Vulkan. -- ... and many more features! A cross-platform, Vulkan-based 3D visualizer, [differentiable programming](https://docs.taichi.graphics/lang/articles/differentiable_programming), [quantized computation](https://github.com/taichi-dev/quantaichi) (experimental), etc. +- ... and many more features! A cross-platform, Vulkan-based 3D visualizer, [differentiable programming](https://docs.taichi-lang.org/lang/articles/differentiable_programming), [quantized computation](https://github.com/taichi-dev/quantaichi) (experimental), etc. ## Getting Started @@ -120,17 +120,17 @@ for i in range(1000000): -See [Get started](https://docs.taichi.graphics) for more information. +See [Get started](https://docs.taichi-lang.org) for more information. ### Build from source -If you wish to try our our experimental features or build Taichi for your own environments, see [Developer installation](https://docs.taichi.graphics/lang/articles/dev_install). +If you wish to try our our experimental features or build Taichi for your own environments, see [Developer installation](https://docs.taichi-lang.org/lang/articles/dev_install). ## Documentation -- [Technical documents](https://docs.taichi.graphics/) -- [API Reference](https://docs.taichi.graphics/api/) -- [Blog](https://docs.taichi.graphics/blog) +- [Technical documents](https://docs.taichi-lang.org/) +- [API Reference](https://docs.taichi-lang.org/api/) +- [Blog](https://docs.taichi-lang.org/blog) ## Contributing diff --git a/benchmarks/microbenchmarks/_items.py b/benchmarks/microbenchmarks/_items.py index 4c3ec40bba48f..c8b8651f322ea 100644 --- a/benchmarks/microbenchmarks/_items.py +++ b/benchmarks/microbenchmarks/_items.py @@ -67,7 +67,7 @@ def __init__(self): class MathOps(BenchmarkItem): name = 'math_op' - #reference: https://docs.taichi.graphics/lang/articles/operator + #reference: https://docs.taichi-lang.org/lang/articles/operator def __init__(self): self._items = { # Trigonometric diff --git a/docs/lang/articles/basic/overview.md b/docs/lang/articles/basic/overview.md index 1ecd751c7a23b..6944c360a3f90 100644 --- a/docs/lang/articles/basic/overview.md +++ b/docs/lang/articles/basic/overview.md @@ -17,7 +17,7 @@ To be fair, a domain-specific language (DSL) with a Python frontend is not somet * Taichi heavily optimizes the source code using various compiler technologies: common subexpression elimination, dead code elimination, control flow graph analysis, etc. These optimizations are backend neutral, because Taichi hosts its own intermediate representation (IR) layer. * JIT compilation provides additional optimization opportunities. -That said, Taichi goes beyond a Python JIT transpiler. One of the initial design goals is to *decouple the computation from the data structures*. The mechanism that Taichi provides is a set of generic data containers, called *SNode* (/ˈsnoʊd/). SNodes can be used to compose hierarchical, dense or sparse, multi-dimensional fields conveniently. Switching between array-of-structures and structure-of-arrays layouts is usually a matter of ≤10 lines of code. This has sparked many use cases in numerical simulation. If you are interested to learn them, please check out [Fields (advanced)](https://docs.taichi.graphics/lang/articles/layout), [Sparse spatial data structures](https://docs.taichi.graphics/lang/articles/sparse), or [the original Taichi paper](https://yuanming.taichi.graphics/publication/2019-taichi/taichi-lang.pdf). +That said, Taichi goes beyond a Python JIT transpiler. One of the initial design goals is to *decouple the computation from the data structures*. The mechanism that Taichi provides is a set of generic data containers, called *SNode* (/ˈsnoʊd/). SNodes can be used to compose hierarchical, dense or sparse, multi-dimensional fields conveniently. Switching between array-of-structures and structure-of-arrays layouts is usually a matter of ≤10 lines of code. This has sparked many use cases in numerical simulation. If you are interested to learn them, please check out [Fields (advanced)](https://docs.taichi-lang.org/lang/articles/layout), [Sparse spatial data structures](https://docs.taichi-lang.org/lang/articles/sparse), or [the original Taichi paper](https://yuanming.taichi.graphics/publication/2019-taichi/taichi-lang.pdf). The concept of decoupling is further extended to the type system. With GPU memory capacity and bandwidth becoming the major bottlenecks nowadays, it is vital to be able to pack more data per memory unit. Since 2021, Taichi has introduced customizable quantized types, allowing for the definition of fixed point or floating point numbers with arbitrary bits (still needs to be under 64). This has allowed an MPM simulation of over 400 million particles on a single GPU device. Learn more details in [the QuanTaichi paper](https://yuanming.taichi.graphics/publication/2021-quantaichi/quantaichi.pdf). diff --git a/docs/lang/articles/contribution/contributor_guide.md b/docs/lang/articles/contribution/contributor_guide.md index 933f826ced42c..04f16bbaa64f4 100644 --- a/docs/lang/articles/contribution/contributor_guide.md +++ b/docs/lang/articles/contribution/contributor_guide.md @@ -87,7 +87,7 @@ As part of the effort to increase visibility of the community and to improve dev - Documentation source files are hosted under [docs/](https://github.com/taichi-dev/taichi/blob/master/docs/). - We use GitHub Flavored Markdown (GFM) and [Docusaurus](https://docusaurus.io/) to build our documentation site. For information on the supported Markdown syntax, see the [Documentation Writing Guide](./doc_writing). - When it comes to writing, we adhere to the [Google Developer Documentation Style Guide](https://developers.google.com/style/). -- For instructions on setting up a local server and previewing your updated documentation in real-time, see the [Local Development](https://github.com/taichi-dev/docs.taichi.graphics#local-development). +- For instructions on setting up a local server and previewing your updated documentation in real-time, see the [Local Development](https://github.com/taichi-dev/docs.taichi-lang.org#local-development). ## Add test cases for your local changes diff --git a/docs/lang/articles/contribution/dev_install.md b/docs/lang/articles/contribution/dev_install.md index 81e8e8dfdea36..106a6d7fc395a 100644 --- a/docs/lang/articles/contribution/dev_install.md +++ b/docs/lang/articles/contribution/dev_install.md @@ -12,7 +12,7 @@ Developers who are interested in the compiler, computer graphics, or high-perfor This installation guide is *NOT* intended for end users who only wish to do simulation or high performance numerical computation. We recommend that end users install Taichi via `pip install taichi` and that there is no need for you to build Taichi from source. -See the [Get Started](https://docs.taichi.graphics/) for more information on quickly setting up Taichi for end users. +See the [Get Started](https://docs.taichi-lang.org/) for more information on quickly setting up Taichi for end users. ::: diff --git a/docs/lang/articles/reference.md b/docs/lang/articles/reference.md index 82ed4f4515d1f..41026974fc8c6 100644 --- a/docs/lang/articles/reference.md +++ b/docs/lang/articles/reference.md @@ -235,8 +235,8 @@ attributeref ::= primary "." identifier Attribute references are evaluated at compile time. The `primary` must be evaluated to a Python value with an attribute named `identifier`. Common use cases in Taichi include metadata queries of -[field](https://docs.taichi.graphics/lang/articles/meta#field-metadata) and -[matrices](https://docs.taichi.graphics/lang/articles/meta#matrix--vector-metadata). +[field](https://docs.taichi-lang.org/lang/articles/meta#field-metadata) and +[matrices](https://docs.taichi-lang.org/lang/articles/meta#matrix--vector-metadata). #### Subscriptions diff --git a/python/taichi/_lib/utils.py b/python/taichi/_lib/utils.py index 2917f2cf8a347..542dde54e6c9c 100644 --- a/python/taichi/_lib/utils.py +++ b/python/taichi/_lib/utils.py @@ -46,7 +46,7 @@ def import_ti_core(): if isinstance(e, ImportError): print(Fore.YELLOW + "Share object taichi_core import failed, " "check this page for possible solutions:\n" - "https://docs.taichi.graphics/lang/articles/install" + + "https://docs.taichi-lang.org/lang/articles/install" + Fore.RESET) if get_os_name() == 'win': # pylint: disable=E1101 diff --git a/python/taichi/_logging.py b/python/taichi/_logging.py index 96a857545034a..b58c1b5b76f66 100644 --- a/python/taichi/_logging.py +++ b/python/taichi/_logging.py @@ -36,7 +36,7 @@ def set_logging_level(level): also be effective. For example if `level` is set to 'warn', then the levels below it, which are 'error' and 'critical' in this case, will also be effective. - See also https://docs.taichi.graphics/lang/articles/utilities#logging. + See also https://docs.taichi-lang.org/lang/articles/utilities#logging. Args: level (str): Logging level. @@ -53,7 +53,7 @@ def is_logging_effective(level): All levels below current level will be effective. The default level is 'info'. - See also https://docs.taichi.graphics/lang/articles/utilities#logging. + See also https://docs.taichi-lang.org/lang/articles/utilities#logging. Args: level (str): The string represents logging level. \ diff --git a/python/taichi/_main.py b/python/taichi/_main.py index 51ee384da02e9..aa2f9b2be600d 100644 --- a/python/taichi/_main.py +++ b/python/taichi/_main.py @@ -91,7 +91,7 @@ def __call__(self): @staticmethod def _get_friend_links(): return '\n' \ - 'Docs: https://docs.taichi.graphics/\n' \ + 'Docs: https://docs.taichi-lang.org/\n' \ 'GitHub: https://github.com/taichi-dev/taichi/\n' \ 'Forum: https://forum.taichi.graphics/\n' diff --git a/python/taichi/lang/impl.py b/python/taichi/lang/impl.py index f091cc06e1ec0..d9e02ee2e6131 100644 --- a/python/taichi/lang/impl.py +++ b/python/taichi/lang/impl.py @@ -482,7 +482,7 @@ def __repr__(self): root = _Root() """Root of the declared Taichi :func:`~taichi.lang.impl.field`s. -See also https://docs.taichi.graphics/lang/articles/layout +See also https://docs.taichi-lang.org/lang/articles/layout Example:: @@ -524,7 +524,7 @@ def field(dtype, shape=None, name="", offset=None, needs_grad=False): actually defined. The data in a Taichi field can be directly accessed by a Taichi :func:`~taichi.lang.kernel_impl.kernel`. - See also https://docs.taichi.graphics/lang/articles/field + See also https://docs.taichi-lang.org/lang/articles/field Args: dtype (DataType): data type of the field. @@ -779,7 +779,7 @@ def static(x, *xs): `static()` is what enables the so-called metaprogramming in Taichi. It is in many ways similar to ``constexpr`` in C++. - See also https://docs.taichi.graphics/lang/articles/meta. + See also https://docs.taichi-lang.org/lang/articles/meta. Args: x (Any): an expression to be evaluated diff --git a/python/taichi/lang/kernel_impl.py b/python/taichi/lang/kernel_impl.py index 3137fcdf39467..58d02087c064b 100644 --- a/python/taichi/lang/kernel_impl.py +++ b/python/taichi/lang/kernel_impl.py @@ -836,7 +836,7 @@ def kernel(fn): Kernel's gradient kernel would be generated automatically by the AutoDiff system. - See also https://docs.taichi.graphics/lang/articles/syntax#kernel. + See also https://docs.taichi-lang.org/lang/articles/syntax#kernel. Args: fn (Callable): the Python function to be decorated @@ -885,7 +885,7 @@ def data_oriented(cls): To allow for modularized code, Taichi provides this decorator so that Taichi kernels can be defined inside a class. - See also https://docs.taichi.graphics/lang/articles/odop + See also https://docs.taichi-lang.org/lang/articles/odop Example:: diff --git a/python/taichi/lang/matrix.py b/python/taichi/lang/matrix.py index bff32dc969255..fd17ac9db1c67 100644 --- a/python/taichi/lang/matrix.py +++ b/python/taichi/lang/matrix.py @@ -235,7 +235,7 @@ def __init__(self, arr, dt=None, suppress_warning=False, is_ref=False): ' So the compilation time could be extremely long if the matrix size is too big.' ' You may use a field to store a large matrix like this, e.g.:\n' f' x = ti.field(ti.f32, ({self.n}, {self.m})).\n' - ' See https://docs.taichi.graphics/lang/articles/field#matrix-size' + ' See https://docs.taichi-lang.org/lang/articles/field#matrix-size' ' for more details.', UserWarning, stacklevel=2) @@ -321,7 +321,7 @@ def _linearize_entry_id(self, *args): 'If you want to *iterate through matrix elements*, use a static range:\n' ' for i in ti.static(range(3)):\n' ' print(i, "-th component is", vec[i])\n' - 'See https://docs.taichi.graphics/lang/articles/meta#when-to-use-tistatic-with-for-loops for more details.' + 'See https://docs.taichi-lang.org/lang/articles/meta#when-to-use-tistatic-with-for-loops for more details.' 'Or turn on ti.init(..., dynamic_index=True) to support indexing with variables!' ) assert 0 <= args[0] < self.n, \ diff --git a/python/taichi/lang/misc.py b/python/taichi/lang/misc.py index 13438ddfd31b8..0bfebb2da5973 100644 --- a/python/taichi/lang/misc.py +++ b/python/taichi/lang/misc.py @@ -341,7 +341,7 @@ def init(arch=None, * ``cpu_max_num_threads`` (int): Sets the number of threads used by the CPU thread pool. * ``debug`` (bool): Enables the debug mode, under which Taichi does a few more things like boundary checks. * ``print_ir`` (bool): Prints the CHI IR of the Taichi kernels. - * ``packed`` (bool): Enables the packed memory layout. See https://docs.taichi.graphics/lang/articles/layout. + * ``packed`` (bool): Enables the packed memory layout. See https://docs.taichi-lang.org/lang/articles/layout. """ # Check version for users every 7 days if not disabled by users. _version_check.start_version_check_thread() @@ -476,7 +476,7 @@ def no_activate(*args): def block_local(*args): """Hints Taichi to cache the fields and to enable the BLS optimization. - Please visit https://docs.taichi.graphics/lang/articles/performance + Please visit https://docs.taichi-lang.org/lang/articles/performance for how BLS is used. Args: diff --git a/python/taichi/lang/snode.py b/python/taichi/lang/snode.py index 82a997fdd4a09..3b0c7460f3958 100644 --- a/python/taichi/lang/snode.py +++ b/python/taichi/lang/snode.py @@ -11,7 +11,7 @@ class SNode: For more information on Taichi's SNode system, please check out these references: - * https://docs.taichi.graphics/lang/articles/sparse + * https://docs.taichi-lang.org/lang/articles/sparse * https://yuanming.taichi.graphics/publication/2019-taichi/taichi-lang.pdf Arg: diff --git a/python/taichi/profiler/kernel_metrics.py b/python/taichi/profiler/kernel_metrics.py index a4ff09bbe6c41..2e5bbd7036959 100644 --- a/python/taichi/profiler/kernel_metrics.py +++ b/python/taichi/profiler/kernel_metrics.py @@ -44,7 +44,7 @@ class CuptiMetric: >>> ti.profiler.print_kernel_profiler_info('trace') Note: - For details about using CUPTI in Taichi, please visit https://docs.taichi.graphics/lang/articles/profiler#advanced-mode. + For details about using CUPTI in Taichi, please visit https://docs.taichi-lang.org/lang/articles/profiler#advanced-mode. """ def __init__(self, name='', diff --git a/python/taichi/profiler/kernel_profiler.py b/python/taichi/profiler/kernel_profiler.py index d176304537067..5dea618b1a286 100644 --- a/python/taichi/profiler/kernel_profiler.py +++ b/python/taichi/profiler/kernel_profiler.py @@ -46,7 +46,7 @@ class KernelProfiler: This mode is only available for the CUDA backend with CUPTI toolkit, i.e. you need ``ti.init(kernel_profiler=True, arch=ti.cuda)``. Note: - For details about using CUPTI in Taichi, please visit https://docs.taichi.graphics/lang/articles/profiler#advanced-mode. + For details about using CUPTI in Taichi, please visit https://docs.taichi-lang.org/lang/articles/profiler#advanced-mode. """ def __init__(self): self._profiling_mode = False @@ -388,7 +388,7 @@ def print_kernel_profiler_info(mode='count'): Currently the result of `KernelProfiler` could be incorrect on OpenGL backend due to its lack of support for `ti.sync()`. - For advanced mode of `KernelProfiler`, please visit https://docs.taichi.graphics/lang/articles/profiler#advanced-mode. + For advanced mode of `KernelProfiler`, please visit https://docs.taichi-lang.org/lang/articles/profiler#advanced-mode. """ get_default_kernel_profiler().print_info(mode) diff --git a/python/taichi/types/annotations.py b/python/taichi/types/annotations.py index 56a27d8e8d969..4476fe628d5f0 100644 --- a/python/taichi/types/annotations.py +++ b/python/taichi/types/annotations.py @@ -2,7 +2,7 @@ class Template: """Type annotation for template kernel parameter. Useful for passing parameters to kernels by reference. - See also https://docs.taichi.graphics/lang/articles/meta. + See also https://docs.taichi-lang.org/lang/articles/meta. Args: tensor (Any): unused diff --git a/taichi/analysis/verify.cpp b/taichi/analysis/verify.cpp index 66e5e7b9aa5e1..cc9df7fde4860 100644 --- a/taichi/analysis/verify.cpp +++ b/taichi/analysis/verify.cpp @@ -49,7 +49,7 @@ class IRVerifier : public BasicStmtVisitor { found, "IR broken: stmt {} {} cannot have operand {} {}." " If you are using autodiff, please check" - " https://docs.taichi.graphics/lang/articles/" + " https://docs.taichi-lang.org/lang/articles/" "differences_between_taichi_and_python_programs" " If it doesn't help, please report this bug by opening an issue at" " https://github.com/taichi-dev/taichi to help us improve." diff --git a/taichi/backends/cuda/cupti_toolkit.cpp b/taichi/backends/cuda/cupti_toolkit.cpp index 64cd4f21ae5c8..02698e9857e70 100644 --- a/taichi/backends/cuda/cupti_toolkit.cpp +++ b/taichi/backends/cuda/cupti_toolkit.cpp @@ -39,7 +39,7 @@ bool check_cupti_availability() { "7.0 , fallback to default kernel profiler"); TI_WARN( "See also: " - "https://docs.taichi.graphics/lang/articles/profiler"); + "https://docs.taichi-lang.org/lang/articles/profiler"); return false; } return true; @@ -106,7 +106,7 @@ bool check_cupti_privileges() { "================================================================="); TI_WARN( "See also: " - "https://docs.taichi.graphics/lang/articles/profiler"); + "https://docs.taichi-lang.org/lang/articles/profiler"); return false; } // For other errors , CuptiToolkit::init_cupti() will send error message. diff --git a/taichi/gui/x11.cpp b/taichi/gui/x11.cpp index fff0c9e6f9b86..da82e5b08fcf5 100644 --- a/taichi/gui/x11.cpp +++ b/taichi/gui/x11.cpp @@ -151,7 +151,7 @@ void GUI::create_window() { "Taichi fails to create a window." " This is probably due to the lack of an X11 GUI environment." " Consider using the `ti.GUI(show_gui=False)` option, see" - " https://docs.taichi.graphics/lang/articles/gui_system"); + " https://docs.taichi-lang.org/lang/articles/gui_system"); visual = DefaultVisual(display, 0); window = XCreateSimpleWindow((Display *)display, RootWindow((Display *)display, 0), diff --git a/taichi/ir/snode.h b/taichi/ir/snode.h index ff31f11d385c0..b6698574c8c32 100644 --- a/taichi/ir/snode.h +++ b/taichi/ir/snode.h @@ -119,7 +119,7 @@ class SNode { std::string name; // Product of the |shape| of all the activated axes identified by // |extractors|. - // See https://docs.taichi.graphics/lang/articles/internal for terms + // See https://docs.taichi-lang.org/lang/articles/internal for terms // like cell and container. int64 num_cells_per_container{1}; int total_num_bits{0}; diff --git a/taichi/system/traceback.cpp b/taichi/system/traceback.cpp index 4ba1c365e2034..5ba9ae8a44975 100644 --- a/taichi/system/traceback.cpp +++ b/taichi/system/traceback.cpp @@ -370,7 +370,7 @@ void print_traceback() { fmt::print( fg(fmt::color::orange), "\nInternal error occurred. Check out this page for possible solutions:\n" - "https://docs.taichi.graphics/lang/articles/install\n"); + "https://docs.taichi-lang.org/lang/articles/install\n"); } TI_NAMESPACE_END diff --git a/taichi/transforms/reverse_segments.cpp b/taichi/transforms/reverse_segments.cpp index 7dd3e4ae3bf71..c6148746a7c47 100644 --- a/taichi/transforms/reverse_segments.cpp +++ b/taichi/transforms/reverse_segments.cpp @@ -70,7 +70,7 @@ void reverse_segments(IRNode *root) { "Mixed usage of for-loops and statements without looping. \n" "Please split them into two kernels " "and check the documentation for more details:\n" - "https://docs.taichi.graphics/lang/articles/" + "https://docs.taichi-lang.org/lang/articles/" "differentiable_programming"); } for (auto &sblock : statement_blocks) { diff --git a/tests/python/test_fields_builder.py b/tests/python/test_fields_builder.py index d02fb249531ea..0e95bdb93d8e1 100644 --- a/tests/python/test_fields_builder.py +++ b/tests/python/test_fields_builder.py @@ -139,7 +139,7 @@ def assign_field_multiple_struct_for(): # We currently only consider data types that all platforms support. -# See https://docs.taichi.graphics/lang/articles/type#primitive-types for more details. +# See https://docs.taichi-lang.org/lang/articles/type#primitive-types for more details. @pytest.mark.parametrize('test_1d_size', [1, 10, 100]) @pytest.mark.parametrize('field_type', [ti.f32, ti.i32]) @test_utils.test(arch=[ti.cpu, ti.cuda, ti.vulkan, ti.metal]) diff --git a/tests/python/test_simple_matrix_slice.py b/tests/python/test_simple_matrix_slice.py index 6334ee6769880..68f1461b6b1de 100644 --- a/tests/python/test_simple_matrix_slice.py +++ b/tests/python/test_simple_matrix_slice.py @@ -62,7 +62,7 @@ def test_one_col_slice() -> ti.types.matrix(1, 3, dtype=ti.i32): r'If you want to \*iterate through matrix elements\*, use a static range:\n' r' for i in ti.static\(range\(3\)\):\n' r' print\(i, "-th component is", vec\[i\]\)\n' - r'See https://docs.taichi.graphics/lang/articles/meta#when-to-use-tistatic-with-for-loops for more details.' + r'See https://docs.taichi-lang.org/lang/articles/meta#when-to-use-tistatic-with-for-loops for more details.' r'Or turn on ti.init\(..., dynamic_index=True\) to support indexing with variables!' ): test_one_col_slice() From baedd61f833e57b8d883436fd3814b3fc21f30c9 Mon Sep 17 00:00:00 2001 From: Yi Xu Date: Tue, 26 Apr 2022 16:27:33 +0800 Subject: [PATCH 03/12] [Bug] [ir] Fix and refactor type check for atomic ops (#4858) * [Bug] [ir] Fix and refactor type check for atomic ops * [pre-commit.ci] auto fixes from pre-commit.com hooks for more information, see https://pre-commit.ci * Remove accidental change Co-authored-by: pre-commit-ci[bot] <66853113+pre-commit-ci[bot]@users.noreply.github.com> --- taichi/ir/frontend_ir.cpp | 1 + taichi/transforms/type_check.cpp | 69 +++++++++++++------------------- 2 files changed, 29 insertions(+), 41 deletions(-) diff --git a/taichi/ir/frontend_ir.cpp b/taichi/ir/frontend_ir.cpp index 757b4b33e949d..bedd7f5fb5fd5 100644 --- a/taichi/ir/frontend_ir.cpp +++ b/taichi/ir/frontend_ir.cpp @@ -493,6 +493,7 @@ void AtomicOpExpression::flatten(FlattenContext *ctx) { ctx->push_back(op_type, dest->stmt, expr->stmt); } stmt = ctx->back_stmt(); + stmt->tb = tb; } void SNodeOpExpression::type_check(CompileConfig *) { diff --git a/taichi/transforms/type_check.cpp b/taichi/transforms/type_check.cpp index 0048cf2706a89..cf1d0b3080702 100644 --- a/taichi/transforms/type_check.cpp +++ b/taichi/transforms/type_check.cpp @@ -18,6 +18,28 @@ class TypeCheck : public IRVisitor { private: CompileConfig config_; + Type *type_check_store(Stmt *stmt, + Stmt *dst, + Stmt *&val, + const std::string &stmt_name) { + auto dst_type = dst->ret_type.ptr_removed(); + if (dst_type->is() || dst_type->is()) { + // We force the value type to be the compute_type of the bit pointer. + // Casting from compute_type to physical_type is handled in codegen. + dst_type = dst_type->get_compute_type(); + } + if (dst_type != val->ret_type) { + auto promoted = promoted_type(dst_type, val->ret_type); + if (dst_type != promoted) { + TI_WARN("[{}] {} may lose precision: {} <- {}\n{}", stmt->name(), + stmt_name, dst_type->to_string(), val->ret_data_type_name(), + stmt->tb); + } + val = insert_type_cast_before(stmt, val, dst_type); + } + return dst_type; + } + public: explicit TypeCheck(const CompileConfig &config) : config_(config) { allow_undefined_visitor = true; @@ -57,20 +79,9 @@ class TypeCheck : public IRVisitor { void visit(AtomicOpStmt *stmt) override { TI_ASSERT(stmt->width() == 1); // TODO(type): test_ad_for fails if we assume dest is a pointer type. - auto dst_type = stmt->dest->ret_type.ptr_removed(); - if (auto cit = dst_type->cast()) { - dst_type = cit->get_physical_type(); - } else if (auto cft = dst_type->cast()) { - auto cit = cft->get_digits_type()->as(); - dst_type = cit->get_physical_type(); - } else if (stmt->val->ret_type != dst_type) { - TI_WARN("[{}] Atomic {} ({} to {}) may lose precision\n{}", stmt->name(), - atomic_op_type_name(stmt->op_type), - data_type_name(stmt->val->ret_type), data_type_name(dst_type), - stmt->tb); - stmt->val = insert_type_cast_before(stmt, stmt->val, dst_type); - } - stmt->ret_type = dst_type; + stmt->ret_type = type_check_store( + stmt, stmt->dest, stmt->val, + fmt::format("Atomic {}", atomic_op_type_name(stmt->op_type))); } void visit(LocalLoadStmt *stmt) override { @@ -106,17 +117,8 @@ class TypeCheck : public IRVisitor { // Infer data type for alloca stmt->dest->ret_type = stmt->val->ret_type; } - auto dst_value_type = stmt->dest->ret_type.ptr_removed(); - if (dst_value_type != stmt->val->ret_type) { - auto promoted = promoted_type(dst_value_type, stmt->val->ret_type); - if (dst_value_type != promoted) { - TI_WARN("[{}] Local store may lose precision {} <- {}\n{}", - stmt->name(), dst_value_type->to_string(), - stmt->val->ret_data_type_name(), stmt->tb); - } - stmt->val = insert_type_cast_before(stmt, stmt->val, dst_value_type); - } - stmt->ret_type = dst_value_type; + stmt->ret_type = + type_check_store(stmt, stmt->dest, stmt->val, "Local store"); } void visit(GlobalLoadStmt *stmt) override { @@ -180,22 +182,7 @@ class TypeCheck : public IRVisitor { } void visit(GlobalStoreStmt *stmt) override { - auto dst_value_type = stmt->dest->ret_type.ptr_removed(); - if (dst_value_type->is() || - dst_value_type->is()) { - // We force the value type to be the compute_type of the bit pointer. - // Casting from compute_type to physical_type is handled in codegen. - dst_value_type = dst_value_type->get_compute_type(); - } - if (dst_value_type != stmt->val->ret_type) { - auto promoted = promoted_type(dst_value_type, stmt->val->ret_type); - if (dst_value_type != promoted) { - TI_WARN("[{}] Global store may lose precision: {} <- {}\n{}", - stmt->name(), dst_value_type->to_string(), - stmt->val->ret_data_type_name(), stmt->tb); - } - stmt->val = insert_type_cast_before(stmt, stmt->val, dst_value_type); - } + type_check_store(stmt, stmt->dest, stmt->val, "Global store"); } void visit(RangeForStmt *stmt) override { From 433b0e35fc3486e0cac3d6b80ba65d5645656c5a Mon Sep 17 00:00:00 2001 From: yekuang Date: Tue, 26 Apr 2022 16:52:32 +0800 Subject: [PATCH 04/12] [lang] Add better error detection for swizzle patterens (#4860) * [lang] Add better error detection for swizzle patterens * [pre-commit.ci] auto fixes from pre-commit.com hooks for more information, see https://pre-commit.ci * fix Co-authored-by: pre-commit-ci[bot] <66853113+pre-commit-ci[bot]@users.noreply.github.com> --- python/taichi/lang/matrix.py | 58 +++++++++++++++++++---------- tests/python/test_vector_swizzle.py | 53 ++++++++++++++++++++++++++ 2 files changed, 92 insertions(+), 19 deletions(-) diff --git a/python/taichi/lang/matrix.py b/python/taichi/lang/matrix.py index fd17ac9db1c67..5f1f22db25489 100644 --- a/python/taichi/lang/matrix.py +++ b/python/taichi/lang/matrix.py @@ -1,4 +1,3 @@ -import functools import numbers from collections.abc import Iterable @@ -24,34 +23,54 @@ def _gen_swizzles(cls): swizzle_gen = SwizzleGenerator() # https://www.khronos.org/opengl/wiki/Data_Type_(GLSL)#Swizzling - KEYMAP_SET = ['xyzw', 'rgba', 'stpq'] + KEYGROUP_SET = ['xyzw', 'rgba', 'stpq'] + + def make_valid_attribs_checker(key_group): + def check(instance, pattern): + valid_attribs = set(key_group[:instance.n]) + pattern_set = set(pattern) + diff = pattern_set - valid_attribs + if len(diff): + valid_attribs = tuple(sorted(valid_attribs)) + pattern = tuple(pattern) + raise TaichiSyntaxError( + f'vec{instance.n} only has ' + f'attributes={valid_attribs}, got={pattern}') - def add_single_swizzle_attrs(cls): - """Add property getter and setter for a single character in "xyzwrgbastpq". - """ - def prop_getter(index, instance): - return instance(index) + return check + + for key_group in KEYGROUP_SET: + for index, attr in enumerate(key_group): - @python_scope - def prop_setter(index, instance, value): - instance[index] = value + def gen_property(attr, attr_idx, key_group): + checker = make_valid_attribs_checker(key_group) - for key_group in KEYMAP_SET: - for index, key in enumerate(key_group): - prop = property(functools.partial(prop_getter, index), - functools.partial(prop_setter, index)) - setattr(cls, key, prop) + def prop_getter(instance): + checker(instance, attr) + return instance._get_entry_and_read([attr_idx]) + + @python_scope + def prop_setter(instance, value): + checker(instance, attr) + instance[attr_idx] = value - add_single_swizzle_attrs(cls) + return property(prop_getter, prop_setter) - for key_group in KEYMAP_SET: + prop = gen_property(attr, index, key_group) + setattr(cls, attr, prop) + + for key_group in KEYGROUP_SET: sw_patterns = swizzle_gen.generate(key_group, required_length=4) # len=1 accessors are handled specially above sw_patterns = filter(lambda p: len(p) > 1, sw_patterns) for pat in sw_patterns: # Create a function for value capturing def gen_property(pattern, key_group): + checker = make_valid_attribs_checker(key_group) + prop_key = ''.join(pattern) + def prop_getter(instance): + checker(instance, pattern) res = [] for ch in pattern: res.append(instance._get_entry(key_group.index(ch))) @@ -60,7 +79,9 @@ def prop_getter(instance): def prop_setter(instance, value): if len(pattern) != len(value): raise TaichiCompilationError( - 'values does not match the attribute') + f'value len does not match the swizzle pattern={prop_key}' + ) + checker(instance, pattern) for ch, val in zip(pattern, value): if in_python_scope(): instance[key_group.index(ch)] = val @@ -68,7 +89,6 @@ def prop_setter(instance, value): instance(key_group.index(ch))._assign(val) prop = property(prop_getter, prop_setter) - prop_key = ''.join(pattern) return prop_key, prop prop_key, prop = gen_property(pat, key_group) diff --git a/tests/python/test_vector_swizzle.py b/tests/python/test_vector_swizzle.py index 83883f3566b5d..588b3f1372460 100644 --- a/tests/python/test_vector_swizzle.py +++ b/tests/python/test_vector_swizzle.py @@ -1,3 +1,5 @@ +import re + import pytest import taichi as ti @@ -48,6 +50,38 @@ def foo(): foo() +@test_utils.test(debug=True) +def test_vector_swizzle2_taichi(): + @ti.kernel + def foo(): + v = ti.math.vec3(0, 0, 0) + v.brg += 1 + assert all(v.xyz == (1, 1, 1)) + v.x = 1 + v.g = 2 + v.p = 3 + v123 = ti.math.vec3(1, 2, 3) + v231 = ti.math.vec3(2, 3, 1) + v113 = ti.math.vec3(1, 1, 3) + assert all(v == v123) + assert all(v.xyz == v123) + assert all(v.rgb == v123) + assert all(v.stp == v123) + assert all(v.yzx == v231) + assert all(v.gbr == v231) + assert all(v.tps == v231) + assert all(v.xxz == v113) + assert all(v.rrb == v113) + assert all(v.ssp == v113) + v.bgr = v123 + v321 = ti.math.vec3(3, 2, 1) + assert all(v.xyz == v321) + assert all(v.rgb == v321) + assert all(v.stp == v321) + + foo() + + @test_utils.test(debug=True) def test_vector_dtype(): @ti.kernel @@ -59,3 +93,22 @@ def foo(): assert all(b == (1, 2, 3)) foo() + + +@test_utils.test() +def test_vector_invalid_swizzle_patterns(): + a = ti.math.vec2(1, 2) + with pytest.raises(ti.TaichiSyntaxError, + match=re.escape( + "vec2 only has attributes=('x', 'y'), got=('z',)")): + a.z = 3 + with pytest.raises( + ti.TaichiSyntaxError, + match=re.escape( + "vec2 only has attributes=('x', 'y'), got=('x', 'y', 'z')")): + a.xyz = [1, 2, 3] + + with pytest.raises(ti.TaichiCompilationError, + match=re.escape( + "value len does not match the swizzle pattern=xy")): + a.xy = [1, 2, 3] From 1c3619d9a115e760da841e5148e2b9f683bc9b68 Mon Sep 17 00:00:00 2001 From: Yi Xu Date: Tue, 26 Apr 2022 16:52:48 +0800 Subject: [PATCH 05/12] [gui] Avoid implicit type casts in staging_buffer (#4861) --- python/taichi/ui/staging_buffer.py | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/python/taichi/ui/staging_buffer.py b/python/taichi/ui/staging_buffer.py index f7377c9843002..7056ed9007cae 100644 --- a/python/taichi/ui/staging_buffer.py +++ b/python/taichi/ui/staging_buffer.py @@ -89,10 +89,10 @@ def copy_image_f32_to_u8(src: ti.template(), dst: ti.template(), c = src[i, j][k] c = max(0.0, min(1.0, c)) c = c * 255 - dst[i, j][k] = int(c) + dst[i, j][k] = ti.cast(c, u8) if num_components < 4: # alpha channel - dst[i, j][3] = 255 + dst[i, j][3] = u8(255) @ti.kernel @@ -103,7 +103,7 @@ def copy_image_u8_to_u8(src: ti.template(), dst: ti.template(), dst[i, j][k] = src[i, j][k] if num_components < 4: # alpha channel - dst[i, j][3] = 255 + dst[i, j][3] = u8(255) # ggui renderer always assumes the input image to be u8 RGBA From e346e9f87b1fcf1151a7135248fd1981d3f477f8 Mon Sep 17 00:00:00 2001 From: Frost Ming Date: Wed, 27 Apr 2022 09:07:11 +0800 Subject: [PATCH 06/12] [Build] Switch to scikit-build as the build backend (#4624) * switch to skbuild * Switch the build system to scikit-build * include bc and libmolten * find llvm runtime bc * fix bc files installation * install bc after compile * Add more message * Auto Format * fix findpython * Kickstart CI * add empty line * add missing dependency * fix python args * start CI * Fix clang tidy run * [pre-commit.ci] auto fixes from pre-commit.com hooks for more information, see https://pre-commit.ci Co-authored-by: Taichi Gardener Co-authored-by: Ailing Co-authored-by: pre-commit-ci[bot] <66853113+pre-commit-ci[bot]@users.noreply.github.com> --- .github/workflows/scripts/check_clang_tidy.sh | 7 +- .github/workflows/scripts/win_build.ps1 | 1 - .gitignore | 1 + CMakeLists.txt | 2 +- MANIFEST.in | 5 - cmake/PythonNumpyPybind11.cmake | 99 +------- cmake/TaichiCore.cmake | 15 +- pyproject.toml | 6 +- requirements_dev.txt | 3 + scripts/run_clang_tidy.py | 15 +- setup.py | 231 ++++-------------- 11 files changed, 103 insertions(+), 282 deletions(-) diff --git a/.github/workflows/scripts/check_clang_tidy.sh b/.github/workflows/scripts/check_clang_tidy.sh index d9db1c9a3433f..4155421d86716 100755 --- a/.github/workflows/scripts/check_clang_tidy.sh +++ b/.github/workflows/scripts/check_clang_tidy.sh @@ -5,8 +5,5 @@ CI_SETUP_CMAKE_ARGS=$1 cd taichi python3 -m pip install -r requirements_dev.txt -rm -rf build && mkdir build && cd build -cmake $CI_SETUP_CMAKE_ARGS .. - -cd .. -python3 ./scripts/run_clang_tidy.py $PWD/taichi -clang-tidy-binary clang-tidy-10 -checks=-*,performance-inefficient-string-concatenation,readability-identifier-naming -header-filter=$PWD/taichi -p $PWD/build -j2 +export CI_SETUP_CMAKE_ARGS +python3 ./scripts/run_clang_tidy.py $PWD/taichi -clang-tidy-binary clang-tidy-10 -checks=-*,performance-inefficient-string-concatenation,readability-identifier-naming -header-filter=$PWD/taichi -j2 diff --git a/.github/workflows/scripts/win_build.ps1 b/.github/workflows/scripts/win_build.ps1 index e58c179ee6952..86ad4243742e5 100644 --- a/.github/workflows/scripts/win_build.ps1 +++ b/.github/workflows/scripts/win_build.ps1 @@ -76,7 +76,6 @@ python -m venv venv . venv\Scripts\activate.ps1 python -m pip install wheel python -m pip install -r requirements_dev.txt -python -m pip install -r requirements_test.txt if (-not $?) { exit 1 } WriteInfo("Building Taichi") $env:TAICHI_CMAKE_ARGS += " -DCLANG_EXECUTABLE=$libsDir\\taichi_clang\\bin\\clang++.exe" diff --git a/.gitignore b/.gitignore index fd39d08f9acea..958272e9a7b4d 100644 --- a/.gitignore +++ b/.gitignore @@ -85,3 +85,4 @@ _build !docs/**/*.json imgui.ini /venv/ +/_skbuild/ diff --git a/CMakeLists.txt b/CMakeLists.txt index 17db545ac8194..7387e8b648b56 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -49,7 +49,6 @@ else () endif () set(CMAKE_RUNTIME_OUTPUT_DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/build") -set(CMAKE_LIBRARY_OUTPUT_DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/build") find_program(CCACHE_PROGRAM ccache) if(CCACHE_PROGRAM) @@ -157,6 +156,7 @@ foreach(arch IN LISTS HOST_ARCH CUDA_ARCH) WORKING_DIRECTORY "${PROJECT_SOURCE_DIR}/taichi/runtime/llvm" ) add_dependencies(${CORE_LIBRARY_NAME} "generate_llvm_runtime_${arch}") + install(FILES "${PROJECT_SOURCE_DIR}/taichi/runtime/llvm/runtime_${arch}.bc" DESTINATION ${CMAKE_INSTALL_PREFIX}/python/taichi/_lib/runtime) endforeach() configure_file(taichi/common/version.h.in ${CMAKE_SOURCE_DIR}/taichi/common/version.h) diff --git a/MANIFEST.in b/MANIFEST.in index eff0f4f8f88c7..3c1e64aa11c27 100644 --- a/MANIFEST.in +++ b/MANIFEST.in @@ -1,8 +1,3 @@ -include MANIFEST.in -include version.txt -include python/*.txt -include python/*.py -include *.cfg include python/taichi/*.md recursive-include python/taichi/assets * recursive-include python/taichi/examples *.py diff --git a/cmake/PythonNumpyPybind11.cmake b/cmake/PythonNumpyPybind11.cmake index 311630dba74a8..65a231e04f64b 100644 --- a/cmake/PythonNumpyPybind11.cmake +++ b/cmake/PythonNumpyPybind11.cmake @@ -1,93 +1,16 @@ # Python, numpy, and pybind11 +execute_process(COMMAND ${PYTHON_EXECUTABLE} -m pybind11 --cmake + OUTPUT_VARIABLE pybind11_DIR OUTPUT_STRIP_TRAILING_WHITESPACE) +execute_process(COMMAND ${PYTHON_EXECUTABLE} -c "import numpy;print(numpy.get_include())" + OUTPUT_VARIABLE NUMPY_INCLUDE_DIR OUTPUT_STRIP_TRAILING_WHITESPACE) -if (PYTHON_EXECUTABLE) - message("Using ${PYTHON_EXECUTABLE} as python executable.") -else () - if (WIN32) - message("Using 'python' as python interpreter.") - set(PYTHON_EXECUTABLE python) - else () - message("Using 'python3' as python interpreter.") - set(PYTHON_EXECUTABLE python3) - endif() -endif () +message("-- Python: Using ${PYTHON_EXECUTABLE} as the interpreter") +message(" version: ${PYTHON_VERSION_STRING}") +message(" include: ${PYTHON_INCLUDE_DIR}") +message(" library: ${PYTHON_LIBRARY}") +message(" numpy include: ${NUMPY_INCLUDE_DIR}") -if (WIN32) - execute_process(COMMAND where ${PYTHON_EXECUTABLE} - OUTPUT_VARIABLE PYTHON_EXECUTABLE_PATHS) - if (${PYTHON_EXECUTABLE_PATHS}) - string(FIND ${PYTHON_EXECUTABLE_PATHS} "\n" _LINE_BREAK_LOC) - string(SUBSTRING ${PYTHON_EXECUTABLE_PATHS} 0 ${_LINE_BREAK_LOC} PYTHON_EXECUTABLE_PATH) - else () - set(PYTHON_EXECUTABLE_PATH ${PYTHON_EXECUTABLE}) - endif () -else () - execute_process(COMMAND which ${PYTHON_EXECUTABLE} - OUTPUT_VARIABLE PYTHON_EXECUTABLE_PATH) -endif() -execute_process(COMMAND ${PYTHON_EXECUTABLE} -c - "import sys;\ - from distutils import sysconfig;\ - sys.stdout.write(sysconfig.get_python_version())" - OUTPUT_VARIABLE PYTHON_VERSION) -execute_process(COMMAND ${PYTHON_EXECUTABLE} --version) -execute_process(COMMAND ${PYTHON_EXECUTABLE} -c - "import sys;\ - from distutils import sysconfig;\ - sys.stdout.write(\ - (sysconfig.get_config_var('INCLUDEPY')\ - if sysconfig.get_config_var('INCLUDEDIR') is not None else None)\ - or sysconfig.get_python_inc())" - OUTPUT_VARIABLE PYTHON_INCLUDE_DIRS) -execute_process(COMMAND ${PYTHON_EXECUTABLE} -c - "import sys;\ - from distutils import sysconfig;\ - sys.stdout.write((sysconfig.get_config_var('LIBDIR') or sysconfig.get_python_lib()).replace('\\\\','/'))" - OUTPUT_VARIABLE PYTHON_LIBRARY_DIR) +include_directories(${NUMPY_INCLUDE_DIR}) -execute_process(COMMAND ${PYTHON_EXECUTABLE} -c - "import sys;\ - sys.stdout.write(str(sys.version_info[1]))" - OUTPUT_VARIABLE PYTHON_MINOR_VERSION) - - -if (WIN32) - execute_process(COMMAND ${PYTHON_EXECUTABLE} -c - "import sys;sys.stdout.write(sys.base_prefix.replace('\\\\', '/'))" - OUTPUT_VARIABLE PYTHON_BASE_PREFIX) - link_directories(${PYTHON_BASE_PREFIX}/libs) - set(PYTHON_LIBRARIES ${PYTHON_BASE_PREFIX}/libs/python3.lib) - set(PYTHON_LIBRARIES ${PYTHON_BASE_PREFIX}/libs/python3${PYTHON_MINOR_VERSION}.lib) -else() - find_library(PYTHON_LIBRARY NAMES python${PYTHON_VERSION} python${PYTHON_VERSION}m PATHS ${PYTHON_LIBRARY_DIR} - NO_DEFAULT_PATH NO_SYSTEM_ENVIRONMENT_PATH PATH_SUFFIXES x86_64-linux-gnu) - set(PYTHON_LIBRARIES ${PYTHON_LIBRARY}) -endif() - - -include_directories(${PYTHON_INCLUDE_DIRS}) -message(" version: ${PYTHON_VERSION}") -message(" include: ${PYTHON_INCLUDE_DIRS}") -message(" library: ${PYTHON_LIBRARIES}") - -execute_process(COMMAND ${PYTHON_EXECUTABLE} -c - "import numpy.distutils, sys;\ - sys.stdout.write(':'.join(numpy.distutils.misc_util.get_numpy_include_dirs()))" - OUTPUT_VARIABLE PYTHON_NUMPY_INCLUDE_DIR) - -message(" numpy include: ${PYTHON_NUMPY_INCLUDE_DIR}") -include_directories(${PYTHON_NUMPY_INCLUDE_DIR}) - -execute_process(COMMAND ${PYTHON_EXECUTABLE} -c - "import sys; import pybind11; sys.stdout.write(pybind11.get_include() + ';' + pybind11.get_include(True))" - OUTPUT_VARIABLE PYBIND11_INCLUDE_DIR - RESULT_VARIABLE PYBIND11_IMPORT_RET) -if (NOT PYBIND11_IMPORT_RET) - # returns zero if success - message(" pybind11 include: ${PYBIND11_INCLUDE_DIR}") -else () - message(FATAL_ERROR "Cannot import pybind11. Please install. ([sudo] pip3 install --user pybind11)") -endif () - -include_directories(${PYBIND11_INCLUDE_DIR}) +find_package(pybind11 CONFIG REQUIRED) diff --git a/cmake/TaichiCore.cmake b/cmake/TaichiCore.cmake index 9bc218d9cc9ea..687130d9e19d5 100644 --- a/cmake/TaichiCore.cmake +++ b/cmake/TaichiCore.cmake @@ -16,6 +16,7 @@ option(TI_EMSCRIPTENED "Build using emscripten" OFF) # projects. set(CMAKE_CXX_VISIBILITY_PRESET hidden) set(CMAKE_VISIBILITY_INLINES_HIDDEN ON) +set(INSTALL_LIB_DIR ${CMAKE_INSTALL_PREFIX}/python/taichi/_lib) if(ANDROID) set(TI_WITH_VULKAN ON) @@ -384,6 +385,9 @@ if (TI_WITH_VULKAN) find_library(MOLTEN_VK libMoltenVK.dylib PATHS $HOMEBREW_CELLAR/molten-vk $VULKAN_SDK REQUIRED) configure_file(${MOLTEN_VK} ${CMAKE_BINARY_DIR}/libMoltenVK.dylib COPYONLY) message(STATUS "MoltenVK library ${MOLTEN_VK}") + if (EXISTS ${CMAKE_BINARY_DIR}/libMoltenVK.dylib) + install(FILES ${CMAKE_BINARY_DIR}/libMoltenVK.dylib DESTINATION ${INSTALL_LIB_DIR}/runtime) + endif() endif() endif () @@ -437,7 +441,7 @@ if(NOT TI_EMSCRIPTENED) # Cannot compile Python source code with Android, but TI_EXPORT_CORE should be set and # Android should only use the isolated library ignoring those source code. if (NOT ANDROID) - add_library(${CORE_WITH_PYBIND_LIBRARY_NAME} SHARED ${TAICHI_PYBIND_SOURCE}) + pybind11_add_module(${CORE_WITH_PYBIND_LIBRARY_NAME} ${TAICHI_PYBIND_SOURCE}) else() add_library(${CORE_WITH_PYBIND_LIBRARY_NAME} SHARED) endif () @@ -459,6 +463,10 @@ if(NOT TI_EMSCRIPTENED) set_target_properties(${CORE_WITH_PYBIND_LIBRARY_NAME} PROPERTIES RUNTIME_OUTPUT_DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/runtimes") endif () + + install(TARGETS ${CORE_WITH_PYBIND_LIBRARY_NAME} + RUNTIME DESTINATION ${INSTALL_LIB_DIR}/core + LIBRARY DESTINATION ${INSTALL_LIB_DIR}/core) endif() if(TI_EMSCRIPTENED) @@ -487,3 +495,8 @@ endif() target_link_libraries(${CORE_LIBRARY_NAME} imgui) endif() + +if (NOT APPLE) + install(FILES ${CMAKE_SOURCE_DIR}/external/cuda_libdevice/slim_libdevice.10.bc + DESTINATION ${INSTALL_LIB_DIR}/runtime) +endif() diff --git a/pyproject.toml b/pyproject.toml index c4a857f234460..f5d511fad16dc 100644 --- a/pyproject.toml +++ b/pyproject.toml @@ -1,5 +1,9 @@ [build-system] -requires = ["setuptools", "wheel", "numpy", "pybind11", "cmake"] +requires = [ + "setuptools", "wheel", + "numpy", "pybind11", "cmake", + "scikit-build", "ninja; platform_system != 'Windows'", +] build-backend = "setuptools.build_meta" [tool.pytest.ini_options] diff --git a/requirements_dev.txt b/requirements_dev.txt index e8849dc3cb1aa..2cc165f61719d 100644 --- a/requirements_dev.txt +++ b/requirements_dev.txt @@ -14,3 +14,6 @@ twine wheel astunparse pre-commit +scikit-build +numpy +ninja; platform_system != 'Windows' diff --git a/scripts/run_clang_tidy.py b/scripts/run_clang_tidy.py index 229d91bfdf4d0..4e0abeb45fcc4 100644 --- a/scripts/run_clang_tidy.py +++ b/scripts/run_clang_tidy.py @@ -77,6 +77,18 @@ def make_absolute(f, directory): return os.path.normpath(os.path.join(directory, f)) +def cmake_configure(source_path='.'): + import shlex + + from skbuild.cmaker import CMaker + from skbuild.constants import CMAKE_BUILD_DIR + + cmaker = CMaker() + cmake_args = shlex.split(os.getenv('CI_SETUP_CMAKE_ARGS', '')) + cmaker.configure(cmake_args) + return CMAKE_BUILD_DIR() + + def get_tidy_invocation(f, clang_tidy_binary, checks, tmpdir, build_path, header_filter, extra_arg, extra_arg_before, quiet, config): @@ -265,8 +277,7 @@ def main(): if args.build_path is not None: build_path = args.build_path else: - # Find our database - build_path = find_compilation_database(db_path) + build_path = cmake_configure('.') try: invocation = [args.clang_tidy_binary, '-list-checks'] diff --git a/setup.py b/setup.py index 2110a34697234..a776e2d18d26f 100644 --- a/setup.py +++ b/setup.py @@ -8,17 +8,17 @@ import glob import multiprocessing import os -import platform import shutil import subprocess import sys from distutils.command.clean import clean from distutils.dir_util import remove_tree -from setuptools import Extension, find_packages, setup -from setuptools.command.build_ext import build_ext -from setuptools.command.build_py import build_py -from setuptools.command.egg_info import egg_info +from setuptools import find_packages +from skbuild import setup +from skbuild.command.egg_info import egg_info + +root_dir = os.path.dirname(os.path.abspath(__file__)) classifiers = [ 'Development Status :: 2 - Pre-Alpha', @@ -58,43 +58,11 @@ def get_version(): # Our python package root dir is python/ package_dir = 'python' -root_dir = os.path.abspath(os.path.dirname(__file__)) - - -def get_python_executable(): - return sys.executable.replace('\\', '/') - - -def get_os_name(): - name = platform.platform() - # in python 3.8, platform.platform() uses mac_ver() on macOS - # it will return 'macOS-XXXX' instead of 'Darwin-XXXX' - if name.lower().startswith('darwin') or name.lower().startswith('macos'): - return 'osx' - elif name.lower().startswith('windows'): - return 'win' - elif name.lower().startswith('linux'): - return 'linux' - elif 'bsd' in name.lower(): - return 'unix' - assert False, "Unknown platform name %s" % name - def remove_tmp(taichi_dir): shutil.rmtree(os.path.join(taichi_dir, 'assets'), ignore_errors=True) -def remove_files_with_extension(dir_name, extension): - for file in os.listdir(dir_name): - if file.endswith(extension): - os.remove(os.path.join(dir_name, file)) - - -class CMakeExtension(Extension): - def __init__(self, name): - Extension.__init__(self, name, sources=[]) - - class EggInfo(egg_info): def run(self): taichi_dir = os.path.join(package_dir, 'taichi') @@ -105,148 +73,15 @@ def run(self): egg_info.run(self) -# python setup.py build runs the following commands in order: -# python setup.py build_py -# python setup.py build_ext -class BuildPy(build_py): - def run(self): - build_py.run(self) - taichi_dir = os.path.join(package_dir, 'taichi') - remove_tmp(taichi_dir) - - -class CMakeBuild(build_ext): - def parse_cmake_args_from_env(self): - # Source: TAICHI_CMAKE_ARGS=... python setup.py ... - import shlex - cmake_args = os.getenv('TAICHI_CMAKE_ARGS', '') - return shlex.split(cmake_args.strip()) - - def run(self): - try: - subprocess.check_call(['cmake', '--version']) - except OSError: - raise RuntimeError( - "CMake must be installed to build the following extensions: " + - ", ".join(e.name for e in self.extensions)) - - # CMakeLists.txt is in the same directory as this setup.py file - cmake_list_dir = root_dir - self.build_temp = os.path.join(cmake_list_dir, 'build') - - build_directory = os.path.abspath(self.build_temp) - - cmake_args = self.parse_cmake_args_from_env() - - cmake_args += [ - f'-DCMAKE_LIBRARY_OUTPUT_DIRECTORY={build_directory}', - f'-DPYTHON_EXECUTABLE={get_python_executable()}', - f'-DTI_VERSION_MAJOR={TI_VERSION_MAJOR}', - f'-DTI_VERSION_MINOR={TI_VERSION_MINOR}', - f'-DTI_VERSION_PATCH={TI_VERSION_PATCH}', - ] - - emscriptened = os.getenv('TI_EMSCRIPTENED', '0') in ('1', 'ON') - if emscriptened: - cmake_args += ['-DTI_EMSCRIPTENED=ON'] - - if shutil.which('ninja'): - cmake_args += ['-GNinja'] - - cfg = 'Release' - if (os.getenv('DEBUG', '0') in ('1', 'ON')): - cfg = 'Debug' - elif (os.getenv('RELWITHDEBINFO', '0') in ('1', 'ON')): - cfg = 'RelWithDebInfo' - elif (os.getenv('MINSIZEREL', '0') in ('1', 'ON')): - cfg = 'MinSizeRel' - - build_args = ['--config', cfg] - - cmake_args += ['-DCMAKE_BUILD_TYPE=' + cfg] - - # Assuming Makefiles - if get_os_name() != 'win': - num_threads = os.getenv('BUILD_NUM_THREADS', - multiprocessing.cpu_count()) - build_args += ['--', f'-j{num_threads}'] - - self.build_args = build_args - - env = os.environ.copy() - os.makedirs(self.build_temp, exist_ok=True) - - print('-' * 10, 'Running CMake prepare', '-' * 40) - print(' '.join(['cmake', cmake_list_dir] + cmake_args)) - subprocess.check_call(['cmake', cmake_list_dir] + cmake_args, - cwd=self.build_temp, - env=env) - - print('-' * 10, 'Building extensions', '-' * 40) - cmake_cmd = ['cmake', '--build', '.'] + self.build_args - subprocess.check_call(cmake_cmd, cwd=self.build_temp) - - self.prepare_package() - - def prepare_package(self): - # We need to make sure these additional files are ready for - # - develop mode: must exist in local python/taichi/lib/ folder - # - install mode: must exist in self.build_lib/taichi/lib - base_dir = package_dir if self.inplace else self.build_lib - taichi_lib_dir = os.path.join(base_dir, 'taichi', '_lib') - - runtime_dir = os.path.join(taichi_lib_dir, "runtime") - core_dir = os.path.join(taichi_lib_dir, "core") - os.makedirs(runtime_dir, exist_ok=True) - os.makedirs(core_dir, exist_ok=True) - - if (get_os_name() == 'linux' or get_os_name() == 'unix' - or get_os_name() == 'osx'): - remove_files_with_extension(core_dir, ".so") - else: - remove_files_with_extension(core_dir, ".pyd") - if get_os_name() == 'osx': - remove_files_with_extension(runtime_dir, ".dylib") - remove_files_with_extension(runtime_dir, ".bc") - - if get_os_name() == 'linux' or get_os_name() == 'unix': - self.copy_file(os.path.join(self.build_temp, 'libtaichi_core.so'), - os.path.join(core_dir, 'taichi_core.so')) - elif get_os_name() == 'osx': - self.copy_file( - os.path.join(self.build_temp, 'libtaichi_core.dylib'), - os.path.join(core_dir, 'taichi_core.so')) - moltenvk_path = os.path.join(self.build_temp, 'libMoltenVK.dylib') - if os.path.exists(moltenvk_path): - self.copy_file(moltenvk_path, - os.path.join(runtime_dir, 'libMoltenVK.dylib')) - else: - self.copy_file('runtimes/taichi_core.dll', - os.path.join(core_dir, 'taichi_core.pyd')) - - if get_os_name() != 'osx': - libdevice_path = 'external/cuda_libdevice/slim_libdevice.10.bc' - print("copying libdevice:", libdevice_path) - assert os.path.exists(libdevice_path) - self.copy_file(libdevice_path, - os.path.join(runtime_dir, 'slim_libdevice.10.bc')) - - llvm_runtime_dir = 'taichi/runtime/llvm' - for f in os.listdir(llvm_runtime_dir): - if f.startswith('runtime_') and f.endswith('.bc'): - print(f"Fetching runtime file {f} to {taichi_lib_dir} folder") - self.copy_file(os.path.join(llvm_runtime_dir, f), runtime_dir) - - class Clean(clean): def run(self): super().run() - self.build_temp = os.path.join(root_dir, 'build') + self.build_temp = os.path.join(root_dir, '_skbuild') if os.path.exists(self.build_temp): remove_tree(self.build_temp, dry_run=self.dry_run) generated_folders = ('bin', 'dist', 'python/taichi/assets', - 'python/taichi/_lib/runtime', - 'python/taichi.egg-info') + 'python/taichi/_lib/runtime', 'taichi.egg-info', + 'python/taichi.egg-info', 'build') for d in generated_folders: if os.path.exists(d): remove_tree(d, dry_run=self.dry_run) @@ -263,6 +98,45 @@ def run(self): os.remove(f) +def get_cmake_args(): + import shlex + + num_threads = os.getenv('BUILD_NUM_THREADS', multiprocessing.cpu_count()) + cmake_args = shlex.split(os.getenv('TAICHI_CMAKE_ARGS', '').strip()) + + if (os.getenv('DEBUG', '0') in ('1', 'ON')): + cfg = 'Debug' + elif (os.getenv('RELWITHDEBINFO', '0') in ('1', 'ON')): + cfg = 'RelWithDebInfo' + elif (os.getenv('MINSIZEREL', '0') in ('1', 'ON')): + cfg = 'MinSizeRel' + else: + cfg = None + if cfg: + sys.argv[2:2] = ['--build-type', cfg] + + cmake_args += [ + f'-DTI_VERSION_MAJOR={TI_VERSION_MAJOR}', + f'-DTI_VERSION_MINOR={TI_VERSION_MINOR}', + f'-DTI_VERSION_PATCH={TI_VERSION_PATCH}', + ] + emscriptened = os.getenv('TI_EMSCRIPTENED', '0') in ('1', 'ON') + if emscriptened: + cmake_args += ['-DTI_EMSCRIPTENED=ON'] + + if sys.platform != 'win32': + os.environ['SKBUILD_BUILD_OPTIONS'] = f'-j{num_threads}' + return cmake_args + + +def exclude_paths(manifest_files): + return [ + f for f in manifest_files + if f.endswith(('.so', 'pyd', + '.bc')) or os.path.basename(f) == 'libMoltenVK.dylib' + ] + + setup(name=project_name, packages=packages, package_dir={"": package_dir}, @@ -287,9 +161,10 @@ def run(self): ], }, classifiers=classifiers, - ext_modules=[CMakeExtension('taichi_core')], - cmdclass=dict(egg_info=EggInfo, - build_py=BuildPy, - build_ext=CMakeBuild, - clean=Clean), + cmake_args=get_cmake_args(), + cmake_process_manifest_hook=exclude_paths, + cmdclass={ + 'egg_info': EggInfo, + 'clean': Clean + }, has_ext_modules=lambda: True) From 5bb28e25c22669b6f0da950557c5b171ce19dadd Mon Sep 17 00:00:00 2001 From: Frost Ming Date: Wed, 27 Apr 2022 12:49:07 +0800 Subject: [PATCH 07/12] [build] Install export core library to build dir (#4866) --- cmake/TaichiExportCore.cmake | 1 + 1 file changed, 1 insertion(+) diff --git a/cmake/TaichiExportCore.cmake b/cmake/TaichiExportCore.cmake index c7fc0335948bf..5ffaa960a963e 100644 --- a/cmake/TaichiExportCore.cmake +++ b/cmake/TaichiExportCore.cmake @@ -4,3 +4,4 @@ set(TAICHI_EXPORT_CORE_NAME taichi_export_core) add_library(${TAICHI_EXPORT_CORE_NAME} SHARED) target_link_libraries(${TAICHI_EXPORT_CORE_NAME} taichi_isolated_core) +install(TARGETS ${TAICHI_EXPORT_CORE_NAME} DESTINATION ${CMAKE_CURRENT_SOURCE_DIR}/build) From 5d20a1d7a796418175009399606477a32ddc5f82 Mon Sep 17 00:00:00 2001 From: Taichi Gardener <62079278+taichi-gardener@users.noreply.github.com> Date: Wed, 27 Apr 2022 13:47:38 +0800 Subject: [PATCH 08/12] [misc] Bump version to v1.0.2 (#4867) --- version.txt | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/version.txt b/version.txt index b18d46540b351..570c796513fb7 100644 --- a/version.txt +++ b/version.txt @@ -1 +1 @@ -v1.0.1 +v1.0.2 From 92f8464f23c9561c28df1950cf960978dd7e3b97 Mon Sep 17 00:00:00 2001 From: Zhanlue Yang Date: Wed, 27 Apr 2022 15:45:19 +0800 Subject: [PATCH 09/12] [Bug] Remove redundant AllocStmt when lowering FrontendWhileStmt (#4870) --- taichi/transforms/lower_ast.cpp | 1 - 1 file changed, 1 deletion(-) diff --git a/taichi/transforms/lower_ast.cpp b/taichi/transforms/lower_ast.cpp index 5562ec9bb9188..5141724e19deb 100644 --- a/taichi/transforms/lower_ast.cpp +++ b/taichi/transforms/lower_ast.cpp @@ -156,7 +156,6 @@ class LowerAST : public IRVisitor { stmts->insert( std::make_unique(new_while->mask, cond_stmt), fctx.stmts.size()); - stmt->insert_before_me(std::make_unique(PrimitiveType::i32)); auto &&const_stmt = std::make_unique(TypedConstant((int32)0xFFFFFFFF)); auto const_stmt_ptr = const_stmt.get(); From 5bb3b0ee758dc6062089b8441010e08356c1005a Mon Sep 17 00:00:00 2001 From: Frost Ming Date: Wed, 27 Apr 2022 18:41:44 +0800 Subject: [PATCH 10/12] [build] [bug] Fix a bug of skbuild that loses the root package_dir (#4875) --- setup.py | 6 ++++++ 1 file changed, 6 insertions(+) diff --git a/setup.py b/setup.py index a776e2d18d26f..53752a499083f 100644 --- a/setup.py +++ b/setup.py @@ -64,6 +64,12 @@ def remove_tmp(taichi_dir): class EggInfo(egg_info): + def finalize_options(self, *args, **kwargs): + if '' not in self.distribution.package_dir: + # XXX: skbuild loses the root package dir + self.distribution.package_dir[''] = package_dir + return super().finalize_options(*args, **kwargs) + def run(self): taichi_dir = os.path.join(package_dir, 'taichi') remove_tmp(taichi_dir) From 477c996e84f4d08e70674e367982f1f394623efb Mon Sep 17 00:00:00 2001 From: Ailing Date: Wed, 27 Apr 2022 21:49:43 +0800 Subject: [PATCH 11/12] [ci] Add libtaichi_export_core build for desktop in CI (#4871) --- .github/workflows/scripts/unix_build.sh | 19 ++++++--- .github/workflows/testing.yml | 54 +++++++++++++++++++++++++ cmake/TaichiExportCore.cmake | 1 - 3 files changed, 68 insertions(+), 6 deletions(-) diff --git a/.github/workflows/scripts/unix_build.sh b/.github/workflows/scripts/unix_build.sh index a4c1f5679662d..0ad894d176fbd 100755 --- a/.github/workflows/scripts/unix_build.sh +++ b/.github/workflows/scripts/unix_build.sh @@ -49,7 +49,7 @@ setup_python() { python3 -m pip install -r requirements_dev.txt } -build() { +build_taichi_wheel() { git fetch origin master PROJECT_TAGS="" EXTRA_ARGS="" @@ -69,12 +69,21 @@ build() { sccache -s } +build_libtaichi_export() { + git fetch origin master + python3 setup.py build_ext +} + setup_sccache setup_python -build -cat "$SCCACHE_ERROR_LOG" || true -NUM_WHL=$(ls dist/*.whl | wc -l) -if [ $NUM_WHL -ne 1 ]; then echo "ERROR: created more than 1 whl." && exit 1; fi + +if [ "$EXPORT_CORE" == "1" ]; then + build_libtaichi_export +else + build_taichi_wheel + NUM_WHL=$(ls dist/*.whl | wc -l) + if [ $NUM_WHL -ne 1 ]; then echo "ERROR: created more than 1 whl." && exit 1; fi +fi chmod -R 777 "$SCCACHE_DIR" rm -f python/CHANGELOG.md diff --git a/.github/workflows/testing.yml b/.github/workflows/testing.yml index ccc2b08d2ffc7..00edaaa750b9f 100644 --- a/.github/workflows/testing.yml +++ b/.github/workflows/testing.yml @@ -415,3 +415,57 @@ jobs: PY: ${{ matrix.python }} PLATFORM: 'm1' TI_CI: 1 + + build_libtaichi_export: + name: Build libtaichi_export.so(GPU) + needs: check_files + runs-on: [self-hosted, cuda, vulkan, cn] + timeout-minutes: 60 + strategy: + matrix: + include: + - os: ubuntu-latest + python: py39 + with_cc: ON + permissions: + packages: read + contents: read + steps: + - uses: actions/checkout@v2 + with: + submodules: "recursive" + + - name: Get sccache cache + uses: actions/cache@v2 + with: + path: sccache_cache + key: sccache-linux-${{matrix.with_cc}}-${{ github.sha }} + restore-keys: | + sccache-linux-${{matrix.with_cc}}- + + - name: Build For Desktop + run: | + if [[ ${{needs.check_files.outputs.run_job}} == false ]]; then + exit 0 + fi + docker create --user dev --name taichi_build_desktop --gpus all -v /tmp/.X11-unix:/tmp/.X11-unix \ + -e PY -e GPU_BUILD -e PROJECT_NAME -e TAICHI_CMAKE_ARGS -e DISPLAY -e EXPORT_CORE\ + registry.taichigraphics.com/taichidev-ubuntu18.04:v0.2.1 \ + /home/dev/taichi/.github/workflows/scripts/unix_build.sh + # A tarball is needed because sccache needs some permissions that only the file owner has. + # 1000 is the uid and gid of user "dev" in the container. + # If the uid or gid of the user inside the docker changes, please change the uid and gid in the following line. + tar -cf - ../${{ github.event.repository.name }} --mode u=+rwx,g=+rwx,o=+rwx --owner 1000 --group 1000 | docker cp - taichi_build_desktop:/home/dev/ + docker start -a taichi_build_desktop + env: + PY: ${{ matrix.python }} + GPU_BUILD: ON + PROJECT_NAME: taichi + TAICHI_CMAKE_ARGS: -DTI_WITH_VULKAN:BOOL=ON -DTI_WITH_CUDA:BOOL=OFF -DTI_WITH_OPENGL:BOOL=OFF -DTI_WITH_LLVM:BOOL=OFF -DCMAKE_C_COMPILER_LAUNCHER=sccache -DCMAKE_CXX_COMPILER_LAUNCHER=sccache -DTI_EXPORT_CORE:BOOL=ON + EXPORT_CORE: 1 + DISPLAY: :1 + + - name: clean docker container + if: always() + run: | + docker rm taichi_build_desktop -f diff --git a/cmake/TaichiExportCore.cmake b/cmake/TaichiExportCore.cmake index 5ffaa960a963e..c7fc0335948bf 100644 --- a/cmake/TaichiExportCore.cmake +++ b/cmake/TaichiExportCore.cmake @@ -4,4 +4,3 @@ set(TAICHI_EXPORT_CORE_NAME taichi_export_core) add_library(${TAICHI_EXPORT_CORE_NAME} SHARED) target_link_libraries(${TAICHI_EXPORT_CORE_NAME} taichi_isolated_core) -install(TARGETS ${TAICHI_EXPORT_CORE_NAME} DESTINATION ${CMAKE_CURRENT_SOURCE_DIR}/build) From 6e055f0c28c1092af9402e25b1c03325e7458f62 Mon Sep 17 00:00:00 2001 From: Bo Qiao Date: Wed, 27 Apr 2022 23:53:55 +0800 Subject: [PATCH 12/12] [Build] [refactor] Define runtime build target (#4838) * Move LLVM Cmake to its own dir * Suppress warning from submodules * [pre-commit.ci] auto fixes from pre-commit.com hooks for more information, see https://pre-commit.ci * Use current source dir * [pre-commit.ci] auto fixes from pre-commit.com hooks for more information, see https://pre-commit.ci * Separate Vulkan runtime files from codegen * [pre-commit.ci] auto fixes from pre-commit.com hooks for more information, see https://pre-commit.ci Co-authored-by: pre-commit-ci[bot] <66853113+pre-commit-ci[bot]@users.noreply.github.com> --- CMakeLists.txt | 11 +---------- cmake/TaichiCore.cmake | 13 ++++++++----- taichi/backends/dx/dx_program.h | 2 +- .../backends/vulkan/aot_module_builder_impl.h | 2 +- .../vulkan/aot_module_loader_impl.cpp | 2 +- .../backends/vulkan/aot_module_loader_impl.h | 2 +- taichi/backends/vulkan/snode_tree_manager.cpp | 2 +- taichi/backends/vulkan/vulkan_program.h | 2 +- taichi/python/export_misc.cpp | 2 +- taichi/runtime/llvm/CMakeLists.txt | 19 +++++++++++++++++++ taichi/runtime/vulkan/CMakeLists.txt | 13 +++++++++++++ .../{backends => runtime}/vulkan/runtime.cpp | 2 +- taichi/{backends => runtime}/vulkan/runtime.h | 0 13 files changed, 49 insertions(+), 23 deletions(-) create mode 100644 taichi/runtime/llvm/CMakeLists.txt create mode 100644 taichi/runtime/vulkan/CMakeLists.txt rename taichi/{backends => runtime}/vulkan/runtime.cpp (99%) rename taichi/{backends => runtime}/vulkan/runtime.h (100%) diff --git a/CMakeLists.txt b/CMakeLists.txt index 7387e8b648b56..75d71cb604609 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -148,16 +148,7 @@ if (${CLANG_VERSION_MAJOR} VERSION_GREATER ${CLANG_HIGHEST_VERSION}) endif() endif() -# Build llvm-runtime for host arch and cuda (if available) -foreach(arch IN LISTS HOST_ARCH CUDA_ARCH) - add_custom_target( - "generate_llvm_runtime_${arch}" - COMMAND ${CLANG_EXECUTABLE} ${CLANG_OSX_FLAGS} -c runtime.cpp -o "runtime_${arch}.bc" -fno-exceptions -emit-llvm -std=c++17 -D "ARCH_${arch}" -I ${PROJECT_SOURCE_DIR}; - WORKING_DIRECTORY "${PROJECT_SOURCE_DIR}/taichi/runtime/llvm" - ) - add_dependencies(${CORE_LIBRARY_NAME} "generate_llvm_runtime_${arch}") - install(FILES "${PROJECT_SOURCE_DIR}/taichi/runtime/llvm/runtime_${arch}.bc" DESTINATION ${CMAKE_INSTALL_PREFIX}/python/taichi/_lib/runtime) -endforeach() +add_subdirectory(taichi/runtime/llvm) configure_file(taichi/common/version.h.in ${CMAKE_SOURCE_DIR}/taichi/common/version.h) configure_file(taichi/common/commit_hash.h.in ${CMAKE_SOURCE_DIR}/taichi/common/commit_hash.h) diff --git a/cmake/TaichiCore.cmake b/cmake/TaichiCore.cmake index 687130d9e19d5..8b740b5e74f82 100644 --- a/cmake/TaichiCore.cmake +++ b/cmake/TaichiCore.cmake @@ -16,6 +16,9 @@ option(TI_EMSCRIPTENED "Build using emscripten" OFF) # projects. set(CMAKE_CXX_VISIBILITY_PRESET hidden) set(CMAKE_VISIBILITY_INLINES_HIDDEN ON) +# Suppress warnings from submodules introduced by the above symbol visibility change +set(CMAKE_POLICY_DEFAULT_CMP0063 NEW) +set(CMAKE_POLICY_DEFAULT_CMP0077 NEW) set(INSTALL_LIB_DIR ${CMAKE_INSTALL_PREFIX}/python/taichi/_lib) if(ANDROID) @@ -140,10 +143,6 @@ file(GLOB TAICHI_OPENGL_REQUIRED_SOURCE "taichi/backends/opengl/codegen_opengl.*" "taichi/backends/opengl/struct_opengl.*" ) -file(GLOB TAICHI_VULKAN_REQUIRED_SOURCE - "taichi/backends/vulkan/runtime.h" - "taichi/backends/vulkan/runtime.cpp" -) list(REMOVE_ITEM TAICHI_CORE_SOURCE ${TAICHI_BACKEND_SOURCE}) @@ -198,7 +197,7 @@ if (TI_WITH_VULKAN) set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -DTI_WITH_VULKAN") list(APPEND TAICHI_CORE_SOURCE ${TAICHI_VULKAN_SOURCE}) endif() -list(APPEND TAICHI_CORE_SOURCE ${TAICHI_VULKAN_REQUIRED_SOURCE}) + if (TI_WITH_DX11) set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -DTI_WITH_DX11") @@ -389,8 +388,12 @@ if (TI_WITH_VULKAN) install(FILES ${CMAKE_BINARY_DIR}/libMoltenVK.dylib DESTINATION ${INSTALL_LIB_DIR}/runtime) endif() endif() + + add_subdirectory(taichi/runtime/vulkan) + target_link_libraries(${CORE_LIBRARY_NAME} vulkan_runtime) endif () + # Optional dependencies if (APPLE) diff --git a/taichi/backends/dx/dx_program.h b/taichi/backends/dx/dx_program.h index 4974139147b0c..493cb5c62c69a 100644 --- a/taichi/backends/dx/dx_program.h +++ b/taichi/backends/dx/dx_program.h @@ -3,7 +3,7 @@ #ifdef TI_WITH_DX11 #include "taichi/backends/dx/dx_device.h" -#include "taichi/backends/vulkan/runtime.h" +#include "taichi/runtime/vulkan/runtime.h" #include "taichi/backends/vulkan/snode_tree_manager.h" #include "taichi/program/program_impl.h" diff --git a/taichi/backends/vulkan/aot_module_builder_impl.h b/taichi/backends/vulkan/aot_module_builder_impl.h index bbd6b40e4df48..0accfcc203343 100644 --- a/taichi/backends/vulkan/aot_module_builder_impl.h +++ b/taichi/backends/vulkan/aot_module_builder_impl.h @@ -5,7 +5,7 @@ #include "taichi/aot/module_builder.h" #include "taichi/backends/vulkan/aot_utils.h" -#include "taichi/backends/vulkan/runtime.h" +#include "taichi/runtime/vulkan/runtime.h" #include "taichi/codegen/spirv/snode_struct_compiler.h" #include "taichi/codegen/spirv/kernel_utils.h" diff --git a/taichi/backends/vulkan/aot_module_loader_impl.cpp b/taichi/backends/vulkan/aot_module_loader_impl.cpp index a773f71f37def..5f87eb4fc9ca8 100644 --- a/taichi/backends/vulkan/aot_module_loader_impl.cpp +++ b/taichi/backends/vulkan/aot_module_loader_impl.cpp @@ -3,7 +3,7 @@ #include #include -#include "taichi/backends/vulkan/runtime.h" +#include "taichi/runtime/vulkan/runtime.h" namespace taichi { namespace lang { diff --git a/taichi/backends/vulkan/aot_module_loader_impl.h b/taichi/backends/vulkan/aot_module_loader_impl.h index 7d32d991f2e89..b188281cb749d 100644 --- a/taichi/backends/vulkan/aot_module_loader_impl.h +++ b/taichi/backends/vulkan/aot_module_loader_impl.h @@ -5,7 +5,7 @@ #include #include "taichi/backends/vulkan/aot_utils.h" -#include "taichi/backends/vulkan/runtime.h" +#include "taichi/runtime/vulkan/runtime.h" #include "taichi/codegen/spirv/kernel_utils.h" #include "taichi/aot/module_loader.h" diff --git a/taichi/backends/vulkan/snode_tree_manager.cpp b/taichi/backends/vulkan/snode_tree_manager.cpp index 0bfb6d2f01edd..b7d4816ae42d9 100644 --- a/taichi/backends/vulkan/snode_tree_manager.cpp +++ b/taichi/backends/vulkan/snode_tree_manager.cpp @@ -1,6 +1,6 @@ #include "taichi/backends/vulkan/snode_tree_manager.h" -#include "taichi/backends/vulkan/runtime.h" +#include "taichi/runtime/vulkan/runtime.h" namespace taichi { namespace lang { diff --git a/taichi/backends/vulkan/vulkan_program.h b/taichi/backends/vulkan/vulkan_program.h index a94f2abbb2ba7..b3b33348c525f 100644 --- a/taichi/backends/vulkan/vulkan_program.h +++ b/taichi/backends/vulkan/vulkan_program.h @@ -6,7 +6,7 @@ #include "taichi/backends/vulkan/vulkan_device_creator.h" #include "taichi/backends/vulkan/vulkan_utils.h" #include "taichi/backends/vulkan/vulkan_loader.h" -#include "taichi/backends/vulkan/runtime.h" +#include "taichi/runtime/vulkan/runtime.h" #include "taichi/backends/vulkan/snode_tree_manager.h" #include "taichi/backends/vulkan/vulkan_device.h" #include "vk_mem_alloc.h" diff --git a/taichi/python/export_misc.cpp b/taichi/python/export_misc.cpp index e75df0b88d856..a1abb389ab782 100644 --- a/taichi/python/export_misc.cpp +++ b/taichi/python/export_misc.cpp @@ -5,7 +5,7 @@ #include "taichi/backends/metal/api.h" #include "taichi/backends/opengl/opengl_api.h" -#include "taichi/backends/vulkan/runtime.h" +#include "taichi/runtime/vulkan/runtime.h" #include "taichi/backends/dx/dx_api.h" #include "taichi/common/core.h" #include "taichi/common/interface.h" diff --git a/taichi/runtime/llvm/CMakeLists.txt b/taichi/runtime/llvm/CMakeLists.txt new file mode 100644 index 0000000000000..fdb5c723c2fb7 --- /dev/null +++ b/taichi/runtime/llvm/CMakeLists.txt @@ -0,0 +1,19 @@ +# ./taichi/runtime/llvm/CMakeLists.txt + +function(COMPILE_LLVM_RUNTIME rtm_arch) + message(STATUS "Compiling LLVM byte code file for arch ${rtm_arch}") + # Keep this for now, as .bc need to be generated. + add_custom_target( + "generate_llvm_runtime_${rtm_arch}" + COMMAND ${CLANG_EXECUTABLE} ${CLANG_OSX_FLAGS} -c runtime.cpp -o "runtime_${rtm_arch}.bc" -fno-exceptions -emit-llvm -std=c++17 -D "ARCH_${rtm_arch}" -I ${PROJECT_SOURCE_DIR}; + # TODO, it's better to avoid polluting the source dir, keep in build + WORKING_DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}" + ) + add_dependencies(${CORE_LIBRARY_NAME} "generate_llvm_runtime_${rtm_arch}") + install(FILES "${CMAKE_SOURCE_DIR}/taichi/runtime/llvm/runtime_${arch}.bc" DESTINATION ${CMAKE_INSTALL_PREFIX}/python/taichi/_lib/runtime) +endfunction() + +# Build llvm-runtime for host arch and cuda (if available) +foreach(arch IN LISTS HOST_ARCH CUDA_ARCH) + compile_llvm_runtime(${arch}) +endforeach() diff --git a/taichi/runtime/vulkan/CMakeLists.txt b/taichi/runtime/vulkan/CMakeLists.txt new file mode 100644 index 0000000000000..00ecee7a09caf --- /dev/null +++ b/taichi/runtime/vulkan/CMakeLists.txt @@ -0,0 +1,13 @@ +# ./taichi/runtime/vulkan/CMakeLists.txt + +add_library(vulkan_runtime) +target_sources(vulkan_runtime + PRIVATE + runtime.cpp + ) +target_include_directories(vulkan_runtime + PRIVATE + ${PROJECT_SOURCE_DIR}/external/SPIRV-Tools/include + ${PROJECT_SOURCE_DIR}/external/eigen + ${PROJECT_SOURCE_DIR}/external/FP16/include + ) diff --git a/taichi/backends/vulkan/runtime.cpp b/taichi/runtime/vulkan/runtime.cpp similarity index 99% rename from taichi/backends/vulkan/runtime.cpp rename to taichi/runtime/vulkan/runtime.cpp index 7da6b181719b6..82478b05bf427 100644 --- a/taichi/backends/vulkan/runtime.cpp +++ b/taichi/runtime/vulkan/runtime.cpp @@ -1,4 +1,4 @@ -#include "taichi/backends/vulkan/runtime.h" +#include "taichi/runtime/vulkan/runtime.h" #include "taichi/program/program.h" #include diff --git a/taichi/backends/vulkan/runtime.h b/taichi/runtime/vulkan/runtime.h similarity index 100% rename from taichi/backends/vulkan/runtime.h rename to taichi/runtime/vulkan/runtime.h