search

ingowald / cudaKDTree

188 stars 12 forks source link

Is it possible to perform FCP on host? #22

Open ahadji05 opened 3 months ago

ahadji05 commented 3 months ago

Hi,

I have seen the example for building the kd-tree on host using cukd::buildTree_host.

Is it also possible to perform scan for the nearest neighbor on host, via cukd::stackFree::fcp or cukd::stackBased::fcp ?

ingowald commented 3 months ago

have never tried, i think, but see no reason why not. let me know if you run into any issues

ahadji05 commented 3 months ago

I tried the following:


template<typename T, class Mem>
void kdtreeSearch<T,Mem>::find_closet_host( int numQueryData, int *traceIndex, T const* x, T const* y, T const* z, T const* w ){

    for( size_t queryIndex = 0; queryIndex < numQueryData; ++queryIndex ){
        PointPlusIndex queryPoint;
        queryPoint.pos.x = (float) x[queryIndex];
        queryPoint.pos.y = (float) y[queryIndex];
        queryPoint.pos.z = (float) z[queryIndex];
        queryPoint.pos.w = (float) w[queryIndex];

        int closestID = cukd::stackFree::fcp<PointPlusIndex,PointPlusIndex_traits>(queryPoint.pos, m_data, m_numSearchData);

        traceIndex[queryIndex] = m_data[closestID].index;
    }
}

but it looks like cukd::stackFree::fcp cannot compile inside a host function!


error: calling a __device__ function("int cukd::stackFree::fcp<    ::PointPlusIndex,     ::PointPlusIndex_traits> ( ::T2::point_t, const T1 *, int,  ::cukd::FcpSearchParams)") from a __host__ function("find_closet_host") is not allowed
          int closestID = cukd::stackFree::fcp<PointPlusIndex,PointPlusIndex_traits>(queryPoint.pos, m_data, m_numSearchData);