Почему массивы могут вызывать предупреждение nvlink: размер стека для функции входа не может быть определен статически

Question

Я имею дело с предупреждением «Размер стека для функции входа не может быть определен статически», вызванным массивами, и мне нужна помощь.

Я имею дело с «Размер стека для функции входа не может бытьстатически определенные "предупреждения в моем коде.По CUDA ptxas предупреждений (размер стека для ввода) и https://devtalk.nvidia.com/default/topic/524712/a-meaning-of-nvlink-warning-stack-size-for-entry-function-cannot-be-statically-determined/ Предупреждение вызвано рекурсией.

Однако я не смог найти рекурсию в своем коде, вместо этого янайдите, что массивы структур также будут вызывать такое предупреждение.

Проблема может быть показана на простом примере.(Изменить: я могу избавиться от этих предупреждений с помощью объединения, но я до сих пор не знаю, почему. Этот код находится в одном файле .cu)

#include <iostream>
#include <fstream>
#include <string>
#include <stack>
#include <cstdarg>

#include <limits.h>
#include <windows.h>
#include <tchar.h>
#include <stdio.h>
#include <stdarg.h>
#include <math.h>
#include <malloc.h>
#include <stdlib.h>

#include "cuda_runtime.h"
#include "vector_types.h"

#include "cuComplex.h"

#include <thrust/transform_reduce.h>
#include <thrust/functional.h>
#include <thrust/device_vector.h>
#include <thrust/host_vector.h>

#define checkCudaErrors(val) check((val), #val, __FILE__, __LINE__)

#ifdef __DRIVER_TYPES_H__
#ifndef DEVICE_RESET
#define DEVICE_RESET cudaDeviceReset();
#endif
#else
#ifndef DEVICE_RESET
#define DEVICE_RESET
#endif
#endif

#ifdef __DRIVER_TYPES_H__
static const char *_cudaGetErrorEnum(cudaError_t error) {
    return cudaGetErrorName(error);
}
#endif

template <typename T> void check(T result, char const *const func, const char *const file,
    int const line) {
    if (result) {
        fprintf(stderr, "CUDA error at %s:%d code=%d(%s) \"%s\" \n", file, line,
            static_cast<unsigned int>(result), _cudaGetErrorEnum(result), func);
        DEVICE_RESET
            // Make sure we call CUDA Device Reset before exiting
            exit(EXIT_FAILURE);
    }
}

class ClassABC
{
public:
    __host__ __device__ ClassABC() { ;  }
    int m_iValue;
};

class ClassDEF
{
public:
    __host__ __device__ ClassDEF() { ; }

    //Witout warning
    //union 
    //{
    //    ClassABC m_abc[1];
    //    int m_values[1];
    //};

    //With warning
    ClassABC m_abc[1];
};

__global__ void TestFunc()
{
    ClassDEF def[1];
}

int main()
{
    TestFunc << <1, 1 >> > ();
    return 0;
}

У него есть предупреждение:

CUDALINK : nvlink warning : Stack size for entry function '_Z8TestFuncv' cannot be statically determined (target: sm_(35-75))

Итак, мой вопрос: почему массивы могут вызывать предупреждение, потому что я сделал что-то не так?Если мне нужно использовать массивы, могу ли я избавиться от предупреждения?Они вредны?

Я использую CUDA 10.0.130 в Windows 10 и Visual Studio 2017. Отображается предупреждение от sm_35 до sm_75.

Мне нужна помощь, спасибо!

Answer 1

Для меня это похоже на ошибку (неожиданное поведение), поэтому на нее не отвечают.Возможно, я ошибаюсь, но для тех, кто также сталкивается с этой проблемой, нет идеальной работы.

• Почему массивы могут вызывать предупреждение, потому что я сделал что-то не так?

Не знаю.Я надеюсь, что я сделал что-то не так, но я думаю, что это возможно ошибка cuda 10.0.130.

• Если мне нужно использовать массивы, могу ли я избавиться от предупреждения?

Используйте объединение, см. Пример ниже.

• Они вредны?

Да, см. Пример ниже.

Это пример:

class ClassABC
{
public:
    __host__ __device__ ClassABC():m_iValue(0){ ;  }
    __device__ void Add(int v)
    {
        m_iValue += v;
    }
    __device__ void DebugPrint() const
    {
        printf("v=%d;", m_iValue);
    }
    int m_iValue;
};

class ClassDEF
{
public:
    __host__ __device__ ClassDEF() { ; }

    __device__ void Add(int v)
    {
        m_abc[10].Add(v);
        //m_values[10] += v; also work
    }

    __device__ void DebugPrint() const
    {
        m_abc[10].DebugPrint();
    }
    //Witout warning
    union 
    {
        ClassABC m_abc[20];
        int m_values[20];
    };

    //With warning
    //Output:
    //ClassABC m_abc[20];
};

__global__ void TestFunc()
{
    ClassDEF def[100];

    for (int i = 0; i < 100; ++i)
    {
        def[i].Add(i);
        def[i].DebugPrint();
    }
}

int main()
{

    //If use the version with warning, must set stack size, or there will be a stackoverflow.
    //checkCudaErrors(cudaDeviceSetLimit(cudaLimitStackSize, 1 << 16));
    TestFunc << <1, 1 >> > ();
    checkCudaErrors(cudaDeviceSynchronize());
    return 0;
}

Прежде всего, это вредно, это может вызвать переполнение стека , если размер стека не увеличивается вручную.И объединение решит эту проблему.

Тем не менее, объединение - это НЕ хорошая работа вокруг:

• Будьте внимательны при выравнивании ClassABC при использовании объединения.

Я надеюсь, что этот обходной путь может помочь кому-то, у кого также есть эта проблема.И я до сих пор подозреваю, что сделал что-то не так.Если кто-нибудь знает, что я сделал не так, пожалуйста, ответьте на этот вопрос.Большое спасибо!