Commit Graph

140 Commits

Author SHA1 Message Date
Li-Ta Lo
0e97fcb94f handle the case of 0 or 1 element in the input for ScanExclusiveByKey 2017-04-18 16:12:36 -06:00
Li-Ta Lo
a2085abaf6 add DerivedAlgorithm when calling other methods 2017-04-18 09:39:15 -06:00
Li-Ta Lo
a205f21043 make ScanExclusiveByKey return void, rearrange parameter ordering 2017-04-17 16:11:02 -06:00
Li-Ta Lo
7023266585 add both generic and Thrust ScanExclusiveByKey 2017-04-17 15:03:49 -06:00
Li-Ta Lo
e77f9fac6a add CUDA implementation of ScanInclusiveByKey using Thrust library 2017-04-14 11:25:25 -06:00
Li-Ta Lo
da8a2315ce add wip for exclusive scan by key 2017-04-13 11:50:56 -06:00
Li-Ta Lo
847c73b7d4 first attemp at segmented inclusive scan 2017-03-29 09:29:11 -06:00
Kenneth Moreland
8b33f2bc79 Fix device adapter algorithms for small arrays
There were some issues for device adapter algorithms (like scan and
reduce) for empty arrays or arrays of size 1. This adds tests for these
short arrays to the device adapter algorithm tests and fixes these
problems.
2017-03-08 19:05:28 -05:00
Sujin Philip
9eddce6c99 Rename StreamCompact to CopyIf
Plus, removes the version that uses one array as both input and stencil.
2017-03-06 11:08:27 -05:00
Robert Maynard
a74bab02f4 Merge topic 'allow_reduction_to_different_T_type'
b97b4cc7 Allow thrust::reduce to work when iterator and initial value types differ.
64bcc343 Refactor MinAndMax to use vtkm::Vec<T,2> instead of Pair.
8d60ed57 Refactor MinAndMax to be a shared binary operator.
18375b54 Update Bound computations to always use a single Reduce call
2cfc9743 Reduce can support reduce to a T type that isn't the arrayhandles T type.

Acked-by: Kitware Robot <kwrobot@kitware.com>
Merge-request: !614
2016-12-01 07:56:30 -05:00
Robert Maynard
6b1094c767 Consistently include windows.h by making a wrapper header.
We previously included windows.h in numerous locations using different
techniques to guard against bringing in parts of the file that are bad
(min/max macros, etc). This solves the problem by consistently using
vtkm/internal/Windows.h to setup everything.
2016-11-28 09:54:37 -05:00
Robert Maynard
2cfc9743e3 Reduce can support reduce to a T type that isn't the arrayhandles T type.
This has been done so that operations such as computing the Min/Max of
an array can be done in a single reduce step.
2016-11-25 11:40:46 -05:00
Kenneth Moreland
fdaccc22db Remove exports for header-only functions/methods
Change the VTKM_CONT_EXPORT to VTKM_CONT. (Likewise for EXEC and
EXEC_CONT.) Remove the inline from these macros so that they can be
applied to everything, including implementations in a library.

Because inline is not declared in these modifies, you have to add the
keyword to functions and methods where the implementation is not inlined
in the class.
2016-11-15 22:22:13 -07:00
Christopher Sewell
f779e8a1bd Removing streaming scan inclusive for now 2016-11-09 15:43:57 -07:00
Christopher Sewell
faa69e539d vtkm::internal::Add to vtkm::Add 2016-11-09 14:21:50 -07:00
Christopher Sewell
8a8b409d4c Merge remote-tracking branch 'upstream/master' into StreamingArray 2016-11-09 12:15:58 -07:00
Robert Maynard
f31d6c2258 Refactor vtkm::Types to be concise and move math helpers out of internal.
I have verified that the optimized assembly for Vec<3> and Vec<4> are consistent
with what we generated before.
2016-10-28 14:57:16 -04:00
Christopher Sewell
9afbdffbf5 Attempt fix warning in streaming 2016-10-25 08:36:24 -06:00
Christopher Sewell
a5a487b365 Attempt 15 to resolve Windows compiler warning with streaming storage 2016-10-23 07:24:28 -06:00
Christopher Sewell
b9d8172635 Attempt 14 to resolve Windows compiler warning with streaming storage 2016-10-21 17:31:45 -06:00
Christopher Sewell
93d7956daf Attempt 13 to resolve Windows compiler warning with streaming storage 2016-10-21 16:08:13 -06:00
Christopher Sewell
92d686fccd Attempt 12 to resolve Windows compiler warning with streaming storage 2016-10-21 15:12:22 -06:00
Christopher Sewell
45208e83cc Attempt 9 to resolve Windows compiler warning with streaming storage 2016-10-21 12:41:23 -06:00
Christopher Sewell
f1af30804e Attempt 8 to resolve Windows compiler warning with streaming storage 2016-10-21 10:58:14 -06:00
Christopher Sewell
c540a699f2 Attempt 7 to resolve Windows compiler warning with streaming storage 2016-10-21 10:12:15 -06:00
Christopher Sewell
6096b6f12a Attempt 6 to resolve Windows compiler warning with streaming storage 2016-10-20 15:35:54 -06:00
Christopher Sewell
2c8d2d3006 Attempt 5 to resolve Windows compiler warning with streaming storage 2016-10-20 14:27:07 -06:00
Christopher Sewell
4af0cd4bfb Attempt 4 to resolve Windows compiler warning with streaming storage 2016-10-20 13:00:43 -06:00
Christopher Sewell
05975a2325 Attempt 3 to resolve Windows compiler warning with streaming storage 2016-10-20 10:32:30 -06:00
Christopher Sewell
278db37c08 Attempt 2 to resolve Windows compiler warning with streaming storage 2016-10-19 14:44:48 -06:00
Christopher Sewell
c5bd2a3e80 Attempt 1 to resolve Windows compiler warning with streaming storage 2016-10-19 12:17:01 -06:00
Christopher Sewell
1aa092e027 Removing unnecessary allocation in streaming scan 2016-10-07 18:09:50 -06:00
Christopher Sewell
58952e1465 Adding streaming reduce 2016-09-29 15:06:34 -06:00
Christopher Sewell
bf9e1366f5 Fixing warning about char* 2016-09-29 14:33:21 -06:00
Christopher Sewell
9e2210927e Adding streaming exclusive scan and fixing bugs with streaming inclusive scan, but still need to make it treat input as const 2016-09-28 18:29:04 -06:00
Christopher Sewell
d92f39df12 Merge branch 'master' into StreamingArray 2016-09-15 17:54:59 -06:00
Christopher Sewell
610f96a831 Adding streaming inclusive scan 2016-09-15 17:46:09 -06:00
Robert Maynard
51e50d2933 Add DeviceAdapter::CopySubRange to all device adapters.
This allows callers to copy a subsection of an array into another array,
without clearing the contents of the destination array if a resize
is required.
2016-08-24 15:42:51 -04:00
Robert Maynard
1e19101eee Obey VisualStudio checked iterator levels, and NULL checked iterators
When compiling under VisualStudio we need to first determine if checked
iterators are enabled ( _ITERATOR_DEBUG_LEVEL ). We don't want to use the
NDEBUG key, as we could be inside a project that is in Debug mode with
disabled checked iterators.

Secondly if they are enabled we need to handle the use case of NULL iterators
that get advanced by length zero. This last case is valid, but isn't supported
by the checked iterators so we need to work around it
2016-06-15 13:13:02 -04:00
Kenneth Moreland
0af7188708 Remove warning about loss of precision. 2016-05-03 16:13:09 -06:00
Kenneth Moreland
cc497e6a1b Remove cont/Assert.h and exec/Assert.h
These asserts are consolidated into the unified Assert.h. Also made some
minor edits to add asserts where appropriate and a little bit of
reconfiguring as found.
2016-04-20 15:41:14 -06:00
Robert Maynard
589da1f66a Simplify the General DeviceAdapter Unique method. 2016-03-18 16:39:21 -04:00
Robert Maynard
ee4e490f1d Moved SMP atomic operations from TBB to General.
Even when running with the serial backend, the compiler might enable SIMD
vectorization when optimizations are turned on. When this occurs, we need
to use properly atomic Add's and CAE's.
2016-03-14 10:40:42 -04:00
Matt Larsen
249cce352b Adding type restrictions to serial atomics 2016-03-08 10:39:23 -08:00
Matt Larsen
3b46706e1f Adding compare and swap and removing unsigned atomics 2016-03-08 09:41:02 -08:00
mclarsen
ca37b2e9cb Fixed compiler error with atomics 2016-02-10 08:21:38 -08:00
Matt Larsen
2baac9cd8b initial commit of atomic adds 2016-02-10 07:51:31 -08:00
Robert Maynard
8816642dfd Move algorithms out of DeviceAdapterAlgorithmGeneral to reduce compilation size 2015-10-26 17:22:49 -04:00
Kenneth Moreland
891182ee19 Add ArrayHandleIndex class.
This is the most common use case for ArrayHandleCounting, and this class
is a bit easier to use and a bit faster.
2015-09-14 22:11:09 -06:00
Sujin Philip
514ac54e59 Add custom operator and initial value support to ExclusiveScan 2015-08-28 09:56:04 -04:00
Robert Maynard
bae6ff7f55 Merge branch 'introduce_binary_and_unary_operators' into 'master'
Introduce binary and unary operators

See merge request !94
2015-08-06 15:14:28 -04:00
Robert Maynard
d3fd571ef2 Add vtkm/UnaryPredicates header.
Currently includes the following predicates:
  - IsDefaultConstructor
  - NotDefaultConstructor
  - LogicalNot
2015-07-30 13:12:59 -04:00
Kenneth Moreland
85a2545c1e Fix various MSVC warnings.
Most warnings had to do with implicit type conversions and some on
treating an integer (actually a pointer) as a boolean.
2015-07-27 15:56:04 -06:00
Will Usher
a6078a6413 It looks like the type here should actually be vtkm::Id
Encountered this causing a compilation error when trying to run some
benchmarks with FloatDefault input and values since it was using the
wrong type for some iterators iirc and couldn't cast to the vtkm::Id
type. This should fix that by using the correct type for the output
portal.
2015-07-08 14:01:30 -06:00
Robert Maynard
b18f59f10c Remove unused typedef in DeviceAdapterAlgorithmGeneral. 2015-06-30 08:38:25 -04:00
Robert Maynard
459552d795 Update the name of the binary functor for UpperBounds 2015-06-23 09:19:09 -04:00
Robert Maynard
3cbac03810 Update the name of the binary functor Unique 2015-06-23 08:58:53 -04:00
Robert Maynard
e6cadef457 Update the name of the unary functor for StreamCompact. 2015-06-23 08:56:18 -04:00
Robert Maynard
7ee3084868 Update the name of the binary functor for Sort. 2015-06-23 08:36:42 -04:00
Robert Maynard
604289c1ce Update the name of the binary functor for ScanInclusive 2015-06-23 08:26:04 -04:00
Robert Maynard
8ba0dacc65 Update the name of the binary functor for Reduce. 2015-06-22 11:55:02 -04:00
Robert Maynard
0901e71574 Update the name of the binary comparison for LowerBounds. 2015-06-22 11:49:11 -04:00
Robert Maynard
49467f3c33 Clean up indentation and remove dead code. 2015-06-19 15:30:03 -04:00
Robert Maynard
2ce14c7760 DeviceAdapter's now properly init all value types.
This is needed when use vtkm::Vec, as it doesn't zero the memory it contains.
2015-06-16 08:35:04 -04:00
Chun-Ming Chen
32f3fa6d87 Fix unmerged codes from the previous merge. Use multiply in customed operator ReduceByKey test 2015-06-15 11:12:51 -04:00
Chun-Ming Chen
4d9c25deff Merge remote-tracking branch 'robert/reduce_by_key_custom_binary_function' into ReduceByKey_fix 2015-06-15 10:55:44 -04:00
Chun-Ming Chen
dcc12e733f Fix ReduceByKey general algorithm to work with parallel ScanInclusive. Use bit representation for states. 2015-06-13 02:46:36 -04:00
Robert Maynard
53d7053d7a Fix a bug where DeviceAdapterGeneral::ReduceByKey ignored custom binary functor 2015-06-12 16:11:28 -04:00
Robert Maynard
50ffcec3bb Merge branch 'tbb_SortByKey_ZipArray_fix' into 'master'
TBB sort by key zip array fix

The original codes do not accept when the zip handle array contains two different storage types

See merge request !25
2015-06-12 15:43:09 -04:00
Chun-Ming Chen
360279c29a A more general Copy that allows copying into a diff. array type. Testing codes are added. 2015-06-12 12:42:43 -04:00
Chun-Ming Chen
1088e855c3 Add ReduceByKey support for zip array input. A testing codelet to demo the problem. 2015-06-11 11:29:51 -04:00
Chun-Ming Chen
274ed989c4 change KeyCompare() in DeviceAdapterAlgorithmGeneral from private to protected, to be used in DeviceAdapterAlgorithmSerial 2015-06-08 14:19:37 -04:00
Sujin Philip
08f88b1cb9 Add TBB backend. 2015-06-01 13:57:37 -04:00
Robert Maynard
d54aee7eb5 Merge branch 'fix_typo_in_copyright' 2015-05-21 10:32:08 -04:00
Robert Maynard
6b8e7822be The Copyright statement now has all the periods in the correct location. 2015-05-21 10:30:11 -04:00
Robert Maynard
f2b47ffd4a Add ReduceByKey to the DeviceAdapter. 2015-05-21 08:23:28 -04:00
Robert Maynard
be193542ac Introduce StreamCompact with a custom stencil unary predicate. 2015-05-21 08:23:28 -04:00
Robert Maynard
959d89f77d Correct DeviceAdapterAlgorithmGeneral ScanInclusive with length 1.
Previously we returned a sum of 0 when the length is 1, when really
we need to return the value that is in side the single location of the
array.
2015-05-20 09:30:36 -04:00
Robert Maynard
e38caafe37 Adding Reduce with custom operator to the DeviceAdapterAlgorithm. 2015-05-14 15:16:49 -04:00
Robert Maynard
5d9f369d0c Adding ScanInclusive with custom binary operator to DeviceAdapterAlgorithm. 2015-05-14 15:16:49 -04:00
Robert Maynard
9519737b3a Adding Reduce to the DeviceAdapterAlgorithm. 2015-05-14 15:16:49 -04:00
Robert Maynard
6564d7af1c Enable SortByKey Test on the Device Adapter. 2015-05-06 09:27:59 -04:00
Robert Maynard
d9270e408d Adding a cuda device adapter to vtkm.
Porting the dax device adapter over to vtkm. Unlike the dax version, doesn't
use the thrust::device_vector, but instead uses thrust::system calls so that
we can support multiple thrust based backends.

Also this has Texture Memory support for input array handles. Some more work
will need to be done to ArrayHandle so that everything works when using an
ArrayHandle inplace with texture memory bindings.
2014-12-19 13:47:28 -05:00
Kenneth Moreland
8839e615e6 Remove GetIteratorBegin/End from all ArrayPortal classes.
Replace them (and their use) with the new ArrayPortalToIterators
functionality.
2014-09-08 14:59:11 -06:00
Kenneth Moreland
21823500c3 Change ArrayContainerControl to Storage.
After a talk with Robert Maynard, we decided to change the name
ArrayContainerControl to Storage. There are several reasons for this
change.

1. The name ArrayContainerControl is unwieldy. It is long, hard for
humans to parse, and makes for long lines and wraparound. It is also
hard to distinguish from other names like ArrayHandleFoo and
ArrayExecutionManager.

2. The word container is getting overloaded. For example, there is a
SimplePolymorphicContainer. Container is being used for an object that
literally acts like a container for data. This class really manages
data.

3. The data does not necessarily have to be on the control side.
Implicit containers store the data nowhere. Derivative containers might
have all the real data on the execution side. It is possible in the
future to have storage on the execution environment instead of the
control (think interfacing with a simulator on the GPU).

Storage is not a perfect word (what does implicit storage really mean?),
but its the best English word we came up with.
2014-06-24 09:58:32 -06:00
Kenneth Moreland
ed9cf46a17 Make sure all header wrapper macros start with vtk_m_ rather than vtkm_
We made this change a while ago to help with completion in IDEs.
(Completion was matching a bunch of wrapper macros that were almost
never used anywhere.) Most of the changes are in comments, but there are
a few bad macro definitions.
2014-06-11 10:43:36 -06:00
Kenneth Moreland
b012668345 Add a FunctorBase class for scheduling non-worklets
Whenever creating a functor to be launched in the execution environment
using the device adapter Schedule algorithm, you had to also create a
couple of methods to handle error message buffers. For convenience, lots
of code started to just inherit from WorkletBase. Although this worked,
it was a misnomer (and might cause problems in the future if worklets
later require different things from its base). To get around this
problem, add a FunctorBase class that is intended to be used as the
superclass to functors called with Schedule.
2014-06-10 11:35:13 -06:00
Robert Maynard
2b7a0e0490 revise the header guard naming convention to not conflict with macro names. 2014-03-07 10:22:36 -05:00
Robert Maynard
c80fb9259f Update the initial repository to use the correct indentation style. 2014-02-11 16:20:30 -05:00
Robert Maynard
c2101b8ffc Add in a serial device adapter and required supporting classes.
We now can verify that the array handle is usable by a device adapter.
2014-02-11 12:34:56 -05:00