add more operators

Tensor-Array · Jan 25, 2024 · 341cb40 · 341cb40
1 parent c6b7cae
commit 341cb40
Show file tree

Hide file tree

Showing 15 changed files with 472 additions and 150 deletions.
diff --git a/.vscode/tasks.json b/.vscode/tasks.json
@@ -1,18 +1,21 @@
 {
     "version": "2.0.0",
-    "options": {
-        "cwd": "${workspaceFolder}/build"
-    },
     "tasks": [
         {
             "label": "cmake",
+            "options": {
+                "cwd": "${workspaceFolder}/build"
+            },
             "command": "cmake",
             "args": [
                 ".."
             ]
         },
         {
             "label": "cpack",
+            "options": {
+                "cwd": "${workspaceFolder}/build"
+            },
             "command": "cpack",
             "args": [
                 "-C",
@@ -21,13 +24,19 @@
         },
         {
             "label": "make-install",
+            "options": {
+                "cwd": "${workspaceFolder}/build"
+            },
             "command": "make",
             "args": [
                 "install"
             ]
         },
         {
             "label": "make-test",
+            "options": {
+                "cwd": "${workspaceFolder}/build"
+            },
             "command": "make",
             "args": [
                 "test"

diff --git a/docs/1-bit/calculating-loss.md b/docs/1-bit/calculating-loss.md
@@ -1,3 +1,6 @@
+Before you read this, you need to read:<br>
+[operator](operator.md)
+
 ## Calculating loss
 
 Normally, you are calculating loss by:

diff --git a/docs/1-bit/derivative.md b/docs/1-bit/derivative.md
@@ -1,9 +1,16 @@
-Before read this, you need to read:<br>
+Before you read this, you need to read:<br>
 [operator](operator.md)
 
 ## Derivative rules
 
-Here are some derivative function that only using one bit and bitwise operator.
+Before we read the derivative rule tables, we need to know:
+- **c** : constant
+- **x** : variable
+- **f**, **g** : function
+- **’** : derivative of
+
+
+Here are some derivative functions that only using one bit and bitwise operators.
 
 | Common Functions | Function | Derivative |
 |:-|:-:|:-:|
@@ -12,12 +19,12 @@ Here are some derivative function that only using one bit and bitwise operator.
 | Square | x & x | x ^ x |
 | NOT | ~ x | 1 |
 
-Here are some derivative rules that only using one bit and bitwise operator.
+Here are some derivative rules that only using one bit and bitwise operators.
 
 | Rules | Function | Derivative |
 |:-|:-:|:-:|
 | AND by constant | c & f(x) | c & f’(x) |
 | AND rule | f(x) & g(x) | f’(x) & g’(x) |
 | XOR rule | f(x) ^ g(x) | (f’(x) & g(x)) ^ (f(x) & g’(x)) |
 | NOT rule | ~ f(x) | f’(x) |
-| Chain Rule | f(g(x))| 	f’(g(x)) ^ g’(x) |
+| Chain rule | f(g(x)) | f’(g(x)) & g’(x) |
diff --git a/docs/1-bit/linear.md b/docs/1-bit/linear.md
@@ -0,0 +1,17 @@
+Before you read this, you need to read:<br>
+[operator](operator.md)
+
+## Linear
+
+In normal, we are calulating the linear by using this:
+
+```
+y = ax + b
+y = (a * x) + b
+```
+
+But in 1-bit, we need faster, so we try to make it to:
+
+```
+y = (a & x) ^ b
+```
diff --git a/docs/README → docs/README.md b/docs/README → docs/README.md
diff --git a/src/binary_tensor/core/CMakeLists.txt b/src/binary_tensor/core/CMakeLists.txt
@@ -10,7 +10,7 @@ set(CMAKE_CUDA_ARCHITECTURES 52 75 89)
 set(CMAKE_CUDA_SEPARABLE_COMPILATION TRUE)
 set(CMAKE_CUDA_FLAGS ${CMAKE_CUDA_FLAGS} "--default-stream per-thread")
 
-file(GLOB BinaryTensor_src "dtype/*.cu" "dtype/*.cc" "*.cc" "*.cu")
+file(GLOB BinaryTensor_src "dtype/*.cu" "*.cc" "*.cu")
 
 # file(MAKE_DIRECTORY "include/tensor_array/core")
 

diff --git a/src/binary_tensor/core/dtype/uint1_t_x8.cc b/src/binary_tensor/core/dtype/uint1_t_x8.cc
diff --git a/src/binary_tensor/core/dtype/uint1_t_x8.cu b/src/binary_tensor/core/dtype/uint1_t_x8.cu
@@ -34,77 +34,79 @@ __host__ __device__ uint1_t_x8::operator unsigned char() const
 
 __device__ uint1_t_x8 atomicAdd(uint1_t_x8*, uint1_t_x8);
 
+__host__ __device__ uint1_t_x8 operator&(const uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return static_cast<const unsigned char&>(a) & static_cast<const unsigned char&>(b);
+}
+
+__host__ __device__ uint1_t_x8 operator|(const uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return static_cast<const unsigned char&>(a) | static_cast<const unsigned char&>(b);
+}
+
+__host__ __device__ uint1_t_x8 operator^(const uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return static_cast<const unsigned char&>(a) ^ static_cast<const unsigned char&>(b);
+}
+
+__host__ __device__ uint1_t_x8 operator~(const uint1_t_x8& values)
+{
+    return ~static_cast<const unsigned char&>(values);
+}
+
+__host__ __device__ uint1_t_x8 operator+(const uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return a ^ b;
+}
+
+__host__ __device__ uint1_t_x8 operator-(const uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return a ^ b;
+}
+
+__host__ __device__ uint1_t_x8 operator*(const uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return a & b;
+}
+
+__host__ __device__ uint1_t_x8 operator/(const uint1_t_x8& a, const uint1_t_x8& b);
+
+__host__ __device__ uint1_t_x8& operator&=(uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return (a = a & b);
+}
+
+__host__ __device__ uint1_t_x8& operator|=(uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return (a = a | b);
+}
+
+__host__ __device__ uint1_t_x8& operator^=(uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return (a = a ^ b);
+}
+
+__host__ __device__ uint1_t_x8& operator+=(uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return (a = a + b);
+}
+
+__host__ __device__ uint1_t_x8& operator-=(uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return (a = a - b);
+}
+
+__host__ __device__ uint1_t_x8& operator*=(uint1_t_x8& a, const uint1_t_x8& b)
+{
+    return (a = a * b);
+}
+
+__host__ __device__ uint1_t_x8& operator/=(uint1_t_x8& a, const uint1_t_x8& b);
+
 namespace binary_tensor
 {
     namespace dtype
     {
-        __host__ __device__ uint1_t_x8 operator&(const uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return static_cast<const unsigned char&>(a) & static_cast<const unsigned char&>(b);
-        }
-
-        __host__ __device__ uint1_t_x8 operator|(const uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return static_cast<const unsigned char&>(a) | static_cast<const unsigned char&>(b);
-        }
-
-        __host__ __device__ uint1_t_x8 operator^(const uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return static_cast<const unsigned char&>(a) ^ static_cast<const unsigned char&>(b);
-        }
-
-        __host__ __device__ uint1_t_x8 operator~(const uint1_t_x8& values)
-        {
-            return ~static_cast<const unsigned char&>(values);
-        }
-
-        __host__ __device__ uint1_t_x8 operator+(const uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return a ^ b;
-        }
-
-        __host__ __device__ uint1_t_x8 operator-(const uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return a ^ b;
-        }
-
-        __host__ __device__ uint1_t_x8 operator*(const uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return a & b;
-        }
-
-        __host__ __device__ uint1_t_x8 operator/(const uint1_t_x8& a, const uint1_t_x8& b);
-
-        __host__ __device__ uint1_t_x8& operator&=(uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return (a = a & b);
-        }
-
-        __host__ __device__ uint1_t_x8& operator|=(uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return (a = a | b);
-        }
-
-        __host__ __device__ uint1_t_x8& operator^=(uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return (a = a ^ b);
-        }
-
-        __host__ __device__ uint1_t_x8& operator+=(uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return (a = a + b);
-        }
-
-        __host__ __device__ uint1_t_x8& operator-=(uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return (a = a - b);
-        }
-
-        __host__ __device__ uint1_t_x8& operator*=(uint1_t_x8& a, const uint1_t_x8& b)
-        {
-            return (a = a * b);
-        }
-
-        __host__ __device__ uint1_t_x8& operator/=(uint1_t_x8& a, const uint1_t_x8& b);
+
     }
 }