Commit Graph

4527 Commits

Author SHA1 Message Date
Benoit Steiner
bc68fc2fe7 Enable constant expressions when compiling cuda code with clang. 2016-03-30 17:58:32 -07:00
Benoit Jacob
01b5333e44 bug #1186 - vreinterpretq_u64_f64 fails to build on Android/Aarch64/Clang toolchain 2016-03-30 11:02:33 -04:00
Benoit Steiner
1841d6d4c3 Added missing cuda template specializations for numext::ceil 2016-03-29 13:29:34 -07:00
Benoit Steiner
e02b784ec3 Added support for standard mathematical functions and trancendentals(such as exp, log, abs, ...) on fp16 2016-03-29 09:20:36 -07:00
Benoit Steiner
c38295f0a0 Added support for fmod 2016-03-28 15:53:02 -07:00
Konstantinos Margaritis
01e7298fe6 actually include ZVector files, passes most basic tests (float still fails) 2016-03-28 10:58:02 -04:00
Konstantinos Margaritis
f48011119e Merged eigen/eigen into default 2016-03-28 01:48:45 +03:00
Konstantinos Margaritis
ed6b9d08f1 some primitives ported, but missing intrinsics and crash with asm() are a problem 2016-03-27 18:47:49 -04:00
Benoit Steiner
65716e99a5 Improved the cost estimate of the quotient op 2016-03-25 11:13:53 -07:00
Benoit Steiner
d94f6ba965 Started to model the cost of divisions more accurately. 2016-03-25 11:02:56 -07:00
Benoit Steiner
2e4e4cb74d Use numext::abs instead of abs to avoid incorrect conversion to integer of the argument 2016-03-23 16:57:12 -07:00
Benoit Steiner
81d340984a Removed executable bit from header files 2016-03-23 16:15:02 -07:00
Benoit Steiner
bff8cbad06 Removed executable bit from header files 2016-03-23 16:14:23 -07:00
Benoit Steiner
7a570e50ef Fixed contractions of fp16 2016-03-23 16:00:06 -07:00
Benoit Steiner
fc3660285f Made type conversion explicit 2016-03-23 09:56:50 -07:00
Benoit Steiner
0e68882604 Added the ability to divide a half float by an index 2016-03-23 09:46:42 -07:00
Benoit Steiner
6971146ca9 Added more conversion operators for half floats 2016-03-23 09:44:52 -07:00
Benoit Steiner
f9ad25e4d8 Fixed contractions of 16 bit floats 2016-03-22 09:30:23 -07:00
Benoit Steiner
134d750eab Completed the implementation of vectorized type casting of half floats. 2016-03-18 13:36:28 -07:00
Benoit Steiner
7bd551b3a9 Make all the conversions explicit 2016-03-18 12:20:08 -07:00
Benoit Steiner
7b98de1f15 Implemented some of the missing type casting for half floats 2016-03-17 21:45:45 -07:00
Christoph Hertzberg
46aa9772fc Merged in ebrevdo/eigen (pull request PR-169)
Bugfixes to cuda tests, igamma & igammac implemented, & tests for digamma, igamma, igammac on CPU & GPU.
2016-03-16 21:59:08 +01:00
Eugene Brevdo
1f69a1b65f Change the header guard around certain numext functions to be CUDA specific. 2016-03-16 12:44:35 -07:00
Benoit Steiner
5a51366ea5 Fixed a typo. 2016-03-14 09:25:16 -07:00
Benoit Steiner
fcf59e1c37 Properly gate the use of cuda intrinsics in the code 2016-03-14 09:13:44 -07:00
Benoit Steiner
97a1f1c273 Make sure we only use the half float intrinsic when compiling with a version of CUDA that is recent enough to provide them 2016-03-14 08:37:58 -07:00
Benoit Steiner
e29c9676b1 Don't mark the cast operator as explicit, since this is a c++11 feature that's not supported by older compilers. 2016-03-12 00:15:58 -08:00
Benoit Steiner
eecd914864 Also replaced uint32_t with unsigned int to make the code more portable 2016-03-11 19:34:21 -08:00
Benoit Steiner
1ca8c1ec97 Replaced a couple more uint16_t with unsigned short 2016-03-11 19:28:28 -08:00
Benoit Steiner
0423b66187 Use unsigned short instead of uint16_t since they're more portable 2016-03-11 17:53:41 -08:00
Benoit Steiner
048c4d6efd Made half floats usable on hardware that doesn't support them natively. 2016-03-11 17:21:42 -08:00
Benoit Steiner
456e038a4e Fixed the +=, -=, *= and /= operators to return a reference 2016-03-10 15:17:44 -08:00
Eugene Brevdo
836e92a051 Update MathFunctions/SpecialFunctions with intelligent header guards. 2016-03-09 09:04:45 -08:00
Eugene Brevdo
5e7de771e3 Properly fix merge issues. 2016-03-08 17:35:05 -08:00
Eugene Brevdo
73220d2bb0 Resolve bad merge. 2016-03-08 17:28:21 -08:00
Eugene Brevdo
14f0fde51f Add certain functions to numext (log, exp, tan) because CUDA doesn't support std::
Use these in SpecialFunctions.
2016-03-08 17:17:44 -08:00
Eugene Brevdo
0bb5de05a1 Finishing touches on igamma/igammac for GPU. Tests now pass. 2016-03-07 15:35:09 -08:00
Eugene Brevdo
5707004d6b Fix Eigen's building of sharded tests that use CUDA & more igamma/igammac bugfixes.
0. Prior to this PR, not a single sharded CUDA test was actually being *run*.
Fixed that.

GPU tests are still failing for igamma/igammac.

1. Add calls for igamma/igammac to TensorBase
2. Fix up CUDA-specific calls of igamma/igammac
3. Add unit tests for digamma, igamma, igammac in CUDA.
2016-03-07 14:08:56 -08:00
Benoit Steiner
05bbca079a Turn on some of the cxx11 features when compiling with visual studio 2015 2016-03-05 10:52:08 -08:00
Eugene Brevdo
0b9e0abc96 Make igamma and igammac work correctly.
This required replacing ::abs with std::abs.
Modified some unit tests.
2016-03-04 21:12:10 -08:00
Eugene Brevdo
7ea35bfa1c Initial implementation of igamma and igammac. 2016-03-03 19:39:41 -08:00
Benoit Steiner
1032441c6f Enable partial support for half floats on Kepler GPUs. 2016-03-03 10:34:20 -08:00
Benoit Steiner
1da10a7358 Enable the conversion between floats and half floats on older GPUs that support it. 2016-03-03 10:33:20 -08:00
Benoit Steiner
2de8cc9122 Merged in ebrevdo/eigen (pull request PR-167)
Add infinity() support to numext::numeric_limits, use it in lgamma.

I tested the code on my gtx-titan-black gpu, and it appears to work as expected.
2016-03-03 09:42:12 -08:00
Eugene Brevdo
ab3dc0b0fe Small bugfix to numeric_limits for CUDA. 2016-03-02 21:48:46 -08:00
Eugene Brevdo
6afea46838 Add infinity() support to numext::numeric_limits, use it in lgamma.
This makes the infinity access a __device__ function, removing
nvcc warnings.
2016-03-02 21:35:48 -08:00
Gael Guennebaud
3fccef6f50 bug #537: fix compilation with Apples's compiler 2016-03-02 13:22:46 +01:00
Gael Guennebaud
dfa80b2060 Compilation fix 2016-03-01 12:48:56 +01:00
Gael Guennebaud
bee9efc203 Compilation fix 2016-03-01 12:47:27 +01:00
Gael Guennebaud
e9bea614ec Fix shortcoming in fixed-value deduction of startRow/startCol 2016-02-29 10:31:27 +01:00
Gael Guennebaud
8e6faab51e bug #1172: make valuePtr and innderIndexPtr properly return null for empty matrices. 2016-02-27 14:55:40 +01:00
Gael Guennebaud
91e1375ba9 merge 2016-02-23 11:09:05 +01:00
Gael Guennebaud
055000a424 Fix startRow()/startCol() for dense Block with direct access:
the initial implementation failed for empty rows/columns for which are ambiguous.
2016-02-23 11:07:59 +01:00
Benoit Steiner
6270d851e3 Declare the half float type as arithmetic. 2016-02-22 13:59:33 -08:00
Benoit Steiner
584832cb3c Implemented the ptranspose function on half floats 2016-02-21 12:44:53 -08:00
Benoit Steiner
95fceb6452 Added the ability to compute the absolute value of a half float 2016-02-21 20:24:11 +00:00
Benoit Steiner
9ff269a1d3 Moved some of the fp16 operators outside the Eigen namespace to workaround some nvcc limitations. 2016-02-20 07:47:23 +00:00
Gael Guennebaud
d90a2dac5e merge 2016-02-19 23:01:27 +01:00
Gael Guennebaud
6fa35bbd28 bug #1170: skip calls to memcpy/memmove for empty imput. 2016-02-19 22:58:52 +01:00
Gael Guennebaud
6f0992c05b Fix nesting type and complete reflection methods of Block expressions. 2016-02-19 22:21:02 +01:00
Gael Guennebaud
f3643eec57 Add typedefs for the return type of all block methods. 2016-02-19 22:15:01 +01:00
Benoit Steiner
180156ba1a Added support for tensor reductions on half floats 2016-02-19 10:05:59 -08:00
Benoit Steiner
5c4901b83a Implemented the scalar division of 2 half floats 2016-02-19 10:03:19 -08:00
Benoit Steiner
f7cb755299 Added support for operators +=, -=, *= and /= on CUDA half floats 2016-02-19 15:57:26 +00:00
Benoit Steiner
dc26459b99 Implemented protate() for CUDA 2016-02-19 15:16:54 +00:00
Benoit Steiner
ac5d706a94 Added support for simple coefficient wise tensor expression using half floats on CUDA devices 2016-02-19 08:19:12 +00:00
Benoit Steiner
0606a0a39b FP16 on CUDA are only available starting with cuda 7.5. Disable them when using an older version of CUDA 2016-02-18 23:15:23 -08:00
Benoit Steiner
7151bd8768 Reverted unintended changes introduced by a bad merge 2016-02-19 06:20:50 +00:00
Benoit Steiner
17b9fbed34 Added preliminary support for half floats on CUDA GPU. For now we can simply convert floats into half floats and vice versa 2016-02-19 06:16:07 +00:00
Benoit Steiner
8ce46f9d89 Improved implementation of ptanh for SSE and AVX 2016-02-18 13:24:34 -08:00
Eugene Brevdo
832380c455 Merged eigen/eigen into default 2016-02-17 14:44:06 -08:00
Eugene Brevdo
06a2bc7c9c Tiny bugfix in SpecialFunctions: some compilers don't like doubles
implicitly downcast to floats in an array constructor.
2016-02-17 14:41:59 -08:00
Gael Guennebaud
f6f057bb7d bug #1166: fix shortcomming in gemv when the destination is not a vector at compile-time. 2016-02-15 21:43:07 +01:00
Gael Guennebaud
4252af6897 Remove dead code. 2016-02-12 16:13:35 +01:00
Gael Guennebaud
2f5f56a820 Fix usage of evaluator in sparse * permutation products. 2016-02-12 16:13:16 +01:00
Gael Guennebaud
0a537cb2d8 bug #901: fix triangular-view with unit diagonal of sparse rectangular matrices. 2016-02-12 15:58:31 +01:00
Benoit Steiner
17e93ba148 Pulled latest updates from trunk 2016-02-11 15:05:38 -08:00
Benoit Steiner
3628f7655d Made it possible to run the scalar_binary_pow_op functor on GPU 2016-02-11 15:05:03 -08:00
Hauke Heibel
eeac46f980 bug #774: re-added comment referencing equations in the original paper 2016-02-11 19:38:37 +01:00
Benoit Steiner
c569cfe12a Inline the +=, -=, *= and /= operators consistently between DenseBase.h and SelfCwiseBinaryOp.h 2016-02-11 09:33:32 -08:00
Gael Guennebaud
8cc9232b9a bug #774: fix a numerical issue producing unwanted reflections. 2016-02-11 15:32:56 +01:00
Gael Guennebaud
2d35c0cb5f Merged in rmlarsen/eigen (pull request PR-163)
Implement complete orthogonal decomposition in Eigen.
2016-02-11 15:12:34 +01:00
Benoit Steiner
33e2373f01 Merged in nnyby/eigen/nnyby/doc-grammar-fix-linearly-space-linearly-1443742971203 (pull request PR-138)
[doc] grammar fix: "linearly space" -> "linearly spaced"
2016-02-10 23:29:59 -08:00
Benoit Steiner
6d8b1dce06 Avoid implicit cast from double to float. 2016-02-10 18:07:11 -08:00
Rasmus Munk Larsen
b6fdf7468c Rename inverse -> pseudoInverse. 2016-02-10 13:03:07 -08:00
Benoit Jacob
9d6f1ad398 I'm told to use __EMSCRIPTEN__ by an Emscripten dev. 2016-02-10 12:48:34 -05:00
Benoit Steiner
bfb3fcd94f Optimized implementation of the tanh function for SSE 2016-02-10 08:52:30 -08:00
Benoit Steiner
2d523332b3 Optimized implementation of the hyperbolic tangent function for AVX 2016-02-10 08:48:05 -08:00
Benoit Jacob
e6ee18d6b4 Make the GCC workaround for sqrt GCC-only; detect Emscripten as non-GCC 2016-02-10 11:11:49 -05:00
Benoit Jacob
964a95bf5e Work around Emscripten bug - https://github.com/kripken/emscripten/issues/4088 2016-02-10 10:37:22 -05:00
Benoit Steiner
970751ece3 Disabling the nvcc warnings in addition to the clang warnings when clang is used as a frontend for nvcc 2016-02-09 20:55:50 -08:00
Rasmus Munk Larsen
bb8811c655 Enable inverse() method for computing pseudo-inverse. 2016-02-09 20:35:20 -08:00
Benoit Steiner
5cc0dd5f44 Fixed the code that disables the use of variadic templates when compiling with nvcc on ARM devices. 2016-02-09 10:32:01 -08:00
Benoit Steiner
24d291cf16 Worked around nvcc crash when compiling Eigen on Tegra X1 2016-02-09 02:34:02 +00:00
Rasmus Munk Larsen
53f60e0afc Make applyZAdjointOnTheLeftInPlace protected. 2016-02-08 09:01:43 -08:00
Rasmus Munk Larsen
414efa47d3 Add missing calls to tests of COD.
Fix a few mistakes in 3.2 -> 3.3 port.
2016-02-08 08:50:34 -08:00
Gael Guennebaud
c2bf2f56ef Remove custom unaligned loads for SSE. They were only useful for core2 CPU. 2016-02-08 14:29:12 +01:00
Gael Guennebaud
a4c76f8d34 Improve inlining 2016-02-08 11:33:02 +01:00
Rasmus Munk Larsen
16ec450ca1 Nevermind. 2016-02-06 17:54:01 -08:00
Rasmus Munk Larsen
019fff9a00 Add my name to copyright notice in ColPivHouseholder.h, mostly for previous work on stable norm downdate formula. 2016-02-06 17:48:42 -08:00