Commit Graph

136 Commits

Author SHA1 Message Date
Kenneth Moreland
4e34feecb4 Add ability to specialize worklet for device
This adds an ExecutionSignature tag named Device that passes the
DeviceAdapterTag as an argument to the worklet's operator(). This allows
worklets to specialize their code based on the device.
2019-04-01 10:01:54 -06:00
Allison Vacanti
9014de0ebf Suppress new host/device warnings for cuda 10.1. 2019-03-26 14:53:42 -04:00
Allison Vacanti
d1db4ef8b3 Clarify intent of TypeString and TypeName functions.
TypeName is used for logging, and is now TypeToString.

TypeString is used for serialization, and is now SerializableTypeString.
2019-03-01 11:47:53 -05:00
Kenneth Moreland
191d6e5580 Add Mask capabilities to worklets
Mask objects allow you to specify which output values should be
generated when a worklet is run. That is, the Mask allows you to skip
the invocation of a worklet for any number of outputs.
2019-02-25 08:58:39 -07:00
Robert Maynard
d6f66d17a3 Testing run methods now take argc/argv to init logging/runtime device
`vtkm::cont::testing` now initializes with logging enabled and support
for device being passed on the command line, `vtkm::testing` only
enables logging.
2019-01-17 13:16:27 -06:00
Kenneth Moreland
42f810f70e Remove type lists from ControlSignature arguments for arrays
The typelist arguments for ControlSignature tags are antiquated. Remove
them.
2019-01-11 12:15:16 -07:00
ayenpure
2b4801239e Moving ClipTables
moving ClipTables from vtkm/worklet/internal to vtkm/worklet/clip
2018-12-18 09:57:00 +05:30
ayenpure
1ca02af988 Removing whitespaces 2018-12-07 23:33:08 +05:30
ayenpure
cab80e8e3f Adding Clip files 2018-12-07 20:39:21 +05:30
Kenneth Moreland
bddad9b386 Remove TryExecute from filters
Now that the dispatcher does its own TryExecute, filters do not need to
do that. This change requires all worklets called by filters to be able
to execute without knowing the device a priori.
2018-10-16 15:59:53 -06:00
Allison Vacanti
bd337854ec Initial implementation of general logging.
Addresses #291.
2018-10-02 11:37:55 -04:00
Robert Maynard
17a93921c3 Suppress CUDA 10 warnings about __host__ / __device__ 2018-09-20 14:02:16 -04:00
Kenneth Moreland
d879188de0 Make DispatcherBase invoke using a TryExecute
Rather than force all dispatchers to be templated on a device adapter,
instead use a TryExecute internally within the invoke to select a device
adapter.

Because this removes the need to declare a device when invoking a
worklet, this commit also removes the need to declare a device in
several other areas of the code.
2018-08-29 19:18:54 -07:00
haocheng liu
60363fff07 Provide a better error message when ExecutionSig != ControlSig
By facilitating brigand library, we can throw a meaningful compile time
error if the placeholder in execution signature exceeds the bound in
control signature.
2018-08-20 09:33:04 -04:00
Robert Maynard
64958b014b VTK-m now supports passing pointers when invoking worklets.
The original design of invoke and the transport infrastructure
relied on the implementation behavior of vtkm::cont types
such as ArrayHandle that used an internal shared_ptr to managed
state. This allowed passing by value instead of passing by
non-const ref when needing to transfer information to the device.

As VTK-m adds support for classes that use virtuals the ability
to pass by base pointer type allows for us to invoke worklets
using a base type without the risk of type slicing.

Additional by moving over to a non-const ref Invocation we
can update all transports that have 'output' to now be
by ref and therefore support types that can't be copied while
being 'more' correct.
2018-07-06 14:27:36 -04:00
Allison Vacanti
04d9786688 Suppress NVCC warning 2885.
This warning is emitted when constructing a tuple that contains
an object with `__host__`-only constructors.

warning #2885-D: calling a __host__ function(...) from a __host__ __device__ function(...) is not allowed
2018-06-04 14:41:15 -04:00
Allison Vacanti
93506d25e2 Change function signatures to use 'using' aliases.
Also cleaned up some lingering type typedefs.
2018-05-25 17:18:41 -04:00
Kenneth Moreland
0753131a04 Replace ExecutionObjectFactoryBase with ExecutionObjectBase
While making changes to how execution objects work, we had agreed to
name the base object ExecutionObjectBase instead of its original name of
ExecutionObjectFactoryBase. Somehow that change did not make it through.
2018-05-10 17:53:39 -06:00
Matthew Letter
fce5fea31f Changed how TriangulateTables.h was creating ExecutionObjects
abstracted out execution objects for both tetrahedralize and triangulate and removed the device template requirement from the execution object factory for both classes
2018-05-03 14:59:54 -06:00
Matthew Letter
af250acbfc added PrepareForExecution to compile code base
added prepare for execution function to all the execution object factories that did not have it implemented to compile the code base
2018-05-03 14:45:22 -06:00
Matthew Letter
8f622839b6 added PrepareForExecution() to UnitTestDispatcherBase.cxx
added PrepareForExecution() to UnitTestDispatcherBase.cxx to start resolving all the missing implementations of the new execution object architecture.
2018-05-03 14:45:22 -06:00
Matthew Letter
7e5a55881b changes typechecks for execution objects
In order to make the change from the current way execution obejcts are utilized to the new proposed executionObjectFactory process type checks now has to look for the new execution object factory class to check against.
2018-05-03 14:45:21 -06:00
Kenneth Moreland
edc4c85fd9 Move Scatter from Worklet to Dispatcher
Previously, when a Worklet needed a scatter, the scatter object was
stored in the Worklet object. That was problematic because that means
the Scatter, which is a control object, was shoved into the execution
environment.

To prevent that, move the Scatter into the Dispatcher object. The
worklet still declares a ScatterType alias, but no longer has a
GetScatter method. Instead, the Dispatcher now takes a Scatter object in
its constructor. If using the default scatter (ScatterIdentity), the
default constructor is used. If using another type of Scatter that
requires data to set up its state, then the caller of the worklet needs
to provide that to the dispatcher. For convenience, worklets are
encouraged to have a MakeScatter method to help construct a proper
scatter object.
2018-04-27 00:43:51 -04:00
Robert Maynard
73c05c10c7 Suppressions of __host__ warnings shouldn't trigger internal compile errors 2018-04-12 12:04:53 -04:00
Robert Maynard
5100e559c7 Suppress warnings about calling __host__ functions from CUDA 9.0 2018-04-12 10:31:19 -04:00
Robert Maynard
282c515ea8 Suppress warnings about calling __host__ functions from CUDA 8.0 2018-04-12 10:06:57 -04:00
Robert Maynard
8472b64e60 Suppress more CUDA host function from a host/device function warnings 2018-04-05 10:55:43 -04:00
Robert Maynard
8808b41fbd Merge branch 'master' into vtk-m-cmake_refactor 2018-03-29 22:51:26 -04:00
Robert Maynard
7c54125b66 Switch over from static const to static constexpr where possible. 2018-03-10 11:39:58 -05:00
Robert Maynard
601a839d5d VTK-m uses static const/constexpr when supported ( so not on cuda 7.5 )
These changes now allow VTK-m to compile on CUDA 7.5 by using const arrays,
when compiling with CUDA 8+ support we upgrade to static const arrays, and
lastly when CUDA is disabled we fully elevate to static constexpr.
2018-02-26 16:41:29 -05:00
Robert Maynard
a60e378003 CUDA 7.5 doesn't like suppressing warning codes that don't exist
Newer versions of CUDA have no problem suppressing warning codes that don't
exist.
2018-02-26 16:37:57 -05:00
Robert Maynard
ff9dcc4750 Consistently mark variables as static const across all of vtk-m 2018-02-26 16:37:57 -05:00
Robert Maynard
ee69c7a4b7 Remove VS2013 workarounds from VTK-m. 2018-02-23 15:39:39 -05:00
Robert Maynard
e630ac5aa4 Merge branch 'master' into vtk-m-cmake_refactor 2018-02-23 14:52:00 -05:00
Robert Maynard
182f4707e7 vtkm prefers 'using' over typedef. 2018-02-23 10:47:20 -05:00
Robert Maynard
022c987182 Correct warnings and errors found with MSVC2017+CUDA9 2018-01-31 15:58:45 -05:00
luz.paz
80b11afa24 Misc. typos
Found via `codespell -q 3` via downstream VTK
2018-01-30 06:51:47 -05:00
Robert Maynard
5e0d992057 Suppress host/device warnings when building with cuda 9 2018-01-16 15:42:36 -05:00
Kenneth Moreland
fe9594c1d1 Add hint to read source code for help
In trying to give error diagnostics with template definitions of invalid
types, the user encounters some pretty confusing error messages at
first. There is no way to get the compiler to give exactly the
diagnostics we want in a nice readable error message, so we are putting
some verbose instructions as comments in the code. However, a user might
not know to look at the source code since the error happens deep in an
unfamiliar (and complicated) class. Thus, add (yet another) error at the
front that gives a (hopefully) clear indication to look at the source
code for help in understanding the errors.
2018-01-08 10:53:44 -07:00
Kenneth Moreland
c3889a1a67 Give better error reporting when an Invoke parameter is wrong
When one of the parameters to DispatcherBase::Invoke is incorrect, you
get an error in a strange place. It is deep in a call stack using some
heavily templated types rather than where the Invoke is actually called.

Formerly, the code was structured to give a very obfuscated error
message. Try to make this easier on users by first adding helpful hints
in the code around where the error is to explain why the error occured
and helpful advice on how to track down the problem. Second, reconfigure
the static_assert so that the compiler gives a readable error message.
Third, add some auxiliary error messages that provide additional
information about the types and classes involved in the error.
2018-01-08 10:53:43 -07:00
Matthew Letter
c3737c728e Merge branch 'master' into increase-worklet-arguments 2018-01-04 12:07:38 -07:00
Matthew Letter
d9c51d650d increased the number of arguments to worklets
increase the number of arguments to worklets that we support to 10 from 20.
2018-01-04 12:06:16 -07:00
Robert Maynard
93bc0198fe Suppress false positive warnings about calling host device functions. 2018-01-02 10:40:49 -05:00
Robert Maynard
5ca31f7072 Merge topic 'dispatcher_base_proper_type_for_sizeof'
d51ddda9 Since sizeof returns std::size_t we should iterate using unsigned integers

Acked-by: Kitware Robot <kwrobot@kitware.com>
Merge-request: !1016
2017-12-07 08:12:28 -05:00
Robert Maynard
d51ddda9b6 Since sizeof returns std::size_t we should iterate using unsigned integers 2017-12-06 13:57:05 -05:00
Sujin Philip
e28309f09b Remove VTKM_EXEC_CONSTANT
If a global static array is declared with VTKM_EXEC_CONSTANT and the code
is compiled by nvcc (for multibackend code) then the array is only accesible
on the GPU. If for some reason a worklet fails on the cuda backend and it is
re-executed on any of the CPU backends, it will continue to fail.

We couldn't find a simple way to declare the array once and have it available
on both CPU and GPU. The approach we are using here is to declare the arrays
as static inside some "Get" function which is marked as VTKM_EXEC_CONT.
2017-12-05 13:49:55 -05:00
Robert Maynard
f6e18ac452 Remove IntegerSequence.h as we don't need it in vtk-m anymore 2017-11-29 10:55:04 -05:00
Robert Maynard
7f76220427 Redesign the Dispatcher to not need FunctionInterface to convert dynamic types 2017-11-28 11:01:02 -05:00
Kenneth Moreland
c3a3184d51 Update copyright for Sandia
Sandia National Laboratories recently changed management from the
Sandia Corporation to the National Technology & Engineering Solutions
of Sandia, LLC (NTESS). The copyright statements need to be updated
accordingly.
2017-09-20 15:33:44 -06:00
Li-Ta Lo
1d4b43b540 Add documentation
Add doxygen documentations for public interface of KD-Tree construction and search.
Also fixed several const-correctness issues.
2017-09-18 13:16:37 -06:00