Cuda Си, умножение матрицы на вектор AxV. Неправильно считает результирующий вектор

не понимаю и никогда не пойму, но сделать надо. При умножении матрицы А на вектор V на ядре считается одна строка матрицы, остальные нули (хотя при указании разных показателей размера блоков правильное значение повторяется в результате через N/2). Как я понял мы должны отдать ядру в kernel<<<кол-во строк N, и одну нить 1>>>, т.е. каждая строка матрицы это блок и в нем одна нить циклом высчитывает одну ячейку результирующего вектора (попарно умножаем и складываем). Но считается только один раз, не пойму почему, какие размеры блока нужны или кол-ва нитей, или проблема вообще не в ядре. Подскажите пожалуйста.

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>
#include <Windows.h>

__global__ void addKernel(int* dev_C, int* dev_A, int* dev_V, int M)
{
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    for (int k = 0; k < M; k++)
    {
        dev_C[i] += dev_A[i * M + k] * dev_V[k];
    }
}

int main()
{
    int N = 0;
    int M = 0;

    printf("Enter please N (count row matrix): ");
    scanf("%d", &N);
    printf("Enter please M (count col matrix): ");
    scanf("%d", &M);

    int* host_A;
    int* host_V;
    int* host_C;

    host_A = new int[N * M];
    host_V = new int[M];
    host_C = new int[N];

    for (int i = 0; i < N; i++)  // Инициализация массивов A, V, C
    {
        for (int j = 0; j < M; j++)
        {
            host_A[i * M + j] = 1;
            host_V[j] = 1;
        }
        host_C[i] = 0;
    }

    printf("Vector V: { ");
    for (int i = 0; i < M; i++) {
        printf("%d ", host_V[i]);
    }
    printf("}\n");

    printf("Vector C: { ");
    for (int i = 0; i < N; i++) {
        printf("%d ", host_C[i]);
    }
    printf("}\n");

    printf("Matrix A:\n");

    for (int i = 0; i < N; i++) {
        for (int j = 0; j < M; j++) {
            printf("%3d", host_A[i * M + j]);
        }
        printf("\n");
    }
    printf("\n");

    int *dev_A = 0;
    int *dev_V = 0;
    int *dev_C = 0;
    size_t pitch;
    cudaError_t cudaStatus;

    cudaMalloc((void**)&dev_C, N * sizeof(int));
    cudaMallocPitch((void**)&dev_A, &pitch, N * sizeof(int), M);
    cudaMalloc((void**)&dev_V, M * sizeof(int));
    
    cudaEvent_t start, stop;
    cudaGetErrorString(cudaEventCreate(&start));
    cudaEventCreate(&stop);

    cudaGetErrorString(cudaEventRecord(start, 0));

    cudaMemcpy2D(dev_A, pitch, host_A, N * sizeof(int), N * sizeof(int), M, cudaMemcpyHostToDevice);

    cudaMemcpy(dev_V, host_V, M * sizeof(int), cudaMemcpyHostToDevice);

    //dim3 grid(N / (256 + 1));

    addKernel << <N, 1 >> > (dev_C, dev_A, dev_V, M);

    cudaDeviceSynchronize();

    cudaMemcpy(host_C, dev_C, N * sizeof(int), cudaMemcpyDeviceToHost);

    for (int i = 0; i < N; i++) {
        printf("%d ", host_C[i]);
    }
    printf("\n");

    cudaGetErrorString(cudaEventRecord(stop, 0));
    cudaGetErrorString(cudaEventSynchronize(stop));

    float elapsedTime;
    cudaGetErrorString(cudaEventElapsedTime(&elapsedTime, start, stop));
 
    cudaGetErrorString(cudaEventDestroy(start));
    cudaGetErrorString(cudaEventDestroy(stop));

    printf("\nElapsed time = %f\n", elapsedTime);

    cudaFree(dev_C);
    cudaFree(dev_A);
    cudaFree(dev_V);

    system("pause");

    return 0;
}