From 5f35545095ee6790dda25820a2a2f74c67898022 Mon Sep 17 00:00:00 2001 From: rusty1s Date: Wed, 20 Apr 2022 07:54:23 +0200 Subject: [PATCH 1/2] adjust tensor creation --- csrc/cpu/metis_cpu.cpp | 4 ++-- csrc/cpu/relabel_cpu.cpp | 6 +++--- csrc/cpu/sample_cpu.cpp | 6 +++--- csrc/cpu/utils.h | 4 ++-- csrc/cuda/convert_cuda.cu | 4 ++-- csrc/cuda/spspmm_cuda.cu | 6 +++--- 6 files changed, 15 insertions(+), 15 deletions(-) diff --git a/csrc/cpu/metis_cpu.cpp b/csrc/cpu/metis_cpu.cpp index 05fe9dee..8de38b51 100644 --- a/csrc/cpu/metis_cpu.cpp +++ b/csrc/cpu/metis_cpu.cpp @@ -44,7 +44,7 @@ torch::Tensor partition_cpu(torch::Tensor rowptr, torch::Tensor col, vwgt = optional_node_weight.value().data_ptr(); int64_t objval = -1; - auto part = torch::empty(nvtxs, rowptr.options()); + auto part = torch::empty({nvtxs}, rowptr.options()); auto part_data = part.data_ptr(); if (recursive) { @@ -99,7 +99,7 @@ mt_partition_cpu(torch::Tensor rowptr, torch::Tensor col, mtmetis_pid_type nparts = num_parts; mtmetis_wgt_type objval = -1; - auto part = torch::empty(nvtxs, rowptr.options()); + auto part = torch::empty({nvtxs}, rowptr.options()); mtmetis_pid_type *part_data = (mtmetis_pid_type *)part.data_ptr(); double *opts = mtmetis_init_options(); diff --git a/csrc/cpu/relabel_cpu.cpp b/csrc/cpu/relabel_cpu.cpp index b7556e80..6dc94c80 100644 --- a/csrc/cpu/relabel_cpu.cpp +++ b/csrc/cpu/relabel_cpu.cpp @@ -64,7 +64,7 @@ relabel_one_hop_cpu(torch::Tensor rowptr, torch::Tensor col, std::unordered_map n_id_map; std::unordered_map::iterator it; - auto out_rowptr = torch::empty(idx.numel() + 1, rowptr.options()); + auto out_rowptr = torch::empty({idx.numel() + 1}, rowptr.options()); auto out_rowptr_data = out_rowptr.data_ptr(); out_rowptr_data[0] = 0; @@ -76,12 +76,12 @@ relabel_one_hop_cpu(torch::Tensor rowptr, torch::Tensor col, out_rowptr_data[i + 1] = offset; } - auto out_col = torch::empty(offset, col.options()); + auto out_col = torch::empty({offset}, col.options()); auto out_col_data = out_col.data_ptr(); torch::optional out_value = torch::nullopt; if (optional_value.has_value()) { - out_value = torch::empty(offset, optional_value.value().options()); + out_value = torch::empty({offset}, optional_value.value().options()); AT_DISPATCH_ALL_TYPES(optional_value.value().scalar_type(), "relabel", [&] { auto value_data = optional_value.value().data_ptr(); diff --git a/csrc/cpu/sample_cpu.cpp b/csrc/cpu/sample_cpu.cpp index be2dad98..f04e354a 100644 --- a/csrc/cpu/sample_cpu.cpp +++ b/csrc/cpu/sample_cpu.cpp @@ -19,7 +19,7 @@ sample_adj_cpu(torch::Tensor rowptr, torch::Tensor col, torch::Tensor idx, auto col_data = col.data_ptr(); auto idx_data = idx.data_ptr(); - auto out_rowptr = torch::empty(idx.numel() + 1, rowptr.options()); + auto out_rowptr = torch::empty({idx.numel() + 1}, rowptr.options()); auto out_rowptr_data = out_rowptr.data_ptr(); out_rowptr_data[0] = 0; @@ -117,9 +117,9 @@ sample_adj_cpu(torch::Tensor rowptr, torch::Tensor col, torch::Tensor idx, auto out_n_id = torch::from_blob(n_ids.data(), {N}, col.options()).clone(); int64_t E = out_rowptr_data[idx.numel()]; - auto out_col = torch::empty(E, col.options()); + auto out_col = torch::empty({E}, col.options()); auto out_col_data = out_col.data_ptr(); - auto out_e_id = torch::empty(E, col.options()); + auto out_e_id = torch::empty({E}, col.options()); auto out_e_id_data = out_e_id.data_ptr(); i = 0; diff --git a/csrc/cpu/utils.h b/csrc/cpu/utils.h index 0684d1f4..6861f12a 100644 --- a/csrc/cpu/utils.h +++ b/csrc/cpu/utils.h @@ -61,7 +61,7 @@ choice(int64_t population, int64_t num_samples, bool replace = false, return torch::multinomial(weight.value(), num_samples, replace); if (replace) { - const auto out = torch::empty(num_samples, at::kLong); + const auto out = torch::empty({num_samples}, at::kLong); auto *out_data = out.data_ptr(); for (int64_t i = 0; i < num_samples; i++) { out_data[i] = uniform_randint(population); @@ -72,7 +72,7 @@ choice(int64_t population, int64_t num_samples, bool replace = false, // Sample without replacement via Robert Floyd algorithm: // https://www.nowherenearithaca.com/2013/05/ // robert-floyds-tiny-and-beautiful.html - const auto out = torch::empty(num_samples, at::kLong); + const auto out = torch::empty({num_samples}, at::kLong); auto *out_data = out.data_ptr(); std::unordered_set samples; for (int64_t i = population - num_samples; i < population; i++) { diff --git a/csrc/cuda/convert_cuda.cu b/csrc/cuda/convert_cuda.cu index 01177f32..30f7d273 100644 --- a/csrc/cuda/convert_cuda.cu +++ b/csrc/cuda/convert_cuda.cu @@ -27,7 +27,7 @@ torch::Tensor ind2ptr_cuda(torch::Tensor ind, int64_t M) { CHECK_CUDA(ind); cudaSetDevice(ind.get_device()); - auto out = torch::empty(M + 1, ind.options()); + auto out = torch::empty({M + 1}, ind.options()); if (ind.numel() == 0) return out.zero_(); @@ -57,7 +57,7 @@ torch::Tensor ptr2ind_cuda(torch::Tensor ptr, int64_t E) { CHECK_CUDA(ptr); cudaSetDevice(ptr.get_device()); - auto out = torch::empty(E, ptr.options()); + auto out = torch::empty({E}, ptr.options()); auto ptr_data = ptr.data_ptr(); auto out_data = out.data_ptr(); auto stream = at::cuda::getCurrentCUDAStream(); diff --git a/csrc/cuda/spspmm_cuda.cu b/csrc/cuda/spspmm_cuda.cu index 9627344e..9ae4444c 100644 --- a/csrc/cuda/spspmm_cuda.cu +++ b/csrc/cuda/spspmm_cuda.cu @@ -108,7 +108,7 @@ spspmm_cuda(torch::Tensor rowptrA, torch::Tensor colA, cudaMalloc(&buffer, bufferSize); // Step 3: Compute CSR row pointer. - rowptrC = torch::empty(M + 1, rowptrA.options()); + rowptrC = torch::empty({M + 1}, rowptrA.options()); auto rowptrC_data = rowptrC.data_ptr(); cusparseXcsrgemm2Nnz(handle, M, N, K, descr, colA.numel(), rowptrA_data, colA_data, descr, colB.numel(), rowptrB_data, @@ -116,11 +116,11 @@ spspmm_cuda(torch::Tensor rowptrA, torch::Tensor colA, nnzTotalDevHostPtr, info, buffer); // Step 4: Compute CSR entries. - colC = torch::empty(nnzC, rowptrC.options()); + colC = torch::empty({nnzC}, rowptrC.options()); auto colC_data = colC.data_ptr(); if (optional_valueA.has_value()) - optional_valueC = torch::empty(nnzC, optional_valueA.value().options()); + optional_valueC = torch::empty({nnzC}, optional_valueA.value().options()); scalar_t *valA_data = NULL, *valB_data = NULL, *valC_data = NULL; if (optional_valueA.has_value()) { From f2af26f1402ab40ae0a097440703723ebce03680 Mon Sep 17 00:00:00 2001 From: rusty1s Date: Wed, 20 Apr 2022 07:56:01 +0200 Subject: [PATCH 2/2] update --- csrc/cpu/diag_cpu.cpp | 2 +- csrc/cpu/spmm_cpu.cpp | 2 +- csrc/cpu/spspmm_cpu.cpp | 4 ++-- csrc/cuda/diag_cuda.cu | 2 +- csrc/cuda/spmm_cuda.cu | 2 +- csrc/cuda/spspmm_cuda.cu | 4 ++-- 6 files changed, 8 insertions(+), 8 deletions(-) diff --git a/csrc/cpu/diag_cpu.cpp b/csrc/cpu/diag_cpu.cpp index 2f6b8da6..c57ae59c 100644 --- a/csrc/cpu/diag_cpu.cpp +++ b/csrc/cpu/diag_cpu.cpp @@ -13,7 +13,7 @@ torch::Tensor non_diag_mask_cpu(torch::Tensor row, torch::Tensor col, int64_t M, auto row_data = row.data_ptr(); auto col_data = col.data_ptr(); - auto mask = torch::zeros(E + num_diag, row.options().dtype(torch::kBool)); + auto mask = torch::zeros({E + num_diag}, row.options().dtype(torch::kBool)); auto mask_data = mask.data_ptr(); int64_t r, c; diff --git a/csrc/cpu/spmm_cpu.cpp b/csrc/cpu/spmm_cpu.cpp index 090baf97..dc75def0 100644 --- a/csrc/cpu/spmm_cpu.cpp +++ b/csrc/cpu/spmm_cpu.cpp @@ -118,7 +118,7 @@ torch::Tensor spmm_value_bw_cpu(torch::Tensor row, torch::Tensor rowptr, auto K = mat.size(-1); auto B = mat.numel() / (N * K); - auto out = torch::zeros(row.numel(), grad.options()); + auto out = torch::zeros({row.numel()}, grad.options()); auto row_data = row.data_ptr(); auto rowptr_data = rowptr.data_ptr(); diff --git a/csrc/cpu/spspmm_cpu.cpp b/csrc/cpu/spspmm_cpu.cpp index 74ad0a4c..824dc228 100644 --- a/csrc/cpu/spspmm_cpu.cpp +++ b/csrc/cpu/spspmm_cpu.cpp @@ -33,11 +33,11 @@ spspmm_cpu(torch::Tensor rowptrA, torch::Tensor colA, if (!optional_valueA.has_value() && optional_valueB.has_value()) optional_valueA = - torch::ones(colA.numel(), optional_valueB.value().options()); + torch::ones({colA.numel()}, optional_valueB.value().options()); if (!optional_valueB.has_value() && optional_valueA.has_value()) optional_valueB = - torch::ones(colB.numel(), optional_valueA.value().options()); + torch::ones({colB.numel()}, optional_valueA.value().options()); auto scalar_type = torch::ScalarType::Float; if (optional_valueA.has_value()) diff --git a/csrc/cuda/diag_cuda.cu b/csrc/cuda/diag_cuda.cu index c193d07b..7b608b44 100644 --- a/csrc/cuda/diag_cuda.cu +++ b/csrc/cuda/diag_cuda.cu @@ -51,7 +51,7 @@ torch::Tensor non_diag_mask_cuda(torch::Tensor row, torch::Tensor col, auto row_data = row.data_ptr(); auto col_data = col.data_ptr(); - auto mask = torch::zeros(E + num_diag, row.options().dtype(torch::kBool)); + auto mask = torch::zeros({E + num_diag}, row.options().dtype(torch::kBool)); auto mask_data = mask.data_ptr(); if (E == 0) diff --git a/csrc/cuda/spmm_cuda.cu b/csrc/cuda/spmm_cuda.cu index 2f54fb69..2a98ebe2 100644 --- a/csrc/cuda/spmm_cuda.cu +++ b/csrc/cuda/spmm_cuda.cu @@ -213,7 +213,7 @@ torch::Tensor spmm_value_bw_cuda(torch::Tensor row, torch::Tensor rowptr, auto B = mat.numel() / (N * K); auto BLOCKS = dim3((E * 32 + THREADS - 1) / THREADS); - auto out = torch::zeros(row.numel(), grad.options()); + auto out = torch::zeros({row.numel()}, grad.options()); auto row_data = row.data_ptr(); auto rowptr_data = rowptr.data_ptr(); diff --git a/csrc/cuda/spspmm_cuda.cu b/csrc/cuda/spspmm_cuda.cu index 9ae4444c..1b932b33 100644 --- a/csrc/cuda/spspmm_cuda.cu +++ b/csrc/cuda/spspmm_cuda.cu @@ -59,11 +59,11 @@ spspmm_cuda(torch::Tensor rowptrA, torch::Tensor colA, if (!optional_valueA.has_value() && optional_valueB.has_value()) optional_valueA = - torch::ones(colA.numel(), optional_valueB.value().options()); + torch::ones({colA.numel()}, optional_valueB.value().options()); if (!optional_valueB.has_value() && optional_valueA.has_value()) optional_valueB = - torch::ones(colB.numel(), optional_valueA.value().options()); + torch::ones({colB.numel()}, optional_valueA.value().options()); auto scalar_type = torch::ScalarType::Float; if (optional_valueA.has_value())