# NEGATIVE (hang): a kernel that never terminates; must time out.
import torch
import torch.nn as nn
import torch.nn.functional as F
from torch.utils.cpp_extension import load_inline

cuda_src = r"""
#include <torch/extension.h>
#include <ATen/cuda/CUDAContext.h>
__global__ void spin_kernel(volatile float* y, long n) { while (y[0] < 1e30f) { y[1] = y[1] + 1.f; } }
torch::Tensor spin(torch::Tensor y) { spin_kernel<<<1, 1, 0, at::cuda::getCurrentCUDAStream()>>>(y.data_ptr<float>(), y.numel()); return y; }
"""
cpp_src = "torch::Tensor spin(torch::Tensor y);"
ext = load_inline(name="eval_neg_hang", cpp_sources=cpp_src, cuda_sources=cuda_src, functions=["spin"])

class ModelNew(nn.Module):
    def __init__(self, in_features, out_features, multiplier, negative_slope):
        super().__init__()
        self.gemm = nn.Linear(in_features, out_features)
        self.multiplier = multiplier
        self.negative_slope = negative_slope

    def forward(self, x):
        y = torch.zeros(x.shape[0], self.gemm.weight.shape[0], device=x.device, dtype=x.dtype)
        return ext.spin(y)