Commit Graph

553 Commits

Author SHA1 Message Date
Robert Maynard
54a8a713e7 Merge topic 'bidirectional-array-transform'
af32a21b Add support for bidirectional array transforms.

Acked-by: Kitware Robot <kwrobot@kitware.com>
Merge-request: !214
2015-11-10 15:47:06 -05:00
T.J. Corona
af32a21b2e Add support for bidirectional array transforms.
Array transforms can now be created with an inverse functor, allowing for
casts back into the native array type. As a result, array transforms with
both a functor and inverse functor defined can perform read and write
operations. As an example, ArrayHandleCast now supports this operation. The
original implementation of ArrayHandleCast (i.e. read only) has been renamed
'ArrayHandleCastForInput'.
2015-11-10 15:13:02 -05:00
Robert Maynard
b3687c6f3c Workaround inclusive_scan issues in thrust 1.8.X for complex value types.
The original workaround for inclusive_scan bugs in thrust 1.8 only solved the
issue for basic arithmetic types such as int, float, double. Now we go one
step further and fix the problem for all types.

The solution is to provide a proper implementation of destructive_accumulate_n
and make sure it exists before any includes of thrust occur.
2015-11-09 17:14:30 -05:00
Kenneth Moreland
53d8e98440 Handle unstable sort in test for cell-to-point connectivity.
The parallel implementation in CellSetExplicit that builds cell-to-point
connectivity from point-to-cell connectivity uses a parallel sort-by-
key. The sort-by-key in the device adapter is not guaranteed to be
stable, so values associated with a particular key can be in any order.
The test for the result was expecting the connectivity array to be in a
particular order. Change the test to allow any connectivity ordering
that is still valid.
2015-11-06 14:43:02 -07:00
Patricia Kroll Fasel - 090207
4757c0ae9e Merge branch 'master' of gitlab.kitware.com:Fasel/vtk-m into cell_to_point 2015-11-04 13:28:08 -07:00
Patricia Kroll Fasel - 090207
a5f1f823ae Set default device to cuda in unit test of DataSetExplicit to bypass
compiler errors in CellSetExplicit.
2015-11-04 10:16:44 -07:00
Patricia Kroll Fasel - 090207
480f0bd416 Merge branch 'master' of gitlab.kitware.com:Fasel/vtk-m into cell_to_point 2015-11-03 13:48:23 -07:00
Robert Maynard
1b30d6e6de Update Cuda so that SumExclusiveScan supports fancy iterators. 2015-11-03 13:28:07 -05:00
Robert Maynard
97550d5e2d Update Cuda so that UnaryPredictes work with fancy cuda array handles. 2015-11-03 13:28:07 -05:00
Patricia Kroll Fasel - 090207
02e16e7e25 Merge branch 'master' of gitlab.kitware.com:Fasel/vtk-m into cell_to_point 2015-11-03 11:15:26 -07:00
Sujin Philip
1b8fe17f1b Fix for several warnings 2015-11-03 09:11:38 -05:00
Sujin Philip
fd244c4142 Fix errors and warnings caused by recent changes to device adapter tag logic 2015-11-03 09:11:38 -05:00
Robert Maynard
5469b11266 Merge topic 'remove_cuda_reduce_no_return_warning'
71cf2a7d Fix return statement in DeviceAdapterAlgorithmThrust.

Acked-by: Kitware Robot <kwrobot@kitware.com>
Merge-request: !257
2015-11-03 07:50:57 -05:00
Robert Maynard
abe8b4a37a Merge topic 'missing_cellsetperm_methods'
42643260 Add missing methods to CellSetPermutation.

Acked-by: Kitware Robot <kwrobot@kitware.com>
Merge-request: !258
2015-11-03 07:50:49 -05:00
T.J. Corona
0223f69a9c Adding Clear() method to DataSet. 2015-11-02 16:51:04 -05:00
T.J. Corona
4264326047 Add missing methods to CellSetPermutation. 2015-11-02 16:50:05 -05:00
Robert Maynard
71cf2a7d92 Fix return statement in DeviceAdapterAlgorithmThrust. 2015-11-02 16:46:02 -05:00
Patricia Kroll Fasel - 090207
b47a352fcd Merge branch 'master' of gitlab.kitware.com:Fasel/vtk-m into cell_to_point 2015-11-02 13:00:46 -07:00
Patricia Kroll Fasel - 090207
d167c75596 Merge branch 'master' of gitlab.kitware.com:Fasel/vtk-m into cell_to_point 2015-11-02 11:09:07 -07:00
Patricia Kroll Fasel - 090207
498fc682ca Cell to point compiler errors 2015-11-02 11:05:12 -07:00
Kenneth Moreland
6e5f188b7c Add DeviceAdapterSerial.h to data set permuation test
A recent change to the DeviceAdapter header includes the TBB device if
available instead of the serial device. Thus, DeviceAdapterTagSerial was
not defined automatically in all cases for the build of
UnitTestDataSetPermutation. Add the header for that explicitly.
2015-11-02 10:49:51 -07:00
Robert Maynard
5a4ad3eaf7 Merge topic 'remove_cuda_reduce_no_return_warning'
85d28667 Add a return statement to reduce to stop false positive warnings.

Acked-by: Kitware Robot <kwrobot@kitware.com>
Merge-request: !254
2015-11-02 12:46:06 -05:00
Jeremy Meredith
9b534ff43d Merge branch 'exporter' into 'master'
adding VTK file exporter and test cases

This adds a legacy VTK file exporter which supports unstructured, explicit, and point meshes.  (Single Cell Type cell sets are also supported.)


See merge request !247
2015-10-30 17:39:04 -04:00
Patricia Kroll Fasel - 090207
ed0ecf284d Parallel CellToPoint initial code. 2015-10-30 13:59:36 -06:00
Robert Maynard
85d28667c2 Add a return statement to reduce to stop false positive warnings. 2015-10-30 14:55:13 -04:00
Robert Maynard
52467474aa Merge topic 'DeviceAdapter-preprocessor-logic'
04a48129 Add DeviceAdapter preprocessor logic

Acked-by: Kitware Robot <kwrobot@kitware.com>
Acked-by: Kenneth Moreland <kmorel@sandia.gov>
Merge-request: !168
2015-10-29 16:02:01 -04:00
Robert Maynard
8816642dfd Move algorithms out of DeviceAdapterAlgorithmGeneral to reduce compilation size 2015-10-26 17:22:49 -04:00
Robert Maynard
c78e54facc Move algorithms out of DeviceAdapterAlgorithmTBB to reduce compilation size. 2015-10-26 09:47:13 -04:00
Jeremy Meredith
a78e735b3c adding VTK file exporter and test cases. 2015-10-22 17:18:33 -04:00
Kenneth Moreland
dc11d9a917 Merge branch 'cuda-default-constructors' into 'master'
CUDA default constructors, destructors, and assignment operators

Several classes exclusively work in the control environment. However, CUDA likes to add __device__ to constructors, destructors, and assignment operators it automatically creates. This in turn causes warnings about the __device__ function using host-only classes (like boost::shared_ptr). Solve this problem by adding explicit methods for all of these.


See merge request !245
2015-10-22 15:22:43 -04:00
Kenneth Moreland
fec9262099 Add default constructors/destructors/assignment to Dynamic* classes
The DynamicArrayHandle and DynamicCellSet classes exclusively work in
the control environment. However, CUDA likes to add __device__ to
constructors, destructors, and assignment operators it automatically
adds. This in turn causes warnings about the __device__ function using
host-only classes (like boost::shared_ptr). Solve this problem by adding
explicit methods for all of these.
2015-10-22 09:44:51 -06:00
Kenneth Moreland
c7e9c1b67c Add default constructors/destructors/assignment to CellSet classes
The CellSet classes all exclusively work in the control environment.
However, CUDA likes to add __device__ to constructors, destructors, and
assignment operators it automatically adds. This in turn causes warnings
about the __device__ function using host-only classes (like
boost::shared_ptr). Solve this problem by adding explicit methods for
all of these.
2015-10-21 16:28:18 -06:00
Kenneth Moreland
65c2261892 Add default constructors/destructors/assignment to ArrayHandle classes
The ArrayHandle classes all exclusively work in the control environment.
However, CUDA likes to add __device__ to constructors, destructors, and
assignment operators it automatically adds. This in turn causes warnings
about the __device__ function using host-only classes (like
boost::shared_ptr). Solve this problem by adding explicit methods for
all of these.

Implemented this by wrapping up all these default objects in a macro.
This also solved the problem of other constructors that are necessary
for array handles such as a constructor that takes the base array
handle.
2015-10-21 13:36:27 -06:00
Kenneth Moreland
b861209a22 Fix nvcc warnings on MSVC
There is a strange nvcc warning in CUDA 7.5 that sometimes happens on MSVC
that causes it to emit a warning for an undefined method that is clearly
defined. The CUDA development team is aware of the problem and is going
to fix it, but these changes will work around the problem for now.

Thanks to Tom Fogal from NVIDIA for these fixes.
2015-10-21 08:33:15 -06:00
Kenneth Moreland
7ff62d8d6b Explicitly add destructors and copy constructors to ArrayHandle classes
Under CUDA, the default constructors and destructors created are exported
as __host__ and __device__, which causes problems because they used a boost
pointer that only works on the host. The explicit copy constructors and
destructors do the same thing as the default ones except declared to only
work on the host.
2015-10-21 07:50:52 -06:00
Robert Maynard
8de216c088 Propagate vtkm::Id3 scheduling down to the ThreadIndex classes.
This now allows for even more efficient construction of uniform point
coordinates when running under the 3d scheduler, since we don't need to go
from 3d index to flat index to 3d index, instead we stay in 3d index
2015-10-20 09:29:41 -04:00
Kenneth Moreland
1773210635 Merge branch 'shared-indices-in-fetch' into 'master'
Change Fetches to use ThreadIndices instead of Invocation.

Previously, all Fetch objects received an Invocation object in their
Load and Store methods. The point of this was that it allowed the Fetch
to get data from any of the execution objects. However, every Fetch
either just got data directly from its associated execution object or
else used a secondary execution object (the input domain) to get indices
into their own execution object.

This left two potential areas for improvement. First, pulling data out
of the Invocation object was unnecessarily complicated. It would be much
nicer to get data directly from the associated execution object. Second,
when getting index information from the input domain, it was often the
case that extra computations were necessary (particularly on structured
cell sets). There was no way to share the index information among
Fetches, and therefore the computations were replicated.

This change removes the Invocation from the Fetch Load and Store.
Instead, it passes the associated execution object and a new object type
called the ThreadIndices. The ThreadIndices are customized for the input
domain and therefore have all the information needed for a redirected
lookup. It is also a thread-local object so it can cache computed
indices and save on computation time.

See merge request !233
2015-10-19 17:41:43 -04:00
Robert Maynard
95f0ed5016 CellSetSingleType::Fill now works with custom storage types. 2015-10-14 11:03:24 -04:00
Robert Maynard
9006b97432 Manually do copy_backwards since ICC was replacing it with copy
When ICC replaced copy_backwards with copy it causes a failure when the
input and output arrays are the same.
2015-10-12 15:19:54 -04:00
T.J. Corona
b1665dcb32 Add an array handle for bare cuda device pointers.
Array handles for cuda device pointers have been implemented. The data for
these handles exists solely on the exec side (info such as length can be
queried from the cont side).
2015-10-09 12:41:33 -04:00
Kenneth Moreland
99ce66c6fe Change Fetches to use ThreadIndices instead of Invocation.
Previously, all Fetch objects received an Invocation object in their
Load and Store methods. The point of this was that it allowed the Fetch
to get data from any of the execution objects. However, every Fetch
either just got data directly from its associated execution object or
else used a secondary execution object (the input domain) to get indices
into their own execution object.

This left two potential areas for improvement. First, pulling data out
of the Invocation object was unnecessarily complicated. It would be much
nicer to get data directly from the associated execution object. Second,
when getting index information from the input domain, it was often the
case that extra computations were necessary (particularly on structured
cell sets). There was no way to share the index information among
Fetches, and therefore the computations were replicated.

This change removes the Invocation from the Fetch Load and Store.
Instead, it passes the associated execution object and a new object type
called the ThreadIndices. The ThreadIndices are customized for the input
domain and therefore have all the information needed for a redirected
lookup. It is also a thread-local object so it can cache computed
indices and save on computation time.
2015-10-07 17:01:42 -06:00
Robert Maynard
fbddba74a1 Merge topic 'cellset_permutation'
9a8809f9 Add CellSetPermutation which allows custom iteration over a cell set.
66f6db5a IsWriteableArrayHandle now can tell if an array handle can be written too
20f3fb50 Update VertexClustering to use vtkm::cont::CellSetSingleType.
154896b7 Extend the test for DataSetSingleType.

Acked-by: Kitware Robot <kwrobot@kitware.com>
Merge-request: !228
2015-10-07 16:21:40 -04:00
T.J. Corona
a9425ac1bb In DataSet, add the ability to retrieve a cell set by name. 2015-10-02 08:31:57 -04:00
Robert Maynard
f38673f618 Replace ErrorControlOutOfMemory with ErrorControlBadAllocation. 2015-10-01 14:25:28 -04:00
Robert Maynard
085711e2d2 Merge topic 'bad-allocation-error'
a88f43db Add Bad Allocation error.

Acked-by: Kitware Robot <kwrobot@kitware.com>
Acked-by: Kenneth Moreland <kmorel@sandia.gov>
Merge-request: !213
2015-10-01 14:19:57 -04:00
Robert Maynard
9a8809f933 Add CellSetPermutation which allows custom iteration over a cell set.
When you create a CellSetPermutation you provide an array of the cell ids that
you want to iterate. This allows the user to do custom blanking of a data set,
or to do multi iteration over a set of cells.
2015-10-01 09:23:10 -04:00
Robert Maynard
66f6db5a97 IsWriteableArrayHandle now can tell if an array handle can be written too 2015-10-01 09:23:10 -04:00
Robert Maynard
154896b7bc Extend the test for DataSetSingleType. 2015-10-01 08:50:54 -04:00
Robert Maynard
ac48e13c3a Merge topic 'cleanup_cellset_explicit'
8bc40880 Add a test for CellSetExplicit::GetIndices
fa81d1de CellSetExplicit always calls methods using "this->"
9e496306 Allow incremental construction of CellSetSingleType.
3e307879 Mark CellSetExplicit::Fill as host side only.

Acked-by: Kitware Robot <kwrobot@kitware.com>
Merge-request: !227
2015-09-30 13:48:08 -04:00
Sujin Philip
757891cfcc Merge topic 'fix-ArrayHandleTransform'
f3560116 Fix ArrayHandleTransform to use supplied functor object instead of a default

Acked-by: Kitware Robot <kwrobot@kitware.com>
Tested-by: buildbot <buildbot@kitware.com>
Merge-request: !224
2015-09-30 11:58:55 -04:00