diff --git a/src/iluvatar/add/kernel.h b/src/iluvatar/add/kernel.h
new file mode 100644
index 0000000..551544f
--- /dev/null
+++ b/src/iluvatar/add/kernel.h
@@ -0,0 +1,41 @@
+#ifndef INFINI_OPS_ILUVATAR_ADD_KERNEL_H_
+#define INFINI_OPS_ILUVATAR_ADD_KERNEL_H_
+
+#include <utility>
+
+// clang-format off
+#include <cuda_runtime.h>
+// clang-format on
+
+#include "cuda/add/kernel.h"
+
+namespace infini::ops {
+
+namespace add {
+
+struct IluvatarBackend {
+  using stream_t = cudaStream_t;
+
+  static constexpr auto malloc = [](auto&&... args) {
+    return cudaMalloc(std::forward<decltype(args)>(args)...);
+  };
+
+  static constexpr auto memcpy = cudaMemcpy;
+
+  static constexpr auto free = cudaFree;
+
+  static constexpr auto memcpyH2D = cudaMemcpyHostToDevice;
+};
+
+}  // namespace add
+
+template <>
+class Operator<Add, Device::Type::kIluvatar>
+    : public CudaAdd<add::IluvatarBackend> {
+ public:
+  using CudaAdd<add::IluvatarBackend>::CudaAdd;
+};
+
+}  // namespace infini::ops
+
+#endif