Commit 07970cf4 authored by Robert Maynard's avatar Robert Maynard
Browse files

Correct all signed / unsigned and narrowing warnings ( 64bit to 32bit ).

parent 100b894c
......@@ -122,7 +122,8 @@ public:
if (this->DeallocateOnRelease)
{
AllocatorType allocator;
allocator.deallocate(this->Array, this->AllocatedSize);
allocator.deallocate(this->Array,
static_cast<std::size_t>(this->AllocatedSize) );
}
this->Array = NULL;
this->NumberOfValues = 0;
......@@ -150,7 +151,8 @@ public:
if (numberOfValues > 0)
{
AllocatorType allocator;
this->Array = allocator.allocate(numberOfValues);
this->Array = allocator.allocate(
static_cast<std::size_t>(numberOfValues) );
this->AllocatedSize = numberOfValues;
this->NumberOfValues = numberOfValues;
}
......
......@@ -26,26 +26,21 @@
#include <iostream>
// Disable GCC warnings we check vtkmfor but Thrust does not.
#if defined(__GNUC__) && !defined(VTKM_CUDA)
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 6)
// Disable warnings we check vtkm for but Thrust does not.
#if defined(__GNUC__) || defined(____clang__)
#pragma GCC diagnostic push
#endif // gcc version >= 4.6
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 2)
#pragma GCC diagnostic ignored "-Wshadow"
#pragma GCC diagnostic ignored "-Wunused-parameter"
#endif // gcc version >= 4.2
#endif // gcc && !CUDA
#pragma GCC diagnostic ignored "-Wconversion"
#endif // gcc || clang
#include <thrust/system/cuda/vector.h>
#include <thrust/device_malloc_allocator.h>
#include <thrust/copy.h>
#if defined(__GNUC__) && !defined(VTKM_CUDA)
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 6)
#if defined(__GNUC__) || defined(____clang__)
#pragma GCC diagnostic pop
#endif // gcc version >= 4.6
#endif // gcc && !CUDA
#endif // gcc || clang
#include <vtkm/exec/cuda/internal/ArrayPortalFromThrust.h>
......@@ -81,6 +76,7 @@ class ArrayManagerExecutionThrustDevice
{
public:
typedef T ValueType;
typedef typename thrust::system::cuda::pointer<T>::difference_type difference_type;
typedef vtkm::cont::internal::Storage<ValueType, StorageTag> StorageType;
......@@ -104,7 +100,7 @@ public:
///
VTKM_CONT_EXPORT
vtkm::Id GetNumberOfValues() const {
return this->Array.size();
return static_cast<vtkm::Id>(this->Array.size());
}
/// Allocates the appropriate size of the array and copies the given data
......@@ -122,8 +118,9 @@ public:
// The data in this->Array should already be valid.
}
return PortalConstType(this->Array.data(),
this->Array.data() + this->Array.size());
this->Array.data() + static_cast<difference_type>(this->Array.size()));
}
/// Allocates the appropriate size of the array and copies the given data
......@@ -142,7 +139,7 @@ public:
}
return PortalType(this->Array.data(),
this->Array.data() + this->Array.size());
this->Array.data() + static_cast<difference_type>(this->Array.size()));
}
/// Allocates the array to the given size.
......@@ -159,7 +156,7 @@ public:
try
{
this->Array.resize(numberOfValues);
this->Array.resize(static_cast<vtkm::UInt64>(numberOfValues));
}
catch (std::bad_alloc error)
{
......@@ -167,7 +164,7 @@ public:
}
return PortalType(this->Array.data(),
this->Array.data() + this->Array.size());
this->Array.data() + static_cast<difference_type>(this->Array.size()));
}
/// Allocates enough space in \c storage and copies the data in the
......@@ -176,10 +173,10 @@ public:
VTKM_CONT_EXPORT
void RetrieveOutputData(StorageType *storage) const
{
storage->Allocate(this->Array.size());
storage->Allocate(static_cast<vtkm::Id>(this->Array.size()));
::thrust::copy(
this->Array.data(),
this->Array.data() + this->Array.size(),
this->Array.data() + static_cast<difference_type>(this->Array.size()),
vtkm::cont::ArrayPortalToIteratorBegin(storage->GetPortal()));
}
......@@ -189,9 +186,9 @@ public:
{
// The operation will succeed even if this assertion fails, but this
// is still supposed to be a precondition to Shrink.
VTKM_ASSERT_CONT(numberOfValues <= this->Array.size());
VTKM_ASSERT_CONT(numberOfValues <= static_cast<vtkm::Id>(this->Array.size()));
this->Array.resize(numberOfValues);
this->Array.resize(static_cast<vtkm::UInt64>(numberOfValues));
}
......
......@@ -30,16 +30,13 @@
#include <vtkm/exec/internal/ErrorMessageBuffer.h>
#include <vtkm/exec/internal/WorkletInvokeFunctor.h>
// Disable GCC warnings we check vtkmfor but Thrust does not.
#if defined(__GNUC__) && !defined(VTKM_CUDA)
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 6)
// Disable warnings we check vtkm for but Thrust does not.
#if defined(__GNUC__) || defined(____clang__)
#pragma GCC diagnostic push
#endif // gcc version >= 4.6
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 2)
#pragma GCC diagnostic ignored "-Wshadow"
#pragma GCC diagnostic ignored "-Wunused-parameter"
#endif // gcc version >= 4.2
#endif // gcc && !CUDA
#pragma GCC diagnostic ignored "-Wconversion"
#endif // gcc || clang
#include <thrust/advance.h>
#include <thrust/binary_search.h>
......@@ -52,11 +49,9 @@
#include <thrust/iterator/counting_iterator.h>
#if defined(__GNUC__) && !defined(VTKM_CUDA)
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 6)
#if defined(__GNUC__) || defined(____clang__)
#pragma GCC diagnostic pop
#endif // gcc version >= 4.6
#endif // gcc && !CUDA
#endif // gcc || clang
namespace vtkm {
namespace cont {
......@@ -76,28 +71,28 @@ void Schedule1DIndexKernel(FunctorType functor, vtkm::Id length)
template<class FunctorType>
__global__
void Schedule3DIndexKernel(FunctorType functor, vtkm::Id3 size)
void Schedule3DIndexKernel(FunctorType functor, dim3 size)
{
const vtkm::Id x = blockIdx.x*blockDim.x + threadIdx.x;
const vtkm::Id y = blockIdx.y*blockDim.y + threadIdx.y;
const vtkm::Id z = blockIdx.z*blockDim.z + threadIdx.z;
if (x >= size[0] || y >= size[1] || z >= size[2])
if (x >= size.x || y >= size.y || z >= size.z)
{
return;
}
//now convert back to flat memory
const int idx = x + size[0]*(y + size[1]*z);
const int idx = x + size.x*(y + size.y*z);
functor( idx );
}
inline
void compute_block_size(vtkm::Id3 rangeMax, dim3 blockSize3d, dim3& gridSize3d)
void compute_block_size(dim3 rangeMax, dim3 blockSize3d, dim3& gridSize3d)
{
gridSize3d.x = (rangeMax[0] % blockSize3d.x != 0) ? (rangeMax[0] / blockSize3d.x + 1) : (rangeMax[0] / blockSize3d.x);
gridSize3d.y = (rangeMax[1] % blockSize3d.y != 1) ? (rangeMax[1] / blockSize3d.y + 1) : (rangeMax[1] / blockSize3d.y);
gridSize3d.z = (rangeMax[2] % blockSize3d.z != 2) ? (rangeMax[2] / blockSize3d.z + 1) : (rangeMax[2] / blockSize3d.z);
gridSize3d.x = (rangeMax.x % blockSize3d.x != 0) ? (rangeMax.x / blockSize3d.x + 1) : (rangeMax.x / blockSize3d.x);
gridSize3d.y = (rangeMax.y % blockSize3d.y != 1) ? (rangeMax.y / blockSize3d.y + 1) : (rangeMax.y / blockSize3d.y);
gridSize3d.z = (rangeMax.z % blockSize3d.z != 2) ? (rangeMax.z / blockSize3d.z + 1) : (rangeMax.z / blockSize3d.z);
}
class PerfRecord
......@@ -121,14 +116,17 @@ public:
template<class Functor>
static void compare_3d_schedule_patterns(Functor functor, const vtkm::Id3& rangeMax)
{
const dim3 ranges(static_cast<vtkm::UInt32>(rangeMax[0]),
static_cast<vtkm::UInt32>(rangeMax[1]),
static_cast<vtkm::UInt32>(rangeMax[2]) );
std::vector< PerfRecord > results;
int indexTable[16] = {1, 2, 4, 8, 12, 16, 20, 24, 28, 30, 32, 64, 128, 256, 512, 1024};
vtkm::UInt32 indexTable[16] = {1, 2, 4, 8, 12, 16, 20, 24, 28, 30, 32, 64, 128, 256, 512, 1024};
for(int i=0; i < 16; i++)
for(vtkm::UInt32 i=0; i < 16; i++)
{
for(int j=0; j < 16; j++)
for(vtkm::UInt32 j=0; j < 16; j++)
{
for(int k=0; k < 16; k++)
for(vtkm::UInt32 k=0; k < 16; k++)
{
cudaEvent_t start, stop;
cudaEventCreate(&start);
......@@ -150,9 +148,9 @@ static void compare_3d_schedule_patterns(Functor functor, const vtkm::Id3& range
continue;
}
compute_block_size(rangeMax, blockSize3d, gridSize3d);
compute_block_size(ranges, blockSize3d, gridSize3d);
cudaEventRecord(start, 0);
Schedule3DIndexKernel<Functor> <<<gridSize3d, blockSize3d>>> (functor, rangeMax);
Schedule3DIndexKernel<Functor> <<<gridSize3d, blockSize3d>>> (functor, ranges);
cudaEventRecord(stop, 0);
cudaEventSynchronize(stop);
......@@ -169,12 +167,14 @@ static void compare_3d_schedule_patterns(Functor functor, const vtkm::Id3& range
}
std::sort(results.begin(), results.end());
for(int i=results.size()-1; i >= 0; --i)
const vtkm::Int64 size = static_cast<vtkm::Int64>(results.size());
for(vtkm::Int64 i=1; i <= size; i++)
{
int x = results[i].blockSize.x;
int y = results[i].blockSize.y;
int z = results[i].blockSize.z;
float t = results[i].elapsedTime;
vtkm::UInt64 index = static_cast<vtkm::UInt64>(size-i);
vtkm::UInt32 x = results[index].blockSize.x;
vtkm::UInt32 y = results[index].blockSize.y;
vtkm::UInt32 z = results[index].blockSize.z;
float t = results[index].elapsedTime;
std::cout << "BlockSize of: " << x << "," << y << "," << z << " required: " << t << std::endl;
}
......@@ -185,9 +185,9 @@ static void compare_3d_schedule_patterns(Functor functor, const vtkm::Id3& range
cudaEventCreate(&start);
cudaEventCreate(&stop);
const vtkm::Id numInstances = rangeMax[0] * rangeMax[1] * rangeMax[2];
const int blockSize = 128;
const int blocksPerGrid = (numInstances + blockSize - 1) / blockSize;
const vtkm::UInt32 numInstances = static_cast<vtkm::UInt32>(rangeMax[0] * rangeMax[1] * rangeMax[2]);
const vtkm::UInt32 blockSize = 128;
const vtkm::UInt32 blocksPerGrid = (numInstances + blockSize - 1) / blockSize;
cudaEventRecord(start, 0);
Schedule1DIndexKernel<Functor> <<<blocksPerGrid, blockSize>>> (functor, numInstances);
......@@ -212,9 +212,9 @@ static void compare_3d_schedule_patterns(Functor functor, const vtkm::Id3& range
dim3 blockSize3d(64,2,1);
dim3 gridSize3d;
compute_block_size(rangeMax, blockSize3d, gridSize3d);
compute_block_size(ranges, blockSize3d, gridSize3d);
cudaEventRecord(start, 0);
Schedule3DIndexKernel<Functor> <<<gridSize3d, blockSize3d>>> (functor, rangeMax);
Schedule3DIndexKernel<Functor> <<<gridSize3d, blockSize3d>>> (functor, ranges);
cudaEventRecord(stop, 0);
cudaEventSynchronize(stop);
......@@ -337,7 +337,8 @@ private:
binaryPredicate,
binaryOP);
return ::thrust::distance(keys_out_begin, result_iterators.first);
return static_cast<vtkm::Id>( ::thrust::distance(keys_out_begin,
result_iterators.first) );
}
template<class InputPortal, class OutputPortal>
......@@ -443,7 +444,7 @@ private:
outputBegin,
predicate);
return ::thrust::distance(outputBegin, newLast);
return static_cast<vtkm::Id>( ::thrust::distance(outputBegin, newLast) );
}
template<class ValuePortal,
......@@ -471,7 +472,7 @@ private:
IteratorType;
IteratorType begin = IteratorBegin(values);
IteratorType newLast = ::thrust::unique(begin, IteratorEnd(values));
return ::thrust::distance(begin, newLast);
return static_cast<vtkm::Id>( ::thrust::distance(begin, newLast) );
}
template<class ValuesPortal, class Compare>
......@@ -482,7 +483,7 @@ private:
IteratorType;
IteratorType begin = IteratorBegin(values);
IteratorType newLast = ::thrust::unique(begin, IteratorEnd(values), comp);
return ::thrust::distance(begin, newLast);
return static_cast<vtkm::Id>( ::thrust::distance(begin, newLast) );
}
template<class InputPortal, class ValuesPortal, class OutputPortal>
......@@ -728,8 +729,8 @@ public:
functor.SetErrorMessageBuffer(errorMessage);
const int blockSize = 128;
const int blocksPerGrid = (numInstances + blockSize - 1) / blockSize;
const vtkm::UInt32 blockSize = 128;
const vtkm::UInt32 blocksPerGrid = (static_cast<vtkm::UInt32>(numInstances) + blockSize - 1) / blockSize;
Schedule1DIndexKernel<Functor> <<<blocksPerGrid, blockSize>>> (functor, numInstances);
......@@ -768,6 +769,9 @@ public:
//requires the errormessage buffer be set
compare_3d_schedule_patterns(functor,rangeMax);
#endif
const dim3 ranges(static_cast<vtkm::UInt32>(rangeMax[0]),
static_cast<vtkm::UInt32>(rangeMax[1]),
static_cast<vtkm::UInt32>(rangeMax[2]) );
//currently we presume that 3d scheduling access patterns prefer accessing
//memory in the X direction. Also should be good for thin in the Z axis
......@@ -783,8 +787,8 @@ public:
}
dim3 gridSize3d;
compute_block_size(rangeMax, blockSize3d, gridSize3d);
Schedule3DIndexKernel<Functor> <<<gridSize3d, blockSize3d>>> (functor, rangeMax);
compute_block_size(ranges, blockSize3d, gridSize3d);
Schedule3DIndexKernel<Functor> <<<gridSize3d, blockSize3d>>> (functor, ranges);
//sync so that we can check the results of the call.
//In the future I want move this before the schedule call, and throwing
......
......@@ -25,28 +25,22 @@
#include <vtkm/exec/cuda/internal/ArrayPortalFromThrust.h>
// Disable GCC warnings we check vtkmfor but Thrust does not.
#if defined(__GNUC__) && !defined(VTKM_CUDA)
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 6)
// Disable warnings we check vtkm for but Thrust does not.
#if defined(__GNUC__) || defined(____clang__)
#pragma GCC diagnostic push
#endif // gcc version >= 4.6
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 2)
#pragma GCC diagnostic ignored "-Wshadow"
#pragma GCC diagnostic ignored "-Wunused-parameter"
#endif // gcc version >= 4.2
#endif // gcc && !CUDA
#pragma GCC diagnostic ignored "-Wconversion"
#endif // gcc || clang
#include <thrust/system/cuda/memory.h>
#include <thrust/functional.h>
#include <thrust/iterator/counting_iterator.h>
#include <thrust/iterator/transform_iterator.h>
#if defined(__GNUC__) && !defined(VTKM_CUDA)
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 6)
#if defined(__GNUC__) || defined(____clang__)
#pragma GCC diagnostic pop
#endif // gcc version >= 4.6
#endif // gcc && !CUDA
#endif // gcc || clang
namespace vtkm {
namespace cont {
......
......@@ -53,9 +53,9 @@ struct CountingTest
void operator()(const ValueType v) const
{
std::vector< ValueType > inputVector(ARRAY_SIZE);
for(int i=0; i < ARRAY_SIZE; ++i)
for(vtkm::Id i=0; i < ARRAY_SIZE; ++i)
{
inputVector[i] = v + i;
inputVector[static_cast<vtkm::UInt32>(i)] = v + i;
}
vtkm::cont::ArrayHandle< ValueType > input =
......@@ -65,7 +65,7 @@ struct CountingTest
dispatcher.Invoke(input, result);
//verify that the control portal works
for(int i=v; i < ARRAY_SIZE; ++i)
for(vtkm::Id i=static_cast<vtkm::Id>(v); i < ARRAY_SIZE; ++i)
{
const ValueType result_v = result.GetPortalConstControl().Get(i);
const ValueType correct_value = v + ValueType(i);
......
......@@ -97,10 +97,10 @@ public:
VTKM_CONT_EXPORT
detail::IteratorFromArrayPortalValue<ArrayPortalType>
operator[](vtkm::Id idx) const
operator[](std::size_t idx) const
{
return detail::IteratorFromArrayPortalValue<ArrayPortalType>(this->Portal,
idx);
static_cast<vtkm::Id>(idx) );
}
private:
......
......@@ -123,7 +123,7 @@ void CheckInPlaceResult(PortalType portal)
{
// This index was part of the permuted array; has a value changed
T expectedValue = TestValue(permutedIndex, T()) + T(1000);
T retrievedValue = portal.Get(permutedIndex);
T retrievedValue = portal.Get(permutedIndex );
VTKM_TEST_ASSERT(test_equal(expectedValue, retrievedValue),
"Permuted set unexpected value.");
}
......@@ -131,7 +131,7 @@ void CheckInPlaceResult(PortalType portal)
{
// This index was not part of the permuted array; has original value
T expectedValue = TestValue(permutedIndex, T());
T retrievedValue = portal.Get(permutedIndex);
T retrievedValue = portal.Get(permutedIndex );
VTKM_TEST_ASSERT(test_equal(expectedValue, retrievedValue),
"Permuted array modified value it should not have.");
}
......@@ -146,7 +146,7 @@ struct OutputPermutationFunctor : vtkm::exec::FunctorBase
VTKM_EXEC_EXPORT
void operator()(vtkm::Id index) const {
typedef typename PermutedPortalType::ValueType T;
this->PermutedPortal.Set(index, TestValue(index, T()));
this->PermutedPortal.Set(index, TestValue(static_cast<vtkm::Id>(index), T()));
}
};
......@@ -169,7 +169,7 @@ VTKM_CONT_EXPORT
void CheckOutputResult(PortalType portal)
{
typedef typename PortalType::ValueType T;
for (vtkm::Id permutedIndex = 0;
for (vtkm::IdComponent permutedIndex = 0;
permutedIndex < 2*ARRAY_SIZE;
permutedIndex++)
{
......@@ -178,7 +178,7 @@ void CheckOutputResult(PortalType portal)
// This index was part of the permuted array; has a value changed
vtkm::Id originalIndex = permutedIndex/2;
T expectedValue = TestValue(originalIndex, T());
T retrievedValue = portal.Get(permutedIndex);
T retrievedValue = portal.Get(permutedIndex );
VTKM_TEST_ASSERT(test_equal(expectedValue, retrievedValue),
"Permuted set unexpected value.");
}
......@@ -186,7 +186,7 @@ void CheckOutputResult(PortalType portal)
{
// This index was not part of the permuted array; has original value
T expectedValue = TestValue(permutedIndex, T());
T retrievedValue = portal.Get(permutedIndex);
T retrievedValue = portal.Get(permutedIndex );
VTKM_TEST_ASSERT(test_equal(expectedValue, retrievedValue),
"Permuted array modified value it should not have.");
}
......@@ -209,9 +209,10 @@ struct PermutationTests
ValueArrayType MakeValueArray() const {
// Allocate a buffer and set initial values
std::vector<ValueType> buffer(2*ARRAY_SIZE);
for (vtkm::Id index = 0; index < 2*ARRAY_SIZE; index++)
for (vtkm::IdComponent index = 0; index < 2*ARRAY_SIZE; index++)
{
buffer[index] = TestValue(index, ValueType());
vtkm::UInt32 i = static_cast<vtkm::UInt32>(index);
buffer[i] = TestValue(index, ValueType());
}
// Create an ArrayHandle from the buffer
......
......@@ -124,10 +124,12 @@ struct TryDefaultArray
{
std::cout << "Trying basic point coordinates array with a default storage."
<< std::endl;
std::vector<Vector3> buffer(ARRAY_SIZE);
std::vector<Vector3> buffer( static_cast<vtkm::UInt32>(ARRAY_SIZE) );
for (vtkm::Id index = 0; index < ARRAY_SIZE; index++)
{
buffer[index] = ExpectedCoordinates(index);
vtkm::UInt32 i = static_cast< vtkm::UInt32 >(index);
buffer[i] = ExpectedCoordinates(index);
}
vtkm::cont::DynamicPointCoordinates pointCoordinates =
......
......@@ -94,10 +94,12 @@ struct TestPointCoordinatesArray
std::cout << "Testing PointCoordinatesArray" << std::endl;
std::cout << " Creating buffer of data values" << std::endl;
std::vector<Vector3> buffer(ARRAY_SIZE);
std::vector<Vector3> buffer( static_cast<vtkm::UInt32>(ARRAY_SIZE) );
for (vtkm::Id index = 0; index < ARRAY_SIZE; index++)
{
buffer[index] = ExpectedCoordinates(index);
vtkm::UInt32 i = static_cast<vtkm::UInt32>(index);
buffer[i] = ExpectedCoordinates(index);
}
std::cout << " Creating and checking array handle" << std::endl;
......
......@@ -52,7 +52,8 @@ void CheckSame(const vtkm::Vec<TypeId,N> &expected,
for (vtkm::IdComponent index = 0; index < N; index++)
{
VTKM_TEST_ASSERT(expected[index] == found[index],
vtkm::UInt32 i = static_cast<vtkm::UInt32>(index);
VTKM_TEST_ASSERT(expected[index] == found[i],
"Got wrong type.");
}
}
......
......@@ -33,7 +33,7 @@ namespace {
struct TestExecutionObject : public vtkm::exec::ExecutionObjectBase
{
TestExecutionObject() : Number(0xDEADDEAD) { }
TestExecutionObject() : Number( static_cast<vtkm::Int32>(0xDEADDEAD) ) { }
TestExecutionObject(vtkm::Int32 number) : Number(number) { }
vtkm::Int32 Number;
};
......
......@@ -24,24 +24,18 @@
#include <iterator>
// Disable GCC warnings we check vtkmfor but Thrust does not.
#if defined(__GNUC__) && !defined(VTKM_CUDA)
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 6)
// Disable warnings we check vtkm for but Thrust does not.
#if defined(__GNUC__) || defined(____clang__)
#pragma GCC diagnostic push
#endif // gcc version >= 4.6
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 2)
#pragma GCC diagnostic ignored "-Wshadow"
#pragma GCC diagnostic ignored "-Wunused-parameter"
#endif // gcc version >= 4.2
#endif // gcc && !CUDA
#pragma GCC diagnostic ignored "-Wconversion"
#endif // gcc || clang
#include <thrust/system/cuda/memory.h>
#if defined(__GNUC__) && !defined(VTKM_CUDA)
#if (__GNUC__ >= 4) && (__GNUC_MINOR__ >= 6)
#if defined(__GNUC__) || defined(____clang__)
#pragma GCC diagnostic pop
#endif // gcc version >= 4.6
#endif // gcc && !CUDA
#endif // gcc || clang
#include <boost/type_traits/integral_constant.hpp>
#include <boost/utility/enable_if.hpp>
......@@ -135,7 +129,7 @@ public:
VTKM_EXEC_CONT_EXPORT
vtkm::Id GetNumberOfValues() const {
// Not using std::distance because on CUDA it cannot be used on a device.
return (this->EndIterator - this->BeginIterator);
return static_cast<vtkm::Id>( (this->EndIterator - this->BeginIterator) );
}
VTKM_EXEC_EXPORT
......@@ -196,7 +190,7 @@ public:
VTKM_EXEC_CONT_EXPORT
vtkm::Id GetNumberOfValues() const {
// Not using std::distance because on CUDA it cannot be used on a device.
return (this->EndIterator - this->BeginIterator);
return static_cast<vtkm::Id>( (this->EndIterator - this->BeginIterator) );
}
VTKM_EXEC_EXPORT
......
......@@ -90,7 +90,8 @@ void CheckSame(const vtkm::Vec<int,N> &expected,
for (vtkm::IdComponent index = 0; index < N; index++)