Fix several conversion warnings

These are bugs reported by floating point conversion flags.
acts-project · Jul 30, 2024 · 6d00cb3 · 6d00cb3
1 parent b647dee
commit 6d00cb3
Show file tree

Hide file tree

Showing 17 changed files with 149 additions and 69 deletions.
diff --git a/benchmarks/cuda/patterns/euler.hpp b/benchmarks/cuda/patterns/euler.hpp
@@ -92,8 +92,12 @@ struct EulerPattern : covfie::benchmark::AccessPattern<EulerPattern> {
         state.ResumeTiming();
 
         euler_kernel<<<
-            p.agents / p.block_size + (p.agents % p.block_size != 0 ? 1 : 0),
-            p.block_size>>>(device_agents, p.agents, p.steps, f);
+            static_cast<unsigned int>(
+                p.agents / p.block_size + (p.agents % p.block_size != 0 ? 1 : 0)
+            ),
+            static_cast<unsigned int>(p.block_size)>>>(
+            device_agents, p.agents, p.steps, f
+        );
 
         cudaErrorCheck(cudaGetLastError());
         cudaErrorCheck(cudaDeviceSynchronize());

diff --git a/benchmarks/cuda/patterns/lorentz.hpp b/benchmarks/cuda/patterns/lorentz.hpp
@@ -70,8 +70,8 @@ struct Lorentz : covfie::benchmark::AccessPattern<Lorentz<Propagator>> {
         std::vector<covfie::benchmark::lorentz_agent<3>> objs(p.particles);
 
         for (std::size_t i = 0; i < p.particles; ++i) {
-            float theta = std::acos(costheta_dist(e));
-            float phi = phi_dist(e);
+            float theta = std::acos(static_cast<float>(costheta_dist(e)));
+            float phi = static_cast<float>(phi_dist(e));
 
             objs[i].pos[0] = 0.f;
             objs[i].pos[1] = 0.f;
@@ -103,9 +103,13 @@ struct Lorentz : covfie::benchmark::AccessPattern<Lorentz<Propagator>> {
         state.ResumeTiming();
 
         lorentz_kernel<Propagator>
-            <<<p.particles / p.block_size +
-                   (p.particles % p.block_size != 0 ? 1 : 0),
-               p.block_size>>>(device_agents, p.particles, p.steps, f);
+            <<<static_cast<unsigned int>(
+                   p.particles / p.block_size +
+                   (p.particles % p.block_size != 0 ? 1 : 0)
+               ),
+               static_cast<unsigned int>(p.block_size)>>>(
+                device_agents, p.particles, p.steps, f
+            );
 
         cudaErrorCheck(cudaGetLastError());
         cudaErrorCheck(cudaDeviceSynchronize());

diff --git a/benchmarks/cuda/patterns/random.hpp b/benchmarks/cuda/patterns/random.hpp
@@ -92,8 +92,12 @@ struct Random : covfie::benchmark::AccessPattern<Random> {
         state.ResumeTiming();
 
         random_kernel<<<
-            p.agents / p.block_size + (p.agents % p.block_size != 0 ? 1 : 0),
-            p.block_size>>>(device_agents, device_out, p.agents, f);
+            static_cast<unsigned int>(
+                p.agents / p.block_size + (p.agents % p.block_size != 0 ? 1 : 0)
+            ),
+            static_cast<unsigned int>(p.block_size)>>>(
+            device_agents, device_out, p.agents, f
+        );
 
         cudaErrorCheck(cudaGetLastError());
         cudaErrorCheck(cudaDeviceSynchronize());

diff --git a/benchmarks/cuda/patterns/runge_kutta4.hpp b/benchmarks/cuda/patterns/runge_kutta4.hpp
@@ -93,8 +93,12 @@ struct RungeKutta4Pattern
         state.ResumeTiming();
 
         rk4_kernel<<<
-            p.agents / p.block_size + (p.agents % p.block_size != 0 ? 1 : 0),
-            p.block_size>>>(device_agents, p.agents, p.steps, f);
+            static_cast<unsigned int>(
+                p.agents / p.block_size + (p.agents % p.block_size != 0 ? 1 : 0)
+            ),
+            static_cast<unsigned int>(p.block_size)>>>(
+            device_agents, p.agents, p.steps, f
+        );
 
         cudaErrorCheck(cudaGetLastError());
         cudaErrorCheck(cudaDeviceSynchronize());

diff --git a/benchmarks/cuda/patterns/scan.hpp b/benchmarks/cuda/patterns/scan.hpp
@@ -70,14 +70,26 @@ struct Scan : covfie::benchmark::AccessPattern<Scan> {
 
         state.ResumeTiming();
 
-        dim3 block_size(8, 8, 8);
+        dim3 block_size(8u, 8u, 8u);
         dim3 grid_size(
-            p.x / block_size.x + (p.x % block_size.x != 0 ? 1 : 0),
-            p.y / block_size.y + (p.y % block_size.y != 0 ? 1 : 0),
-            p.z / block_size.z + (p.z % block_size.z != 0 ? 1 : 0)
+            static_cast<unsigned int>(
+                p.x / block_size.x + (p.x % block_size.x != 0 ? 1 : 0)
+            ),
+            static_cast<unsigned int>(
+                p.y / block_size.y + (p.y % block_size.y != 0 ? 1 : 0)
+            ),
+            static_cast<unsigned int>(
+                p.z / block_size.z + (p.z % block_size.z != 0 ? 1 : 0)
+            )
         );
 
-        scan_kernel<<<grid_size, block_size>>>(device_out, f, p.x, p.y, p.z);
+        scan_kernel<<<grid_size, block_size>>>(
+            device_out,
+            f,
+            static_cast<int>(p.x),
+            static_cast<int>(p.y),
+            static_cast<int>(p.z)
+        );
 
         cudaErrorCheck(cudaGetLastError());
         cudaErrorCheck(cudaDeviceSynchronize());

diff --git a/cmake/covfie-compiler-options-cpp.cmake b/cmake/covfie-compiler-options-cpp.cmake
@@ -19,6 +19,8 @@ if( ( "${CMAKE_CXX_COMPILER_ID}" MATCHES "GNU" ) OR
    covfie_add_flag( CMAKE_CXX_FLAGS "-Wshadow" )
    covfie_add_flag( CMAKE_CXX_FLAGS "-Wunused-local-typedefs" )
    covfie_add_flag( CMAKE_CXX_FLAGS "-pedantic" )
+   covfie_add_flag( CMAKE_CXX_FLAGS "-Wfloat-conversion" )
+   covfie_add_flag( CMAKE_CXX_FLAGS "-Wconversion" )
 
    # Fail on warnings, if asked for that behaviour.
    if( COVFIE_FAIL_ON_WARNINGS )

diff --git a/cmake/covfie-compiler-options-cuda.cmake b/cmake/covfie-compiler-options-cuda.cmake
@@ -33,6 +33,9 @@ if( "${CMAKE_CUDA_COMPILER_ID}" MATCHES "NVIDIA" )
    covfie_add_flag( CMAKE_CUDA_FLAGS "--expt-relaxed-constexpr" )
 endif()
 
+covfie_add_flag( CMAKE_CUDA_FLAGS "-Wfloat-conversion" )
+covfie_add_flag( CMAKE_CUDA_FLAGS "-Wconversion" )
+
 # Fail on warnings, if asked for that behaviour.
 if( COVFIE_FAIL_ON_WARNINGS )
    if( ( "${CUDAToolkit_VERSION}" VERSION_GREATER_EQUAL "10.2" ) AND

diff --git a/examples/common/bitmap/bitmap.cpp b/examples/common/bitmap/bitmap.cpp
@@ -161,7 +161,7 @@ void render_bitmap(
     for (unsigned int i = 0; i < 256; ++i) {
         float s = 0.7f;
         float v = 1.0f;
-        float hue = (i / 256.0f) * 240.0f;
+        float hue = (static_cast<float>(i) / 256.0f) * 240.0f;
         float c = v * s;
         float hp = hue / 60.0f;
 

diff --git a/examples/core/convert_bfield.cpp b/examples/core/convert_bfield.cpp
@@ -139,9 +139,12 @@ field_t read_bfield(const std::string & fn)
      * Now that we have the limits of our field, compute the size in each
      * dimension.
      */
-    std::size_t sx = std::lround((maxx - minx) / 100.0) + 1;
-    std::size_t sy = std::lround((maxy - miny) / 100.0) + 1;
-    std::size_t sz = std::lround((maxz - minz) / 100.0) + 1;
+    std::size_t sx =
+        static_cast<std::size_t>(std::lround((maxx - minx) / 100.0)) + 1;
+    std::size_t sy =
+        static_cast<std::size_t>(std::lround((maxy - miny) / 100.0)) + 1;
+    std::size_t sz =
+        static_cast<std::size_t>(std::lround((maxz - minz) / 100.0)) + 1;
 
     BOOST_LOG_TRIVIAL(info)
         << "Magnetic field size is " << sx << "x" << sy << "x" << sz;
@@ -151,9 +154,9 @@ field_t read_bfield(const std::string & fn)
     covfie::algebra::affine<3> translation =
         covfie::algebra::affine<3>::translation(-minx, -miny, -minz);
     covfie::algebra::affine<3> scaling = covfie::algebra::affine<3>::scaling(
-        (sx - 1) / (maxx - minx),
-        (sy - 1) / (maxy - miny),
-        (sz - 1) / (maxz - minz)
+        static_cast<float>(sx - 1) / (maxx - minx),
+        static_cast<float>(sy - 1) / (maxy - miny),
+        static_cast<float>(sz - 1) / (maxz - minz)
     );
 
     field_t field(covfie::make_parameter_pack(

diff --git a/examples/core/convert_bfield_csv.cpp b/examples/core/convert_bfield_csv.cpp
@@ -155,9 +155,12 @@ field_t read_bfield(const std::string & fn)
      * Now that we have the limits of our field, compute the size in each
      * dimension.
      */
-    std::size_t sx = std::lround((maxx - minx) / 100.0) + 1;
-    std::size_t sy = std::lround((maxy - miny) / 100.0) + 1;
-    std::size_t sz = std::lround((maxz - minz) / 100.0) + 1;
+    std::size_t sx =
+        static_cast<std::size_t>(std::lround((maxx - minx) / 100.0)) + 1;
+    std::size_t sy =
+        static_cast<std::size_t>(std::lround((maxy - miny) / 100.0)) + 1;
+    std::size_t sz =
+        static_cast<std::size_t>(std::lround((maxz - minz) / 100.0)) + 1;
 
     BOOST_LOG_TRIVIAL(info)
         << "Magnetic field size is " << sx << "x" << sy << "x" << sz;
@@ -167,9 +170,9 @@ field_t read_bfield(const std::string & fn)
     covfie::algebra::affine<3> translation =
         covfie::algebra::affine<3>::translation(-minx, -miny, -minz);
     covfie::algebra::affine<3> scaling = covfie::algebra::affine<3>::scaling(
-        (sx - 1) / (maxx - minx),
-        (sy - 1) / (maxy - miny),
-        (sz - 1) / (maxz - minz)
+        static_cast<float>(sx - 1) / (maxx - minx),
+        static_cast<float>(sy - 1) / (maxy - miny),
+        static_cast<float>(sz - 1) / (maxz - minz)
     );
 
     field_t field(covfie::make_parameter_pack(

diff --git a/examples/core/scaleup_bfield.cpp b/examples/core/scaleup_bfield.cpp
@@ -111,14 +111,14 @@ int main(int argc, char ** argv)
         for (std::size_t y = 0; y < 601; ++y) {
             for (std::size_t z = 0; z < 901; ++z) {
                 ov.at(
-                    -10000.f + x * 33.333333333f,
-                    -10000.f + y * 33.333333333f,
-                    -15000.f + z * 33.333333333f
+                    -10000.f + static_cast<float>(x) * 33.333333333f,
+                    -10000.f + static_cast<float>(y) * 33.333333333f,
+                    -15000.f + static_cast<float>(z) * 33.333333333f
                 ) =
                     iv.at(
-                        -10000.f + x * 33.333333333f,
-                        -10000.f + y * 33.333333333f,
-                        -15000.f + z * 33.333333333f
+                        -10000.f + static_cast<float>(x) * 33.333333333f,
+                        -10000.f + static_cast<float>(y) * 33.333333333f,
+                        -15000.f + static_cast<float>(z) * 33.333333333f
                     );
             }
         }

diff --git a/examples/cpu/render_slice.cpp b/examples/cpu/render_slice.cpp
@@ -128,8 +128,10 @@ int main(int argc, char ** argv)
 
     for (unsigned int x = 0; x < vm["width"].as<unsigned int>(); ++x) {
         for (unsigned int y = 0; y < vm["height"].as<unsigned int>(); ++y) {
-            float fx = x / static_cast<float>(vm["width"].as<unsigned int>());
-            float fy = y / static_cast<float>(vm["height"].as<unsigned int>());
+            float fx = static_cast<float>(x) /
+                       static_cast<float>(vm["width"].as<unsigned int>());
+            float fy = static_cast<float>(y) /
+                       static_cast<float>(vm["height"].as<unsigned int>());
 
             decltype(fv)::output_t p;
 

diff --git a/examples/cuda/render_slice.cu b/examples/cuda/render_slice.cu
@@ -90,8 +90,8 @@ __global__ void render(
     int y = blockDim.y * blockIdx.y + threadIdx.y;
 
     if (x < width && y < height) {
-        float fx = x / static_cast<float>(width);
-        float fy = y / static_cast<float>(height);
+        float fx = static_cast<float>(x) / static_cast<float>(width);
+        float fy = static_cast<float>(y) / static_cast<float>(height);
 
         typename field_t::output_t p =
             vf.at(fx * 20000.f - 10000.f, fy * 20000.f - 10000.f, z);

diff --git a/lib/core/covfie/core/backend/transformer/nearest_neighbour.hpp b/lib/core/covfie/core/backend/transformer/nearest_neighbour.hpp
@@ -140,7 +140,9 @@ struct nearest_neighbour {
 
             for (std::size_t i = 0; i < contravariant_output_t::dimensions; ++i)
             {
-                nc[i] = std::lrintf(c[i]);
+                nc[i] = static_cast<typename contravariant_output_t::scalar_t>(
+                    std::lrintf(c[i])
+                );
             }
 
             return m_backend.at(nc);

diff --git a/lib/core/covfie/core/backend/transformer/strided.hpp b/lib/core/covfie/core/backend/transformer/strided.hpp
@@ -61,7 +61,7 @@ struct strided {
                 std::accumulate(
                     std::begin(sizes),
                     std::end(sizes),
-                    1,
+                    1ul,
                     std::multiplies<std::size_t>()
                 )
             );
@@ -136,7 +136,7 @@ struct strided {
                   std::accumulate(
                       std::begin(m_sizes),
                       std::end(m_sizes),
-                      1,
+                      1ul,
                       std::multiplies<std::size_t>()
                   ),
                   make_strided_copy(o)

diff --git a/lib/cuda/covfie/cuda/backend/primitive/cuda_texture.hpp b/lib/cuda/covfie/cuda/backend/primitive/cuda_texture.hpp
@@ -108,13 +108,13 @@ struct cuda_texture {
                     typename T::parent_t::covariant_output_t::vector_t v =
                         no.at(ia);
 
-                    typename contravariant_input_t::scalar_t idx = 0;
+                    std::size_t idx = 0;
 
                     for (std::size_t k = contravariant_input_t::dimensions - 1;
                          k <= contravariant_input_t::dimensions;
                          --k)
                     {
-                        typename contravariant_input_t::scalar_t tmp = ia[k];
+                        std::size_t tmp = ia[k];
 
                         for (std::size_t l = k - 1; l < k; --l) {
                             tmp *= srcSize[l];
@@ -123,21 +123,26 @@ struct cuda_texture {
                         idx += tmp;
                     }
 
+                    std::size_t idx2 = static_cast<std::size_t>(idx);
+
+                    using stage_scalar_t =
+                        typename covariant_output_t::scalar_t;
+
                     if constexpr (covariant_output_t::dimensions == 1) {
-                        stage[idx] = v[0];
+                        stage[idx2] = static_cast<stage_scalar_t>(v[0]);
                     } else if constexpr (covariant_output_t::dimensions == 2) {
-                        stage[idx].x = v[0];
-                        stage[idx].y = v[1];
+                        stage[idx2].x = static_cast<stage_scalar_t>(v[0]);
+                        stage[idx2].y = static_cast<stage_scalar_t>(v[1]);
                     } else if constexpr (covariant_output_t::dimensions == 3) {
-                        stage[idx].x = v[0];
-                        stage[idx].y = v[1];
-                        stage[idx].z = v[2];
-                        stage[idx].w = 0.f;
+                        stage[idx2].x = static_cast<stage_scalar_t>(v[0]);
+                        stage[idx2].y = static_cast<stage_scalar_t>(v[1]);
+                        stage[idx2].z = static_cast<stage_scalar_t>(v[2]);
+                        stage[idx2].w = static_cast<stage_scalar_t>(0.f);
                     } else if constexpr (covariant_output_t::dimensions == 4) {
-                        stage[idx].x = v[0];
-                        stage[idx].y = v[1];
-                        stage[idx].z = v[2];
-                        stage[idx].w = v[3];
+                        stage[idx2].x = static_cast<stage_scalar_t>(v[0]);
+                        stage[idx2].y = static_cast<stage_scalar_t>(v[1]);
+                        stage[idx2].z = static_cast<stage_scalar_t>(v[2]);
+                        stage[idx2].w = static_cast<stage_scalar_t>(v[3]);
                     }
                 }),
                 std::tuple_cat(srcSize)