 }
 
 inline __host__ __device__ half operator*(half a, half b) {
-  return mul(a, b);
+  return THCNumerics<half>::mul(a, b);
 }
 
 inline __host__ __device__ float operator*(half a, float b) {


#ifdef CUDA_HALF_TENSOR

inline __host__ __device__ half mul(half a, half b) {
  #ifdef __CUDA_ARCH__
  #ifdef CUDA_HALF_INSTRUCTIONS
    return __hmul(a, b);
  #else
    float fa = __half2float(a);
    float fb = __half2float(b);
    return __float2half( fa * fb );
  #endif
  #else // __CUDA_ARCH__
    return THC_float2half(THC_half2float(a) * THC_half2float(b));
  #endif
}

inline __host__ __device__ half div(half a, half b) {
  #ifdef __CUDA_ARCH__
  #ifdef CUDA_HALF_INSTRUCTIONS
    return __hdiv(a, b);
  #else
    float fa = __half2float(a);
    float fb = __half2float(b);
    return __float2half( fa / fb );
  #endif
  #else // __CUDA_ARCH__
    return THC_float2half(THC_half2float(a) / THC_half2float(b));
  #endif
}

// arithmetic functions

inline __host__ __device__ half operator+(half a, half b) {

}

inline __host__ __device__ half operator/(half a, half b) {
  return THCNumerics<half>::div(a, b);
}

inline __host__ __device__ float operator/(float a, half b) {
