diff --git a/build/cpu.o b/build/cpu.o index 9f7b579..7a91d37 100644 Binary files a/build/cpu.o and b/build/cpu.o differ diff --git a/build/libtensor.so b/build/libtensor.so index e6e11d3..5722336 100755 Binary files a/build/libtensor.so and b/build/libtensor.so differ diff --git a/build/tensor.o b/build/tensor.o index a4e4a41..fa22bdf 100644 Binary files a/build/tensor.o and b/build/tensor.o differ diff --git a/norch/__pycache__/tensor.cpython-38.pyc b/norch/__pycache__/tensor.cpython-38.pyc index f61a489..6c43e0a 100644 Binary files a/norch/__pycache__/tensor.cpython-38.pyc and b/norch/__pycache__/tensor.cpython-38.pyc differ diff --git a/norch/autograd/__pycache__/functions.cpython-38.pyc b/norch/autograd/__pycache__/functions.cpython-38.pyc index 033c83c..d4ada96 100644 Binary files a/norch/autograd/__pycache__/functions.cpython-38.pyc and b/norch/autograd/__pycache__/functions.cpython-38.pyc differ diff --git a/norch/autograd/functions.py b/norch/autograd/functions.py index bd52020..7ba219f 100644 --- a/norch/autograd/functions.py +++ b/norch/autograd/functions.py @@ -33,7 +33,7 @@ class MatmulBackward: def backward(self, gradient): x, y = self.input - return [gradient @ y.T, x.T @ gradient] + return [gradient @ y.transpose(-1,-2), x.transpose(-1,-2) @ gradient] class PowBackward: def __init__(self, x, power): diff --git a/norch/csrc/cpu.cpp b/norch/csrc/cpu.cpp index f0ad525..c45a8e9 100644 --- a/norch/csrc/cpu.cpp +++ b/norch/csrc/cpu.cpp @@ -116,7 +116,14 @@ void zeros_like_tensor_cpu(Tensor* tensor, float* result_data) { } } -/*void transpose_tensor_cpu(Tensor* tensor, float* result_data) { +void transpose_1D_tensor_cpu(Tensor* tensor, float* result_data) { + + for (int i = 0; i < tensor->shape[0]; i++) { + result_data[i] = tensor->data[i]; + } +} + +void transpose_2D_tensor_cpu(Tensor* tensor, float* result_data) { int rows = tensor->shape[0]; int cols = tensor->shape[1]; @@ -125,33 +132,17 @@ void zeros_like_tensor_cpu(Tensor* tensor, float* result_data) { result_data[j * rows + i] = tensor->data[i * cols + j]; } } -}*/ +} +void transpose_3D_tensor_cpu(Tensor* tensor, float* result_data) { + int depth = tensor->shape[0]; + int rows = tensor->shape[1]; + int cols = tensor->shape[2]; -void transpose_tensor_cpu(Tensor* tensor, float* result_data) { - int* shape = tensor->shape; - int ndim = tensor->ndim; - int* strides = (int*)malloc(ndim * sizeof(int)); - int* indices = (int*)calloc(ndim, sizeof(int)); - - strides[ndim - 1] = 1; - for (int i = ndim - 2; i >= 0; i--) { - strides[i] = strides[i + 1] * shape[i + 1]; - } - - int idx_result; - for (int idx_source = 0; idx_source < tensor->size; idx_source++) { - idx_result = 0; - for (int dim = 0; dim < ndim; dim++) { - idx_result += indices[dim] * strides[dim]; - } - result_data[idx_result] = tensor->data[idx_source]; - - indices[ndim - 1]++; - for (int dim = ndim - 1; dim > 0; dim--) { - if (indices[dim] == shape[dim]) { - indices[dim] = 0; - indices[dim - 1]++; + for (int i = 0; i < depth; i++) { + for (int j = 0; j < rows; j++) { + for (int k = 0; k < cols; k++) { + result_data[k * rows * depth + j * depth + i] = tensor->data[i * rows * cols + j * cols + k]; } } } diff --git a/norch/csrc/cpu.h b/norch/csrc/cpu.h index ae3eae5..a3c5879 100644 --- a/norch/csrc/cpu.h +++ b/norch/csrc/cpu.h @@ -14,7 +14,10 @@ void pow_tensor_cpu(Tensor* tensor, float power, float* result_data); void scalar_mul_tensor_cpu(Tensor* tensor, float scalar, float* result_data); void ones_like_tensor_cpu(Tensor* tensor, float* result_data); void zeros_like_tensor_cpu(Tensor* tensor, float* result_data); -void transpose_tensor_cpu(Tensor* tensor, float* result_data); +void transpose_1D_tensor_cpu(Tensor* tensor, float* result_data); +void transpose_2D_tensor_cpu(Tensor* tensor, float* result_data); +void transpose_3D_tensor_cpu(Tensor* tensor, float* result_data); +void transpose_axes_cpu(Tensor* tensor, float* result_data, int axis1, int axis2, int* new_shape); void assign_tensor_cpu(Tensor* tensor, float* result_data); #endif /* CPU_H */ diff --git a/norch/csrc/tensor.cpp b/norch/csrc/tensor.cpp index f85759e..03cc657 100644 --- a/norch/csrc/tensor.cpp +++ b/norch/csrc/tensor.cpp @@ -60,6 +60,15 @@ extern "C" { } printf("]\n"); + printf("Strides: ["); + for (int i = 0; i < ndim; i++) { + printf("%d", tensor->strides[i]); + if (i < ndim - 1) { + printf(", "); + } + } + printf("]\n"); + /*printf("Data:\n["); for (int i = 0; i < stride; i++) { printf("%.2f", tensor->data[i]); @@ -715,8 +724,112 @@ extern "C" { fprintf(stderr, "Memory allocation failed\n"); exit(1); } - transpose_tensor_cpu(tensor, result_data); + switch (ndim) { + case 1: + transpose_1D_tensor_cpu(tensor, result_data); + break; + case 2: + transpose_2D_tensor_cpu(tensor, result_data); + break; + case 3: + transpose_3D_tensor_cpu(tensor, result_data); + break; + default: + fprintf(stderr, "Transpose only supports tensors up to 3 dimensions.\n"); + exit(-1); + } return create_tensor(result_data, shape, ndim, device); } } -} + + Tensor* transpose_axes_tensor(Tensor* tensor, int axis1, int axis2) { + char* device = (char*)malloc(strlen(tensor->device) + 1); + if (device != NULL) { + strcpy(device, tensor->device); + } else { + fprintf(stderr, "Memory allocation failed\n"); + exit(-1); + } + + int ndim = tensor->ndim; + int* shape = (int*)malloc(ndim * sizeof(int)); + if (shape == NULL) { + fprintf(stderr, "Memory allocation failed\n"); + exit(-1); + } + + for (int i = 0; i < ndim; i++) { + shape[i] = tensor->shape[i]; + } + + shape[axis1] = tensor->shape[axis2]; + shape[axis2] = tensor->shape[axis1]; + + int size = tensor->size; + + if (strcmp(tensor->device, "cuda") == 0) { + + float* result_data; + cudaMalloc((void **)&result_data, size * sizeof(float)); + //transpose_axes_cuda(tensor, result_data); + return create_tensor(result_data, shape, ndim, device); + } + else { + float* result_data = (float*)malloc(size * sizeof(float)); + if (result_data == NULL) { + fprintf(stderr, "Memory allocation failed\n"); + exit(1); + } + //transpose_axes_cpu(tensor, result_data, axis1, axis2, shape); + assign_tensor_cpu(tensor, result_data); + + Tensor* new_tensor = create_tensor(result_data, shape, ndim, device); + for (int i = 0; i < ndim; i++) { + new_tensor->strides[i] = tensor->strides[i]; + } + new_tensor->strides[axis1] = tensor->strides[axis2]; + new_tensor->strides[axis2] = tensor->strides[axis1]; + make_contiguous(new_tensor); + return new_tensor; + } + } + + void make_contiguous(Tensor* tensor) { + float* new_data = (float*)malloc(tensor->size * sizeof(float)); + if (new_data == NULL) { + // Handle memory allocation failure + return; + } + + int* new_strides = (int*)malloc(tensor->ndim * sizeof(int)); + if (new_strides == NULL) { + free(new_data); + // Handle memory allocation failure + return; + } + + // Calculate new strides assuming C-contiguous order + int stride = 1; + for (int i = tensor->ndim - 1; i >= 0; i--) { + new_strides[i] = stride; + stride *= tensor->shape[i]; + } + + // Rearrange data + for (int i = 0; i < tensor->size; i++) { + int index = 0; + int offset = i; + for (int j = 0; j < tensor->ndim; j++) { + index += (offset / new_strides[j]) * tensor->strides[j]; + offset %= new_strides[j]; + } + new_data[i] = tensor->data[index]; + } + + // Free old data and update tensor properties + free(tensor->data); + free(tensor->strides); + tensor->data = new_data; + tensor->strides = new_strides; + } + } diff --git a/norch/csrc/tensor.h b/norch/csrc/tensor.h index aa8d3e9..dab0b23 100644 --- a/norch/csrc/tensor.h +++ b/norch/csrc/tensor.h @@ -26,6 +26,9 @@ extern "C" { void to_device(Tensor* tensor, char* device); Tensor* ones_like_tensor(Tensor* tensor); Tensor* zeros_like_tensor(Tensor* tensor); + Tensor* transpose_tensor(Tensor* tensor); + Tensor* transpose_axes_tensor(Tensor* tensor, int axis1, int axis2); + void make_contiguous(Tensor* tensor); } #endif /* TENSOR_H */ diff --git a/norch/tensor.py b/norch/tensor.py index d22aa91..2b637c0 100644 --- a/norch/tensor.py +++ b/norch/tensor.py @@ -383,6 +383,28 @@ class Tensor: return result_data + def transpose(self, axis1, axis2): + if axis1 < 0: + axis1 = self.ndim + axis1 + if axis2 < 0: + axis2 = self.ndim + axis2 + + Tensor._C.transpose_axes_tensor.argtypes = [ctypes.POINTER(CTensor), ctypes.c_int, ctypes.c_int] + Tensor._C.transpose_axes_tensor.restype = ctypes.POINTER(CTensor) + + result_tensor_ptr = Tensor._C.transpose_axes_tensor(self.tensor, axis1, axis2) + + result_data = Tensor() + result_data.tensor = result_tensor_ptr + result_data.shape = self.shape.copy() + result_data.shape[axis1] = self.shape[axis2] + result_data.shape[axis2] = self.shape[axis1] + result_data.ndim = self.ndim + result_data.device = self.device + + return result_data + + @property def T(self): Tensor._C.transpose_tensor.argtypes = [ctypes.POINTER(CTensor)] diff --git a/test.py b/test.py index e86e5c2..22a0621 100644 --- a/test.py +++ b/test.py @@ -28,22 +28,72 @@ if __name__ == "__main__": [[7.890, 8.901], [9.012, 1.234], [2.345, 3.456]] ], requires_grad=True) - b = norch.Tensor([ + b = norch.Tensor([[ [1.234, 2.123, 1.5], [5.678, 6.789, 1.293], [3.635, 4.456, 1.0202], [7.890, 8.901, 1.91], - ]) + ],[ + [1.234, 2.123, 1.5], + [5.678, 6.789, 1.293], + [3.635, 4.456, 1.0202], + [7.890, 8.901, 1.91], + ],[ + [1.234, 2.123, 1.5], + [5.678, 6.789, 1.293], + [3.635, 4.456, 1.0202], + [7.890, 8.901, 1.91], + ],[ + [1.234, 2.123, 1.5], + [5.678, 6.789, 1.293], + [3.635, 4.456, 1.0202], + [7.890, 8.901, 1.91], + ],[ + [1.234, 2.123, 1.5], + [5.678, 6.789, 1.293], + [3.635, 4.456, 1.0202], + [7.890, 8.901, 1.91], + ]]) + #print(a.T) + #print(a.T.shape) + #[5, 3, 2] [4, 3] [5, 4, 2] + #print(a.shape, b.shape) #b = norch.Tensor([ # [1.234, 2.123, 1.5]]) + #result = b @ a - #print(a.shape) - c = a.reshape([2,3,5]) - print(b @ c) + + #### testar transpose axes!!!! make it contiguous + + tensor1 = norch.Tensor([[[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]]]) + + # Reshape tensor1 to 2x3x5 + reshaped_tensor = tensor1.transpose(1, 0) + + # Create a 5x4 tensor + tensor2 = norch.Tensor([[1, 2, 3], + [5, 6, 7], + [9, 10, 11], + [13, 14, 15], + [17, 18, 19]]) + + tensor2 = tensor2.transpose(1,0) + + + # Multiply reshaped_tensor by tensor2 + result = tensor2 @ reshaped_tensor + print(result) + + #print(a.shape, b.shape, result.shape) #c = result.sum() #c.backward() #print(a.grad) + #print(a.transpose(2,1)) #a = norch.Tensor([[1, 2, 3], [1, 2, 3], [1, 2, 3]])#.to("cuda") #b = Tensor([[1, 2, 3], [1, 2, 3], [1, 2, 3]])#.to("cuda")