From 98ad84c659d8f26b5bf84eb1f97dcda808f913e4 Mon Sep 17 00:00:00 2001
From: LordMathis <matus@namesny.com>
Date: Tue, 5 Mar 2024 17:38:46 +0100
Subject: [PATCH] Add matrix math kernels

---
 include/kernels/matrix_math.cuh | 19 ++++++++++++++++++
 src/kernels/matrix_math.cu      | 34 +++++++++++++++++++++++++++++++++
 2 files changed, 53 insertions(+)
 create mode 100644 include/kernels/matrix_math.cuh
 create mode 100644 src/kernels/matrix_math.cu

diff --git a/include/kernels/matrix_math.cuh b/include/kernels/matrix_math.cuh
new file mode 100644
index 0000000..20e0956
--- /dev/null
+++ b/include/kernels/matrix_math.cuh
@@ -0,0 +1,19 @@
+#ifndef MATRIX_MATH_H
+#define MATRIX_MATH_H
+
+__global__ void mat_vec_mul_kernel(
+    const float* d_matrix,
+    const float* d_vector,
+    float*       d_output,
+    int          w,
+    int          h
+);
+
+__global__ void vec_vec_add_kernel(
+    const float* d_vector1,
+    const float* d_vector2,
+    float*       d_output,
+    int          w
+);
+
+#endif  // MATRIX_MATH_H
\ No newline at end of file
diff --git a/src/kernels/matrix_math.cu b/src/kernels/matrix_math.cu
new file mode 100644
index 0000000..0c2f48f
--- /dev/null
+++ b/src/kernels/matrix_math.cu
@@ -0,0 +1,34 @@
+#include "matrix_math.cuh"
+
+__global__ void mat_vec_mul_kernel(
+    const float* d_matrix,
+    const float* d_vector,
+    float*       d_output,
+    int          w,
+    int          h
+) {
+    
+    int tid = blockDim.x * blockIdx.x + threadIdx.x;
+
+    if (tid >= w * h) {
+        return;
+    }
+
+    for (int i = 0; i < w; i++) {
+        d_output[tid] += d_matrix[tid * w + i] * d_vector[i];
+    }
+
+}
+
+__global__ void vec_vec_add_kernel(
+    const float* d_vector1,
+    const float* d_vector2,
+    float*       d_output,
+    int          w
+) {
+    int tid = blockDim.x * blockIdx.x + threadIdx.x;
+    if (tid >= w) {
+        return;
+    }
+    d_output[tid] = d_vector1[tid] + d_vector2[tid];
+}