Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

calling a __host__ function from a __host__ __device__ when using gather with zip_iterator for vector of vectors #783

Open
fkallen opened this issue Jul 13, 2021 · 0 comments
Assignees
Labels
thrust For all items related to Thrust.

Comments

@fkallen
Copy link

fkallen commented Jul 13, 2021

I would like to gather data from a struct of arrays, which contains some vector of vectors. thrust::zip_iterator is used to fuse the calls of individual arrays. This causes a series of warnings.

#include <thrust/iterator/zip_iterator.h>
#include <thrust/gather.h>
#include <thrust/host_vector.h>


int main(){
    thrust::host_vector<int> map(42);
    thrust::host_vector<thrust::host_vector<char>> vecvec(42); 
    thrust::host_vector<thrust::host_vector<char>> outvecvec(42); 

    auto in = thrust::make_zip_iterator(thrust::make_tuple(
        vecvec.begin()
    ));

    auto out = thrust::make_zip_iterator(thrust::make_tuple(
        outvecvec.begin()
    ));

    #if 1
   // warning calling a __host__ function from a __host__ __device__
    thrust::gather(
        map.begin(),
        map.end(),
        in,
        out
    );

    #else
    // no warning
    thrust::gather(
        map.begin(),
        map.end(),
        vecvec.begin(),
        outvecvec.begin()
    );

    #endif
}

https://cuda.godbolt.org/z/4Ebo6Yn57


/opt/compiler-explorer/libs/thrustcub/trunk/thrust/detail/tuple.inl(458): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during:
            instantiation of "thrust::detail::cons<HT, thrust::null_type>::cons(const thrust::detail::cons<HT2, thrust::null_type> &) [with HT=thrust::host_vector<char, std::allocator<char>>, HT2=thrust::host_vector<char, std::allocator<char>> &]" 
/opt/compiler-explorer/libs/thrustcub/trunk/thrust/tuple.h(346): here

/opt/compiler-explorer/libs/thrustcub/trunk/thrust/detail/function.h(118): warning: calling a __host__ function("thrust::host_vector<char, ::std::allocator<char> > ::~host_vector()") from a __host__ __device__ function("thrust::detail::cons< ::thrust::host_vector<char, ::std::allocator<char> > ,  ::thrust::null_type> ::~cons") is not allowed

/opt/compiler-explorer/libs/thrustcub/trunk/thrust/detail/tuple.inl(458): warning: calling a __host__ function("thrust::host_vector<char, ::std::allocator<char> > ::host_vector(const thrust::host_vector<char, ::std::allocator<char> > &)") from a __host__ __device__ function("thrust::detail::cons< ::thrust::host_vector<char, ::std::allocator<char> > ,  ::thrust::null_type> ::cons< ::thrust::host_vector<char, ::std::allocator<char> >  &> ") is not allowed

@jrhemstad jrhemstad added the thrust For all items related to Thrust. label Feb 22, 2023
@jarmak-nv jarmak-nv transferred this issue from NVIDIA/thrust Nov 8, 2023
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
thrust For all items related to Thrust.
Projects
Status: No status
Development

No branches or pull requests

3 participants