Commit Graph

124 Commits

Author SHA1 Message Date
Robert Maynard
fbee20fc03 Mark the local value for the result of GetNumberOfValues as const.
We aren't modifying the value, so no reason to not have it as const.
2015-05-14 15:16:48 -04:00
Robert Maynard
59618b3a50 Merge branch 'sort_by_key' 2015-05-11 08:04:09 -04:00
Kenneth Moreland
66aa10aab7 Fix issues with warnings about data loss from type conversion
Fix compile warnings that come up with the flags

  -Wconversion -Wno-sign-conversion

This catches several instances (mostly in the testing framework) where
types are implicitly converted. I expect these changes to fix some of
the warnings we are seeing in MSVC.

I was going to add these flags to the list of extra warning flags, but
unfortunately the Thrust library has several warnings of these types,
and I don't know a good way to turn on the warnings for our code but
turn them off for Thrust.
2015-05-07 16:34:06 -06:00
Robert Maynard
6564d7af1c Enable SortByKey Test on the Device Adapter. 2015-05-06 09:27:59 -04:00
Robert Maynard
c2f2c166c0 Test all Fancy ArrayHandles and Array Portals on every device adapter.
Fancy array handles being Counting, Implicit, Permutation, Transform, and Zip.
2015-05-05 14:25:48 -04:00
Kenneth Moreland
d15074153d Remove comments about textures from ArrayManagerExecutionThrustDevice
Per a discussion with Robert Maynard, we no longer have plans to bind
textures or use texture objects. Instead, we are now using the __ldg
command to load from global memory inside the execution portal, which
gives us most of the benefits of textures, but doesn't incur any
bookkeeping.
2015-05-04 14:18:00 -06:00
Kenneth Moreland
e14fc427f0 Change storage references to storage pointers.
Previously ArrayTransfer and ArrayManagerExecution received a reference
to a Storage class in their constructor and held the reference as an
ivar. In retrospect, this is just asking for trouble. First, it is way
too easy to pass by value when you mean to pass by reference. Second, if
there ever is a bug where the Storage goes out of scope before the
classes holding a reference, it is that much harder to debug.
2015-04-30 21:29:58 -06:00
Kenneth Moreland
ec0adf8b16 Change interface of ArrayTransfer to be more like ArrayHandle.
This includes changing methods like LoadDataForInput to PrepareForInput.
It also changed the interface a bit to save a reference to the storage
object. (Maybe it would be better to save a pointer?) These changes also
extend up to the ArrayManagerExecution class, so it can effect device
adapter implementations.
2015-04-30 21:07:36 -06:00
Kenneth Moreland
c5b831b726 Remove LoadDataForInput that takes portals.
This API change effects both ArrayTransfer and ArrayManagerExecution.
This is in preparation for a future change to make the API more
consistent with ArrayHandle.
2015-04-28 11:30:30 -04:00
Robert Maynard
30c5ee9cf4 Remove SetThrustForCuda as we explicitly use the thrust cuda system.
The only reason SetThrustForCuda was ever needed was that we used
::thrust::device_vector instead of ::thrust::system::cuda::vector.
2015-04-23 09:23:41 -04:00
Robert Maynard
c806348c94 Remove debug code from cuda schedule algorithm. 2015-04-16 08:56:08 -04:00
Robert Maynard
3a10a63e8c make sure the 3d block size makes sense. 2015-04-08 15:57:35 -04:00
Robert Maynard
fdac208acb use cuda scheduling versus thrust scheduling. 2015-04-03 10:18:05 -04:00
Robert Maynard
d340a573d9 don't clear ::thrust::vectors as that is not our intended behavior. 2015-03-23 09:01:42 -04:00
Robert Maynard
86dc8f1d38 Move back to thrust::cuda::vector to properly handle allocating uint8's
Our approach of using the underlying allocator inside thrust was a bad approach,
for some reason it fails to properly allocate uint8's or int8's on the correct
boundaries. I expect that this logic is somewhere else in the code and
instead we should use thrust::system::cuda::vector which does this properly.
2015-03-10 10:10:03 -04:00
Robert Maynard
63b1f03187 Simplify the implementation of loading through textures.
We don't need this super complicated system for texture loading.
2015-03-09 16:37:45 -04:00
Robert Maynard
9b49973621 Use __ldg instead of texture object. 2015-03-05 18:31:44 -05:00
Robert Maynard
3c8ce36666 Properly deallocate cuda memory when we are done with it. 2015-01-28 15:54:45 -05:00
Robert Maynard
46df484ca7 Redesign the way we implement when to choose texture portals.
The logic to select texture portals is easier to understand now.
2015-01-20 13:58:28 -05:00
Robert Maynard
e26040282b Properly map TexturePortals to Iterators.
Also at the same time disable the Texture support by default.
2015-01-20 13:58:28 -05:00
Robert Maynard
04bc41cad3 Fix multiple issue in the cuda array handle unit tests.
The namespaces need to be different for each test, or else only the first
implementation of the function will be used for all tests that call that
function.

Also updated the test to verify that we can count starting from a non zero
number.
2015-01-20 09:03:24 -05:00
Robert Maynard
1b5c5a6ce5 Add in initial support for texture binding of input arrays. 2014-12-19 13:47:28 -05:00
Robert Maynard
a509dae909 Use pinned memory for error reporting on the cuda backend.
This reduces the amount of explicit cuda Device to Host memory copies.
2014-12-19 13:47:28 -05: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