Build for Windows and VS 2022 does not compile CUDA sources
Author: JRGit4UE
Pull request: no
Labels: name: bug

### System Info

I am trying to build the **transformers** package  **v4.48.1** from source  
for platform *Windows 10*  
with python *v3.12*  
*pytorch 2.5.1+cu124*  
and *CUDA 12.4*  
along with *Visual Studio 2022*. 

I have executed  `vcvarsall.bat x64` , furthermore
all known environment variables like
* `set CUDA_HOME=...`
* `set DISTUTILS_USE_SDK=...`
* `set CXX=...`
* `set TORCH_DONT_CHECK_COMPILER_ABI=...`
* `set TORCH_CUDA_ARCH_LIST=...`
are in place.

`python setup.py build` just copies the CUDA files to the build directory like

``` bash
copying src\transformers\kernels\deformable_detr\cuda\ms_deform_attn_cuda.cu -> build\lib\transformers\kernels\deformable_detr\cuda
```
but the *.cu* files are never handled by *nvcc*, as the compiler is never called.

How can I compile the transformers package sources in Windows for CUDA 12.4?

### Who can help?

_No response_

### Information

- [ ] The official example scripts
- [ ] My own modified scripts

### Tasks

- [ ] An officially supported task in the `examples` folder (such as GLUE/SQuAD, ...)
- [ ] My own task or dataset (give details below)

### Reproduction

1) set CUDA_HOME=\<path to cuda 12.4 toolkit\>  
2) set CXX=\<path to vs2022 cl.exe\>
3) make sure pip version  is *v25.0*
4) make sure setuptools version is *v70.0.0*
5) execute `python setup.py build`
6) grep build output for *cuda* or *nvcc*

### Expected behavior

The .cu files  should be compiled for the architectures in the given arch list.

Comments:
Hi @JRGit4UE, Transformers generally does not compile CUDA kernels on installation. Instead, they are [lazily compiled at runtime](https://github.com/huggingface/transformers/blob/cf8091c017533c03be73b84ab535ae9c80924796/src/transformers/models/mra/modeling_mra.py#L59) (thank you @IlyasMoutawwakil for pointing this out). As a result, you should not expect to see any CUDA compilation while installing the package on any platform.

@Rocketknight1 : Thank you for clarifying.   
Does this mean, that every package user must have the proper cuda toolkit installed at runtime, along with the corresponding compiler - cuda 12.4 toolkit and vs 2022 plus env vars in my case?   
And if lazy compilation takes place, are all possible gpu architectures built (i.e. cuda 12.4 supports 50 60 61 70 75 80 86 89 90) which I have properly set in TORCH_CUDA_ARCH_LIST?  
The problem for me is that the code runs without any issue on my workstation with RTX 3090 (=80+ architecture) but fails on my other machine with P4000 (61 arch) with error:
`ms_deformable_im2col_cuda: no kernel image is available for execution on the device`.  



> And if lazy compilation takes place, are all possible gpu architectures built

Normally yes.
Can you please try running something like this example based on [torch testing suite](https://github.com/pytorch/pytorch/blob/main/test/test_cpp_extensions_jit.py#L376)

```python
cuda_source = """
__global__ void cos_add_kernel(
    const float* __restrict__ x,
    const float* __restrict__ y,
    float* __restrict__ output,
    const int size) {
  const auto index = blockIdx.x * blockDim.x + threadIdx.x;
  if (index < size) {
    output[index] = __cosf(x[index]) + __cosf(y[index]);
  }
}

torch::Tensor cos_add(torch::Tensor x, torch::Tensor y) {
  auto output = torch::zeros_like(x);
  const int threads = 1024;
  const int blocks = (output.numel() + threads - 1) / threads;
  cos_add_kernel<<<blocks, threads>>>(x.data<float>(), y.data<float>(), output.data<float>(), output.numel());
  return output;
}
"""

# Here, the C++ source need only declare the function signature.
cpp_source = "torch::Tensor cos_add(torch::Tensor x, torch::Tensor y);"

module = torch.utils.cpp_extension.load_inline(
    name="inline_jit_extension_cuda",
    cpp_sources=cpp_source,
    cuda_sources=cuda_source,
    functions=["cos_add"],
    verbose=True,
)

x = torch.randn(4, 4, device="cuda", dtype=torch.float32)
y = torch.randn(4, 4, device="cuda", dtype=torch.float32)
z = module.cos_add(x, y)
```
and see if it's compiled and works correctly. If not then the issue might be in CUDA 12 compatibility with P4000 (I read somewhere that it dropped support for Pascal but not sure) or in pytorch.
You might also try something like 6.0+PTX in your targeted architectures.

Also see https://github.com/fundamentalvision/Deformable-DETR/issues/88 which seems to be a common issue (this is the repo these kernels were ported from).

@IlyasMoutawwakil  thank you for the testcode.   
I have built it and it turned out that my syntax for defining preferred CUDA architectures was wrong.  
Instead  of 
`set TORCH_CUDA_ARCH_LIST=5.0;6.0;6.1;7.0;7.5;8.0;8.6;8.9;9.0`  
which seemed to work properly, the correct env setting in *Windows* for CUDA 12.4 must be 
`set TORCH_CUDA_ARCH_LIST=Maxwell Pascal Volta Turing Ampere Ada Hopper`
WITHOUT double quotes!


For completeness, here is the runnable testcode.  
Thank you for your help : )

``` python
#(.venv) d:\project\torch-cuda-test>python
#>>> import torch
#>>> torch.__version__
#'2.5.1+cu124'
#>>>

# set TORCH_CUDA_ARCH_LIST=Maxwell Pascal Volta Turing Ampere Ada Hopper
# WORKS!!!


import torch
import torch.utils.cpp_extension as cpp_extension
from typing import Optional
import collections
import os

cuda_source = """
__global__ void cos_add_kernel(
    const float* __restrict__ x,
    const float* __restrict__ y,
    float* __restrict__ output,
    const int size) {
  const auto index = blockIdx.x * blockDim.x + threadIdx.x;
  if (index < size) {
    output[index] = __cosf(x[index]) + __cosf(y[index]);
  }
}

torch::Tensor cos_add(torch::Tensor x, torch::Tensor y) {
  auto output = torch::zeros_like(x);
  const int threads = 1024;
  const int blocks = (output.numel() + threads - 1) / threads;
  cos_add_kernel<<<blocks, threads>>>(x.data<float>(), y.data<float>(), output.data<float>(), output.numel());
  return output;
}
"""

# Here, the C++ source need only declare the function signature.
cpp_source = "torch::Tensor cos_add(torch::Tensor x, torch::Tensor y);"

module = torch.utils.cpp_extension.load_inline(
    name="inline_jit_extension_cuda",
    cpp_sources=cpp_source,
    cuda_sources=cuda_source,
    functions=["cos_add"],
    verbose=True,
)

x = torch.randn(4, 4, device="cuda", dtype=torch.float32)
y = torch.randn(4, 4, device="cuda", dtype=torch.float32)
z = module.cos_add(x, y)
```


This issue has been automatically marked as stale because it has not had recent activity. If you think this still needs to be addressed please comment on this thread.

Please note that issues that do not follow the [contributing guidelines](https://github.com/huggingface/transformers/blob/main/CONTRIBUTING.md) are likely to be ignored.