5 files changed, 59 insertions, 0 deletions
diff --git a/lib/THC/generic/THCTensorMath.cu b/lib/THC/generic/THCTensorMath.cu
index a0e550a..9ffc89b 100644
--- a/lib/THC/generic/THCTensorMath.cu
+++ b/lib/THC/generic/THCTensorMath.cu
@@ -36,6 +36,14 @@ THCTensor_(zero)(THCState *state, THCTensor *self_)
 }
 
 THC_API void
+THCTensor_(mean)(THCState *state, THCudaTensor *self, THCudaTensor *src, long dim)
+{
+  THAssert(THCTensor_(checkGPU)(state, 2, self, src));
+  THCudaTensor_sum(state, self, src, dim);
+  THCudaTensor_div(state, self, self, THCudaTensor_size(state, src, dim));
+}
+
+THC_API void
 THCTensor_(zeros)(THCState *state, THCTensor *r_, THLongStorage *size)
 {
   THAssert(THCTensor_(checkGPU)(state, 1, r_));
diff --git a/lib/THC/generic/THCTensorMath.h b/lib/THC/generic/THCTensorMath.h
index 5c9e66d..6b59262 100644
--- a/lib/THC/generic/THCTensorMath.h
+++ b/lib/THC/generic/THCTensorMath.h
@@ -4,6 +4,7 @@
 
 THC_API void THCTensor_(fill)(THCState *state, THCTensor *self, real value);
 THC_API void THCTensor_(zero)(THCState *state, THCTensor *self);
+THC_API void THCTensor_(mean)(THCState *state, THCudaTensor *self, THCudaTensor *src, long dim);
 
 THC_API void THCTensor_(zeros)(THCState *state, THCTensor *r_, THLongStorage *size);
 THC_API void THCTensor_(ones)(THCState *state, THCTensor *r_, THLongStorage *size);
diff --git a/lib/THC/generic/THCTensorMathPointwise.cu b/lib/THC/generic/THCTensorMathPointwise.cu
index c4e763a..7cbd00f 100644
--- a/lib/THC/generic/THCTensorMathPointwise.cu
+++ b/lib/THC/generic/THCTensorMathPointwise.cu
@@ -101,6 +101,23 @@ void THCTensor_(sigmoid)(THCState* state, THCTensor* self_, THCTensor* src) {
   THCudaCheck(cudaGetLastError());
 }
 
+void THCTensor_pow(THCState *state, THCTensor *self_, THCTensor *src, real value) {
+  THAssert(THCTensor_(checkGPU)(state, 2, self_, src));
+  if (self_ == src) {
+    if (!THC_pointwiseApply1(state, self_, TensorPowOp<real>(value))) {
+      THArgCheck(false, 2, CUTORCH_DIM_WARNING);
+    }
+  } else {
+    THCTensor_(resizeAs)(state, self_, src);
+
+    if (!THC_pointwiseApply2(state, self_, src, TensorPowOp<real>(value))) {
+      THArgCheck(false, 2, CUTORCH_DIM_WARNING);
+    }
+  }
+
+  THCudaCheck(cudaGetLastError());
+}
+
 #endif
 
 THC_API void
diff --git a/lib/THC/generic/THCTensorMathReduce.cu b/lib/THC/generic/THCTensorMathReduce.cu
index 4b3d28b..90cc3e3 100644
--- a/lib/THC/generic/THCTensorMathReduce.cu
+++ b/lib/THC/generic/THCTensorMathReduce.cu
@@ -201,4 +201,36 @@ THCTensor_(min)(THCState *state,
     MinValuePair<typename TensorUtils<THCTensor>::DataType, long>());
 }
 
+#if defined(THC_REAL_IS_FLOAT) || defined(THC_REAL_IS_DOUBLE) || defined(THC_REAL_IS_HALF)
+
+THC_API void THTensor_(norm)(THCState *state, THCTensor* self, THCTensor* src, real value, long dimension)
+{
+  THAssert(THCTensor_(checkGPU)(state, 2, self, src));
+  if (value == 0.0) {
+    THC_reduceDim(state, self, src,
+                  TensorNonZeroOp<real>(), thrust::plus<real>(),
+                  0.0, dimension);
+  } else if (value == 1.0) {
+    THC_reduceDim(state, self, src,
+                  TensorNormOp<real, 1>(), thrust::plus<real>(),
+                  0.0, dimension);
+
+  } else if (value == 2.0) {
+    THC_reduceDim(state, self, src,
+                  TensorNormOp<real, 2>(), thrust::plus<real>(),
+                  0.0, dimension);
+    THCTensor_(pow)(state, self, self, 0.5);
+
+  } else {
+    THC_reduceDim(state, self, src,
+                  TensorNormOp<real, -1>(), thrust::plus<real>(),
+                  0.0, dimension);
+    THCTensor_(pow)(state, self, self, 1.0 / value);
+  }
+
+  THCudaCheck(cudaGetLastError());
+}
+
+#endif
+
 #endif
diff --git a/lib/THC/generic/THCTensorMathReduce.h b/lib/THC/generic/THCTensorMathReduce.h
index 699a89d..d25317e 100644
--- a/lib/THC/generic/THCTensorMathReduce.h
+++ b/lib/THC/generic/THCTensorMathReduce.h
@@ -6,6 +6,7 @@
 
 THC_API void THCTensor_(renorm)(THCState *state, THCTensor* self, THCTensor* src, real value, long dimension, real max_norm);
 THC_API void THCTensor_(std)(THCState *state, THCTensor *self, THCTensor *src, long dim, int flag);
+THC_API void THTensor_(norm)(THCState *state, THCTensor* self, THCTensor* src, real value, long dimension);
 
 #endif