Spaces:
Running
on
Zero
Running
on
Zero
File size: 2,740 Bytes
f29eac5 |
1 2 3 4 5 6 7 8 9 10 11 12 13 14 15 16 17 18 19 20 21 22 23 24 25 26 27 28 29 30 31 32 33 34 35 36 37 38 39 40 41 42 43 44 45 46 47 48 49 50 51 52 53 54 55 56 57 58 59 60 61 62 63 64 65 66 67 68 69 70 71 72 73 74 75 76 77 78 79 80 81 82 83 84 85 86 87 88 89 90 91 92 93 |
#include <torch/extension.h>
#include "api.h"
#include "z_order.h"
#include "hilbert.h"
torch::Tensor
z_order_encode(
const torch::Tensor& x,
const torch::Tensor& y,
const torch::Tensor& z
) {
// Allocate output tensor
torch::Tensor codes = torch::empty_like(x);
// Call CUDA kernel
z_order_encode_cuda<<<(x.size(0) + BLOCK_SIZE - 1) / BLOCK_SIZE, BLOCK_SIZE>>>(
x.size(0),
reinterpret_cast<uint32_t*>(x.contiguous().data_ptr<int>()),
reinterpret_cast<uint32_t*>(y.contiguous().data_ptr<int>()),
reinterpret_cast<uint32_t*>(z.contiguous().data_ptr<int>()),
reinterpret_cast<uint32_t*>(codes.data_ptr<int>())
);
return codes;
}
std::tuple<torch::Tensor, torch::Tensor, torch::Tensor>
z_order_decode(
const torch::Tensor& codes
) {
// Allocate output tensors
torch::Tensor x = torch::empty_like(codes);
torch::Tensor y = torch::empty_like(codes);
torch::Tensor z = torch::empty_like(codes);
// Call CUDA kernel
z_order_decode_cuda<<<(codes.size(0) + BLOCK_SIZE - 1) / BLOCK_SIZE, BLOCK_SIZE>>>(
codes.size(0),
reinterpret_cast<uint32_t*>(codes.contiguous().data_ptr<int>()),
reinterpret_cast<uint32_t*>(x.data_ptr<int>()),
reinterpret_cast<uint32_t*>(y.data_ptr<int>()),
reinterpret_cast<uint32_t*>(z.data_ptr<int>())
);
return std::make_tuple(x, y, z);
}
torch::Tensor
hilbert_encode(
const torch::Tensor& x,
const torch::Tensor& y,
const torch::Tensor& z
) {
// Allocate output tensor
torch::Tensor codes = torch::empty_like(x);
// Call CUDA kernel
hilbert_encode_cuda<<<(x.size(0) + BLOCK_SIZE - 1) / BLOCK_SIZE, BLOCK_SIZE>>>(
x.size(0),
reinterpret_cast<uint32_t*>(x.contiguous().data_ptr<int>()),
reinterpret_cast<uint32_t*>(y.contiguous().data_ptr<int>()),
reinterpret_cast<uint32_t*>(z.contiguous().data_ptr<int>()),
reinterpret_cast<uint32_t*>(codes.data_ptr<int>())
);
return codes;
}
std::tuple<torch::Tensor, torch::Tensor, torch::Tensor>
hilbert_decode(
const torch::Tensor& codes
) {
// Allocate output tensors
torch::Tensor x = torch::empty_like(codes);
torch::Tensor y = torch::empty_like(codes);
torch::Tensor z = torch::empty_like(codes);
// Call CUDA kernel
hilbert_decode_cuda<<<(codes.size(0) + BLOCK_SIZE - 1) / BLOCK_SIZE, BLOCK_SIZE>>>(
codes.size(0),
reinterpret_cast<uint32_t*>(codes.contiguous().data_ptr<int>()),
reinterpret_cast<uint32_t*>(x.data_ptr<int>()),
reinterpret_cast<uint32_t*>(y.data_ptr<int>()),
reinterpret_cast<uint32_t*>(z.data_ptr<int>())
);
return std::make_tuple(x, y, z);
}
|